Commit 7f795e96 authored by Roman Donchenko's avatar Roman Donchenko

Merge commit 'ee36717d' into merge-2.4

parents 9dfb9638 ee36717d
......@@ -213,6 +213,7 @@ OCV_OPTION(ENABLE_VFPV3 "Enable VFPv3-D32 instructions"
OCV_OPTION(ENABLE_NOISY_WARNINGS "Show all warnings even if they are too noisy" OFF )
OCV_OPTION(OPENCV_WARNINGS_ARE_ERRORS "Treat warnings as errors" OFF )
OCV_OPTION(ENABLE_WINRT_MODE "Build with Windows Runtime support" OFF IF WIN32 )
OCV_OPTION(ENABLE_WINRT_MODE_NATIVE "Build with Windows Runtime native C++ support" OFF IF WIN32 )
# ----------------------------------------------------------------------------
......@@ -641,7 +642,7 @@ endif()
if(WIN32)
status("")
status(" Windows RT support:" HAVE_WINRT THEN YES ELSE NO)
if (ENABLE_WINRT_MODE)
if (ENABLE_WINRT_MODE OR ENABLE_WINRT_MODE_NATIVE)
status(" Windows SDK v8.0:" ${WINDOWS_SDK_PATH})
status(" Visual Studio 2012:" ${VISUAL_STUDIO_PATH})
endif()
......
......@@ -9,7 +9,7 @@ set(HAVE_WINRT FALSE)
# search Windows Platform SDK
message(STATUS "Checking for Windows Platform SDK")
GET_FILENAME_COMPONENT(WINDOWS_SDK_PATH "[HKEY_LOCAL_MACHINE\\SOFTWARE\\Microsoft\\Microsoft SDKs\\Windows\\v8.0;InstallationFolder]" ABSOLUTE CACHE)
if (WINDOWS_SDK_PATH STREQUAL "")
if(WINDOWS_SDK_PATH STREQUAL "")
set(HAVE_MSPDK FALSE)
message(STATUS "Windows Platform SDK 8.0 was not found")
else()
......@@ -19,7 +19,7 @@ endif()
#search for Visual Studio 11.0 install directory
message(STATUS "Checking for Visual Studio 2012")
GET_FILENAME_COMPONENT(VISUAL_STUDIO_PATH [HKEY_LOCAL_MACHINE\\SOFTWARE\\Microsoft\\VisualStudio\\11.0\\Setup\\VS;ProductDir] REALPATH CACHE)
if (VISUAL_STUDIO_PATH STREQUAL "")
if(VISUAL_STUDIO_PATH STREQUAL "")
set(HAVE_MSVC2012 FALSE)
message(STATUS "Visual Studio 2012 was not found")
else()
......@@ -30,11 +30,15 @@ try_compile(HAVE_WINRT_SDK
"${OpenCV_BINARY_DIR}"
"${OpenCV_SOURCE_DIR}/cmake/checks/winrttest.cpp")
if (ENABLE_WINRT_MODE AND HAVE_WINRT_SDK AND HAVE_MSVC2012 AND HAVE_MSPDK)
if(ENABLE_WINRT_MODE AND HAVE_WINRT_SDK AND HAVE_MSVC2012 AND HAVE_MSPDK)
set(HAVE_WINRT TRUE)
set(HAVE_WINRT_CX TRUE)
elseif(ENABLE_WINRT_MODE_NATIVE AND HAVE_WINRT_SDK AND HAVE_MSVC2012 AND HAVE_MSPDK)
set(HAVE_WINRT TRUE)
set(HAVE_WINRT_CX FALSE)
endif()
if (HAVE_WINRT)
if(HAVE_WINRT)
add_definitions(/DWINVER=0x0602 /DNTDDI_VERSION=NTDDI_WIN8 /D_WIN32_WINNT=0x0602)
set(CMAKE_SHARED_LINKER_FLAGS "${CMAKE_SHARED_LINKER_FLAGS} /appcontainer")
set(CMAKE_MODULE_LINKER_FLAGS "${CMAKE_MODULE_LINKER_FLAGS} /appcontainer")
......
......@@ -2,8 +2,11 @@ set(the_description "The Core Functionality")
ocv_add_module(core PRIVATE_REQUIRED ${ZLIB_LIBRARIES} "${OPENCL_LIBRARIES}" OPTIONAL opencv_cudev)
ocv_module_include_directories(${ZLIB_INCLUDE_DIRS})
if(HAVE_WINRT_CX)
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} /ZW")
endif()
if(HAVE_WINRT)
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} /ZW /GS /Gm- /AI\"${WINDOWS_SDK_PATH}/References/CommonConfiguration/Neutral\" /AI\"${VISUAL_STUDIO_PATH}/vcpackages\"")
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} /GS /Gm- /AI\"${WINDOWS_SDK_PATH}/References/CommonConfiguration/Neutral\" /AI\"${VISUAL_STUDIO_PATH}/vcpackages\"")
endif()
if(HAVE_CUDA)
......
......@@ -63,7 +63,7 @@ inline float sum(float val)
return val;
}
static float clamp1(float var, float learningRate, float diff, float minVar)
inline float clamp1(float var, float learningRate, float diff, float minVar)
{
return fmax(var + learningRate * (diff * diff - var), minVar);
}
......@@ -96,7 +96,7 @@ inline float sum(const float4 val)
return (val.x + val.y + val.z);
}
static void swap4(__global float4* ptr, int x, int y, int k, int rows, int ptr_step)
inline void swap4(__global float4* ptr, int x, int y, int k, int rows, int ptr_step)
{
float4 val = ptr[(k * rows + y) * ptr_step + x];
ptr[(k * rows + y) * ptr_step + x] = ptr[((k + 1) * rows + y) * ptr_step + x];
......@@ -104,7 +104,7 @@ static void swap4(__global float4* ptr, int x, int y, int k, int rows, int ptr_s
}
static float4 clamp1(const float4 var, float learningRate, const float4 diff, float minVar)
inline float4 clamp1(const float4 var, float learningRate, const float4 diff, float minVar)
{
float4 result;
result.x = fmax(var.x + learningRate * (diff.x * diff.x - var.x), minVar);
......@@ -128,7 +128,7 @@ typedef struct
uchar c_shadowVal;
} con_srtuct_t;
static void swap(__global float* ptr, int x, int y, int k, int rows, int ptr_step)
inline void swap(__global float* ptr, int x, int y, int k, int rows, int ptr_step)
{
float val = ptr[(k * rows + y) * ptr_step + x];
ptr[(k * rows + y) * ptr_step + x] = ptr[((k + 1) * rows + y) * ptr_step + x];
......
......@@ -44,7 +44,7 @@
//
//M*/
static float distance_(__global const float * center, __global const float * src, int feature_length)
inline float distance_(__global const float * center, __global const float * src, int feature_length)
{
float res = 0;
float4 v0, v1, v2;
......
......@@ -46,7 +46,7 @@
//
//M*/
static short2 do_mean_shift(int x0, int y0, __global uchar4* out,int out_step,
inline short2 do_mean_shift(int x0, int y0, __global uchar4* out,int out_step,
__global uchar4* in, int in_step, int dst_off, int src_off,
int cols, int rows, int sp, int sr, int maxIter, float eps)
{
......
......@@ -208,7 +208,7 @@ __kernel void normalize_hists_36_kernel(__global float* block_hists,
//-------------------------------------------------------------
// Normalization of histograms via L2Hys_norm
//
static float reduce_smem(volatile __local float* smem, int size)
inline float reduce_smem(volatile __local float* smem, int size)
{
unsigned int tid = get_local_id(0);
float sum = smem[tid];
......
......@@ -52,7 +52,7 @@
#endif
#ifdef CPU
static void reduce3(float val1, float val2, float val3, __local float* smem1, __local float* smem2, __local float* smem3, int tid)
inline void reduce3(float val1, float val2, float val3, __local float* smem1, __local float* smem2, __local float* smem3, int tid)
{
smem1[tid] = val1;
smem2[tid] = val2;
......@@ -71,7 +71,7 @@ static void reduce3(float val1, float val2, float val3, __local float* smem1,
}
}
static void reduce2(float val1, float val2, volatile __local float* smem1, volatile __local float* smem2, int tid)
inline void reduce2(float val1, float val2, volatile __local float* smem1, volatile __local float* smem2, int tid)
{
smem1[tid] = val1;
smem2[tid] = val2;
......@@ -88,7 +88,7 @@ static void reduce2(float val1, float val2, volatile __local float* smem1, volat
}
}
static void reduce1(float val1, volatile __local float* smem1, int tid)
inline void reduce1(float val1, volatile __local float* smem1, int tid)
{
smem1[tid] = val1;
barrier(CLK_LOCAL_MEM_FENCE);
......@@ -103,7 +103,7 @@ static void reduce1(float val1, volatile __local float* smem1, int tid)
}
}
#else
static void reduce3(float val1, float val2, float val3,
inline void reduce3(float val1, float val2, float val3,
__local volatile float* smem1, __local volatile float* smem2, __local volatile float* smem3, int tid)
{
smem1[tid] = val1;
......@@ -150,7 +150,7 @@ static void reduce3(float val1, float val2, float val3,
barrier(CLK_LOCAL_MEM_FENCE);
}
static void reduce2(float val1, float val2, __local volatile float* smem1, __local volatile float* smem2, int tid)
inline void reduce2(float val1, float val2, __local volatile float* smem1, __local volatile float* smem2, int tid)
{
smem1[tid] = val1;
smem2[tid] = val2;
......@@ -189,7 +189,7 @@ static void reduce2(float val1, float val2, __local volatile float* smem1, __loc
barrier(CLK_LOCAL_MEM_FENCE);
}
static void reduce1(float val1, __local volatile float* smem1, int tid)
inline void reduce1(float val1, __local volatile float* smem1, int tid)
{
smem1[tid] = val1;
barrier(CLK_LOCAL_MEM_FENCE);
......@@ -225,7 +225,7 @@ static void reduce1(float val1, __local volatile float* smem1, int tid)
// Image read mode
__constant sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_LINEAR;
static void SetPatch(image2d_t I, float x, float y,
inline void SetPatch(image2d_t I, float x, float y,
float* Pch, float* Dx, float* Dy,
float* A11, float* A12, float* A22)
{
......@@ -262,7 +262,7 @@ inline void GetError(image2d_t J, const float x, const float y, const float* Pch
*errval += fabs(diff);
}
static void SetPatch4(image2d_t I, const float x, const float y,
inline void SetPatch4(image2d_t I, const float x, const float y,
float4* Pch, float4* Dx, float4* Dy,
float* A11, float* A12, float* A22)
{
......@@ -285,7 +285,7 @@ static void SetPatch4(image2d_t I, const float x, const float y,
*A22 += sqIdx.x + sqIdx.y + sqIdx.z;
}
static void GetPatch4(image2d_t J, const float x, const float y,
inline void GetPatch4(image2d_t J, const float x, const float y,
const float4* Pch, const float4* Dx, const float4* Dy,
float* b1, float* b2)
{
......@@ -297,7 +297,7 @@ static void GetPatch4(image2d_t J, const float x, const float y,
*b2 += xdiff.x + xdiff.y + xdiff.z;
}
static void GetError4(image2d_t J, const float x, const float y, const float4* Pch, float* errval)
inline void GetError4(image2d_t J, const float x, const float y, const float4* Pch, float* errval)
{
float4 diff = read_imagef(J, sampler, (float2)(x,y))-*Pch;
*errval += fabs(diff.x) + fabs(diff.y) + fabs(diff.z);
......
......@@ -97,7 +97,7 @@ inline float pix_diff_1(const uchar4 l, __global const uchar *rs)
return abs((int)(l.x) - *rs);
}
static float pix_diff_4(const uchar4 l, __global const uchar *rs)
inline float pix_diff_4(const uchar4 l, __global const uchar *rs)
{
uchar4 r;
r = *((__global uchar4 *)rs);
......@@ -233,7 +233,7 @@ __kernel void level_up_message(__global T *src, int src_rows, int src_step,
///////////////////////////////////////////////////////////////
//////////////////// calc all iterations /////////////////////
///////////////////////////////////////////////////////////////
static void message(__global T *us_, __global T *ds_, __global T *ls_, __global T *rs_,
inline void message(__global T *us_, __global T *ds_, __global T *ls_, __global T *rs_,
const __global T *dt,
int u_step, int msg_disp_step, int data_disp_step,
float4 cmax_disc_term, float4 cdisc_single_jump)
......
......@@ -62,7 +62,7 @@ __kernel void centeredGradientKernel(__global const float* src, int src_col, int
}
static float bicubicCoeff(float x_)
inline float bicubicCoeff(float x_)
{
float x = fabs(x_);
......@@ -156,7 +156,7 @@ __kernel void warpBackwardKernel(__global const float* I0, int I0_step, int I0_c
}
static float readImage(__global float *image, int x, int y, int rows, int cols, int elemCntPerRow)
inline float readImage(__global float *image, int x, int y, int rows, int cols, int elemCntPerRow)
{
int i0 = clamp(x, 0, cols - 1);
int j0 = clamp(y, 0, rows - 1);
......@@ -284,7 +284,7 @@ __kernel void estimateDualVariablesKernel(__global const float* u1, int u1_col,
}
static float divergence(__global const float* v1, __global const float* v2, int y, int x, int v1_step, int v2_step)
inline float divergence(__global const float* v1, __global const float* v2, int y, int x, int v1_step, int v2_step)
{
if (x > 0 && y > 0)
......
Markdown is supported
0% or
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment