Merge pull request #719 from jet47:gpu-fix-build

modules/core/include/opencv2/core/cvdef.h

@@ -111,6 +111,9 @@
 #define CV_CPU_NEON   11
 #define CV_HARDWARE_MAX_FEATURE 255
 
+// disable SSE/AVX/NEON headers for NVCC compiler
+#ifndef __CUDACC__
+
 #if defined __SSE2__ || defined _M_X64  || (defined _M_IX86_FP && _M_IX86_FP >= 2)
 #  include <emmintrin.h>
 #  define CV_SSE 1
@@ -149,6 +152,8 @@
 #  define CV_NEON 1
 #endif
 
+#endif // __CUDACC__
+
 #ifndef CV_SSE
 #  define CV_SSE 0
 #endif
@@ -336,7 +341,7 @@ typedef signed char schar;
 
 CV_INLINE int cvRound( double value )
 {
-#if (defined _MSC_VER && defined _M_X64) || (defined __GNUC__ && defined __x86_64__ && defined __SSE2__ && !defined __APPLE__)
+#if ((defined _MSC_VER && defined _M_X64) || (defined __GNUC__ && defined __x86_64__ && defined __SSE2__ && !defined __APPLE__)) && !defined(__CUDACC__)
     __m128d t = _mm_set_sd( value );
     return _mm_cvtsd_si32(t);
 #elif defined _MSC_VER && defined _M_IX86
@@ -361,7 +366,7 @@ CV_INLINE int cvRound( double value )
 
 CV_INLINE int cvFloor( double value )
 {
-#if defined _MSC_VER && defined _M_X64 || (defined __GNUC__ && defined __SSE2__ && !defined __APPLE__)
+#if (defined _MSC_VER && defined _M_X64 || (defined __GNUC__ && defined __SSE2__ && !defined __APPLE__)) && !defined(__CUDACC__)
     __m128d t = _mm_set_sd( value );
     int i = _mm_cvtsd_si32(t);
     return i - _mm_movemask_pd(_mm_cmplt_sd(t, _mm_cvtsi32_sd(t,i)));
@@ -377,7 +382,7 @@ CV_INLINE int cvFloor( double value )
 
 CV_INLINE int cvCeil( double value )
 {
-#if defined _MSC_VER && defined _M_X64 || (defined __GNUC__ && defined __SSE2__&& !defined __APPLE__)
+#if (defined _MSC_VER && defined _M_X64 || (defined __GNUC__ && defined __SSE2__&& !defined __APPLE__)) && !defined(__CUDACC__)
     __m128d t = _mm_set_sd( value );
     int i = _mm_cvtsd_si32(t);
     return i + _mm_movemask_pd(_mm_cmplt_sd(_mm_cvtsi32_sd(t,i), t));

modules/core/include/opencv2/core/cvstd.hpp

@@ -52,6 +52,7 @@
 
 #include <cstddef>
 #include <cstring>
+#include <cctype>
 
 #ifndef OPENCV_NOSTL
 #  include <string>
@@ -166,6 +167,8 @@ public:
     friend String operator+ (const String& lhs, char          rhs);
     friend String operator+ (char          lhs, const String& rhs);
 
+    String toLowerCase() const;
+
 #ifndef OPENCV_NOSTL
     String(const std::string& str);
     String(const std::string& str, size_t pos, size_t len = npos);
@@ -482,6 +485,16 @@ inline size_t String::find_last_of(const char* s, size_t pos) const
     return npos;
 }
 
+inline String String::toLowerCase() const
+{
+    String res(cstr_, len_);
+
+    for (size_t i = 0; i < len_; ++i)
+        res.cstr_[i] = (char) ::tolower(cstr_[i]);
+
+    return res;
+}
+
 // ************************* cv::String non-member functions *************************
 
 inline String operator+ (const String& lhs, const String& rhs)

modules/core/include/opencv2/core/utility.hpp

@@ -358,7 +358,7 @@ AutoBuffer<_Tp, fixed_size>::AutoBuffer(const AutoBuffer<_Tp, fixed_size>& abuf
 {
     ptr = buf;
     sz = fixed_size;
-    allocate(abuf.size);
+    allocate(abuf.size());
     for( size_t i = 0; i < sz; i++ )
         ptr[i] = abuf.ptr[i];
 }
@@ -369,7 +369,7 @@ AutoBuffer<_Tp, fixed_size>::operator = (const AutoBuffer<_Tp, fixed_size>& abuf
     if( this != &abuf )
     {
         deallocate();
-        allocate(abuf.size);
+        allocate(abuf.size());
         for( size_t i = 0; i < sz; i++ )
             ptr[i] = abuf.ptr[i];
     }

modules/core/src/cudastream.cpp

@@ -41,7 +41,6 @@
 //M*/
 
 #include "precomp.hpp"
-#include "opencv2/core/gpumat.hpp"
 
 using namespace cv;
 using namespace cv::gpu;
@@ -272,7 +271,7 @@ void cv::gpu::Stream::enqueueConvert(const GpuMat& src, GpuMat& dst, int dtype,
     convertTo(src, dst, alpha, beta, stream);
 }
 
-#if CUDA_VERSION >= 5000
+#if CUDART_VERSION >= 5000
 
 namespace
 {
@@ -295,7 +294,7 @@ namespace
 
 void cv::gpu::Stream::enqueueHostCallback(StreamCallback callback, void* userData)
 {
-#if CUDA_VERSION >= 5000
+#if CUDART_VERSION >= 5000
     CallbackData* data = new CallbackData;
     data->callback = callback;
     data->userData = userData;

modules/core/src/gpumat.cpp

@@ -41,24 +41,6 @@
 //M*/
 
 #include "precomp.hpp"
-#include "opencv2/core/gpumat.hpp"
-#include <cctype>
-
-#ifdef HAVE_CUDA
-    #include <cuda_runtime.h>
-    #include <npp.h>
-
-    #define CUDART_MINIMUM_REQUIRED_VERSION 4020
-    #define NPP_MINIMUM_REQUIRED_VERSION 4200
-
-    #if (CUDART_VERSION < CUDART_MINIMUM_REQUIRED_VERSION)
-        #error "Insufficient Cuda Runtime library version, please update it."
-    #endif
-
-    #if (NPP_VERSION_MAJOR * 1000 + NPP_VERSION_MINOR * 100 + NPP_VERSION_BUILD < NPP_MINIMUM_REQUIRED_VERSION)
-        #error "Insufficient NPP version, please update it."
-    #endif
-#endif
 
 using namespace cv;
 using namespace cv::gpu;
@@ -233,7 +215,7 @@ namespace
                 int cur_value;
                 int chars_read;
                 int args_read = sscanf(set_as_str.c_str() + pos, "%d%n", &cur_value, &chars_read);
-                CV_Assert(args_read == 2);
+                CV_Assert(args_read == 1);
 
                 arr.push_back(cur_value);
                 pos += chars_read;

modules/core/src/precomp.hpp

@@ -51,6 +51,7 @@
 #include "opencv2/core/utility.hpp"
 #include "opencv2/core/core_c.h"
 #include "opencv2/core/internal.hpp"
+#include "opencv2/core/gpumat.hpp"
 
 #include <assert.h>
 #include <ctype.h>
@@ -68,8 +69,20 @@
 #endif
 
 #ifdef HAVE_CUDA
-#  include <cuda_runtime_api.h>
-#  include "opencv2/core/gpumat.hpp"
+
+#  include <cuda_runtime.h>
+#  include <npp.h>
+
+#  define CUDART_MINIMUM_REQUIRED_VERSION 4020
+#  define NPP_MINIMUM_REQUIRED_VERSION 4200
+
+#  if (CUDART_VERSION < CUDART_MINIMUM_REQUIRED_VERSION)
+#    error "Insufficient Cuda Runtime library version, please update it."
+#  endif
+
+#  if (NPP_VERSION_MAJOR * 1000 + NPP_VERSION_MINOR * 100 + NPP_VERSION_BUILD < NPP_MINIMUM_REQUIRED_VERSION)
+#    error "Insufficient NPP version, please update it."
+#  endif
 
 #  if defined(__GNUC__)
 #    define cudaSafeCall(expr)  ___cudaSafeCall(expr, __FILE__, __LINE__, __func__)

modules/gpu/src/cascadeclassifier.cpp

@@ -693,7 +693,7 @@ bool cv::gpu::CascadeClassifier_GPU::load(const String& filename)
     release();
 
     String fext = filename.substr(filename.find_last_of(".") + 1);
-    std::transform(fext.begin(), fext.end(), fext.begin(), ::tolower);
+    fext = fext.toLowerCase();
 
     if (fext == "nvbin")
     {

modules/gpu/src/cu_safe_call.cpp

@@ -51,7 +51,7 @@ namespace
     struct ErrorEntry
     {
         int code;
-        String str;
+        const char* str;
     };
 
     class ErrorEntryComparer
@@ -65,16 +65,14 @@ namespace
         int code_;
     };
 
-    String getErrorString(int code, const ErrorEntry* errors, size_t n)
+    cv::String getErrorString(int code, const ErrorEntry* errors, size_t n)
     {
         size_t idx = std::find_if(errors, errors + n, ErrorEntryComparer(code)) - errors;
 
-        const String& msg = (idx != n) ? errors[idx].str : String("Unknown error code");
+        const char* msg = (idx != n) ? errors[idx].str : "Unknown error code";
+        cv::String str = cv::format("%s [Code = %d]", msg, code);
 
-        std::ostringstream ostr;
-        ostr << msg << " [Code = " << code << "]";
-
-        return ostr.str();
+        return str;
     }
 
     const ErrorEntry cu_errors [] =
@@ -131,7 +129,7 @@ namespace
     const size_t cu_errors_num = sizeof(cu_errors) / sizeof(cu_errors[0]);
 }
 
-String cv::gpu::detail::cuGetErrString(CUresult res)
+cv::String cv::gpu::detail::cuGetErrString(CUresult res)
 {
     return getErrorString(res, cu_errors, cu_errors_num);
 }

modules/gpu/src/cuda/bilateral_filter.cu

@@ -43,8 +43,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
-
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/vec_traits.hpp"
 #include "opencv2/gpu/device/vec_math.hpp"
 #include "opencv2/gpu/device/border_interpolate.hpp"
@@ -198,4 +197,4 @@ OCV_INSTANTIATE_BILATERAL_FILTER(float3)
 OCV_INSTANTIATE_BILATERAL_FILTER(float4)
 
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/blend.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 
 namespace cv { namespace gpu { namespace device
 {
@@ -118,4 +118,4 @@ namespace cv { namespace gpu { namespace device
 }}} // namespace cv { namespace gpu { namespace device
 
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/color.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/transform.hpp"
 #include "opencv2/gpu/device/color.hpp"
 #include "cvt_color_internal.h"

modules/gpu/src/cuda/copy_make_border.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/border_interpolate.hpp"
 
 namespace cv { namespace gpu { namespace device
@@ -128,4 +128,4 @@ namespace cv { namespace gpu { namespace device
     } // namespace imgproc
 }}} // namespace cv { namespace gpu { namespace device
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/disp_bilateral_filter.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/limits.hpp"
 
 namespace cv { namespace gpu { namespace device
@@ -220,4 +220,4 @@ namespace cv { namespace gpu { namespace device
     } // namespace bilateral_filter
 }}} // namespace cv { namespace gpu { namespace device
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/imgproc.cu

@@ -42,11 +42,12 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/vec_traits.hpp"
 #include "opencv2/gpu/device/vec_math.hpp"
 #include "opencv2/gpu/device/saturate_cast.hpp"
 #include "opencv2/gpu/device/border_interpolate.hpp"
+#include "internal_shared.hpp"
 
 namespace cv { namespace gpu { namespace device
 {

modules/gpu/src/cuda/lbp.hpp

@@ -43,8 +43,8 @@
 #ifndef __OPENCV_GPU_DEVICE_LBP_HPP_
 #define __OPENCV_GPU_DEVICE_LBP_HPP_
 
-#include "internal_shared.hpp"
-#include <opencv2/gpu/device/emulation.hpp>
+#include "opencv2/gpu/device/common.hpp"
+#include "opencv2/gpu/device/emulation.hpp"
 
 namespace cv { namespace gpu { namespace device {
 
@@ -72,10 +72,10 @@ namespace lbp {
 
         __device__ __forceinline__ bool operator()(const int4& r1, const int4& r2) const
         {
-            float delta = eps * (min(r1.z, r2.z) + min(r1.w, r2.w)) * 0.5f;
+            float delta = eps * (::min(r1.z, r2.z) + ::min(r1.w, r2.w)) * 0.5f;
 
-            return abs(r1.x - r2.x) <= delta && abs(r1.y - r2.y) <= delta
-                && abs(r1.x + r1.z - r2.x - r2.z) <= delta && abs(r1.y + r1.w - r2.y - r2.w) <= delta;
+            return ::abs(r1.x - r2.x) <= delta && ::abs(r1.y - r2.y) <= delta
+                && ::abs(r1.x + r1.z - r2.x - r2.z) <= delta && ::abs(r1.y + r1.w - r2.y - r2.w) <= delta;
         }
         float eps;
     };
@@ -109,4 +109,4 @@ namespace lbp {
 
 } } }// namespaces
 
-#endif
+#endif
\ No newline at end of file

modules/gpu/src/cuda/match_template.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/vec_math.hpp"
 
 namespace cv { namespace gpu { namespace device
@@ -913,4 +913,4 @@ namespace cv { namespace gpu { namespace device
 }}} // namespace cv { namespace gpu { namespace device
 
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/mathfunc.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 
 namespace cv { namespace gpu { namespace device
 {

modules/gpu/src/cuda/optical_flow_farneback.cu

@@ -42,7 +42,6 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
 #include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/border_interpolate.hpp"
 
@@ -645,4 +644,4 @@ namespace cv { namespace gpu { namespace device { namespace optflow_farneback
 }}}} // namespace cv { namespace gpu { namespace device { namespace optflow_farneback
 
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/pyr_up.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/border_interpolate.hpp"
 #include "opencv2/gpu/device/vec_traits.hpp"
 #include "opencv2/gpu/device/vec_math.hpp"
@@ -193,4 +193,4 @@ namespace cv { namespace gpu { namespace device
     } // namespace imgproc
 }}} // namespace cv { namespace gpu { namespace device
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/remap.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/border_interpolate.hpp"
 #include "opencv2/gpu/device/vec_traits.hpp"
 #include "opencv2/gpu/device/vec_math.hpp"

modules/gpu/src/cuda/resize.cu

@@ -42,14 +42,14 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include <cfloat>
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/border_interpolate.hpp"
 #include "opencv2/gpu/device/vec_traits.hpp"
 #include "opencv2/gpu/device/vec_math.hpp"
 #include "opencv2/gpu/device/saturate_cast.hpp"
 #include "opencv2/gpu/device/filters.hpp"
-#include <cfloat>
-#include <opencv2/gpu/device/scan.hpp>
+#include "opencv2/gpu/device/scan.hpp"
 
 namespace cv { namespace gpu { namespace device
 {
@@ -299,4 +299,4 @@ namespace cv { namespace gpu { namespace device
 }}} // namespace cv { namespace gpu { namespace device
 
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/split_merge.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 
 namespace cv { namespace gpu { namespace device
 {

modules/gpu/src/cuda/stereobm.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 
 namespace cv { namespace gpu { namespace device
 {
@@ -537,4 +537,4 @@ namespace cv { namespace gpu { namespace device
 }}} // namespace cv { namespace gpu { namespace device
 
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/stereobp.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/saturate_cast.hpp"
 #include "opencv2/gpu/device/limits.hpp"
 
@@ -535,4 +535,4 @@ namespace cv { namespace gpu { namespace device
     } // namespace stereobp
 }}} // namespace cv { namespace gpu { namespace device
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/cuda/warp.cu

@@ -42,7 +42,7 @@
 
 #if !defined CUDA_DISABLER
 
-#include "internal_shared.hpp"
+#include "opencv2/gpu/device/common.hpp"
 #include "opencv2/gpu/device/border_interpolate.hpp"
 #include "opencv2/gpu/device/vec_traits.hpp"
 #include "opencv2/gpu/device/vec_math.hpp"

modules/gpu/src/error.cpp

@@ -54,7 +54,7 @@ namespace
     struct ErrorEntry
     {
         int code;
-        String str;
+        const char* str;
     };
 
     struct ErrorEntryComparer
@@ -68,12 +68,10 @@ namespace
     {
         size_t idx = std::find_if(errors, errors + n, ErrorEntryComparer(code)) - errors;
 
-        const String& msg = (idx != n) ? errors[idx].str : String("Unknown error code");
+        const char* msg = (idx != n) ? errors[idx].str : "Unknown error code";
+        String str = cv::format("%s [Code = %d]", msg, code);
 
-        std::ostringstream ostr;
-        ostr << msg << " [Code = " << code << "]";
-
-        return ostr.str();
+        return str;
     }
 
     //////////////////////////////////////////////////////////////////////////

modules/gpu/src/nvidia/NCVHaarObjectDetection.cu

@@ -2099,7 +2099,7 @@ NCVStatus ncvGrowDetectionsVector_host(NCVVector<Ncv32u> &pixelMask,
 }
 
 
-NCVStatus loadFromXML(const String &filename,
+NCVStatus loadFromXML(const cv::String &filename,
                       HaarClassifierCascadeDescriptor &haar,
                       std::vector<HaarStage64> &haarStages,
                       std::vector<HaarClassifierNode128> &haarClassifierNodes,
@@ -2110,7 +2110,7 @@ NCVStatus loadFromXML(const String &filename,
 #define NVBIN_HAAR_VERSION          0x1
 
 
-static NCVStatus loadFromNVBIN(const String &filename,
+static NCVStatus loadFromNVBIN(const cv::String &filename,
                                HaarClassifierCascadeDescriptor &haar,
                                std::vector<HaarStage64> &haarStages,
                                std::vector<HaarClassifierNode128> &haarClassifierNodes,
@@ -2174,14 +2174,14 @@ static NCVStatus loadFromNVBIN(const String &filename,
 }
 
 
-NCVStatus ncvHaarGetClassifierSize(const String &filename, Ncv32u &numStages,
+NCVStatus ncvHaarGetClassifierSize(const cv::String &filename, Ncv32u &numStages,
                                    Ncv32u &numNodes, Ncv32u &numFeatures)
 {
     size_t readCount;
     NCVStatus ncvStat;
 
-    String fext = filename.substr(filename.find_last_of(".") + 1);
-    std::transform(fext.begin(), fext.end(), fext.begin(), ::tolower);
+    cv::String fext = filename.substr(filename.find_last_of(".") + 1);
+    fext = fext.toLowerCase();
 
     if (fext == "nvbin")
     {
@@ -2226,7 +2226,7 @@ NCVStatus ncvHaarGetClassifierSize(const String &filename, Ncv32u &numStages,
 }
 
 
-NCVStatus ncvHaarLoadFromFile_host(const String &filename,
+NCVStatus ncvHaarLoadFromFile_host(const cv::String &filename,
                                    HaarClassifierCascadeDescriptor &haar,
                                    NCVVector<HaarStage64> &h_HaarStages,
                                    NCVVector<HaarClassifierNode128> &h_HaarNodes,
@@ -2238,8 +2238,8 @@ NCVStatus ncvHaarLoadFromFile_host(const String &filename,
 
     NCVStatus ncvStat;
 
-    String fext = filename.substr(filename.find_last_of(".") + 1);
-    std::transform(fext.begin(), fext.end(), fext.begin(), ::tolower);
+    cv::String fext = filename.substr(filename.find_last_of(".") + 1);
+    fext = fext.toLowerCase();
 
     std::vector<HaarStage64> haarStages;
     std::vector<HaarClassifierNode128> haarNodes;
@@ -2272,7 +2272,7 @@ NCVStatus ncvHaarLoadFromFile_host(const String &filename,
 }
 
 
-NCVStatus ncvHaarStoreNVBIN_host(const String &filename,
+NCVStatus ncvHaarStoreNVBIN_host(const cv::String &filename,
                                  HaarClassifierCascadeDescriptor haar,
                                  NCVVector<HaarStage64> &h_HaarStages,
                                  NCVVector<HaarClassifierNode128> &h_HaarNodes,

modules/gpu/src/nvidia/NCVHaarObjectDetection.hpp

@@ -438,18 +438,18 @@ NCV_EXPORTS NCVStatus ncvGrowDetectionsVector_host(NCVVector<Ncv32u> &pixelMask,
                                                    Ncv32f curScale);
 
 
-NCV_EXPORTS NCVStatus ncvHaarGetClassifierSize(const String &filename, Ncv32u &numStages,
+NCV_EXPORTS NCVStatus ncvHaarGetClassifierSize(const cv::String &filename, Ncv32u &numStages,
                                                Ncv32u &numNodes, Ncv32u &numFeatures);
 
 
-NCV_EXPORTS NCVStatus ncvHaarLoadFromFile_host(const String &filename,
+NCV_EXPORTS NCVStatus ncvHaarLoadFromFile_host(const cv::String &filename,
                                                HaarClassifierCascadeDescriptor &haar,
                                                NCVVector<HaarStage64> &h_HaarStages,
                                                NCVVector<HaarClassifierNode128> &h_HaarNodes,
                                                NCVVector<HaarFeature64> &h_HaarFeatures);
 
 
-NCV_EXPORTS NCVStatus ncvHaarStoreNVBIN_host(const String &filename,
+NCV_EXPORTS NCVStatus ncvHaarStoreNVBIN_host(const cv::String &filename,
                                              HaarClassifierCascadeDescriptor haar,
                                              NCVVector<HaarStage64> &h_HaarStages,
                                              NCVVector<HaarClassifierNode128> &h_HaarNodes,
@@ -457,4 +457,4 @@ NCV_EXPORTS NCVStatus ncvHaarStoreNVBIN_host(const String &filename,
 
 
 
-#endif // _ncvhaarobjectdetection_hpp_
+#endif // _ncvhaarobjectdetection_hpp_
\ No newline at end of file

modules/gpu/src/nvidia/core/NCV.cu

@@ -52,16 +52,16 @@
 //==============================================================================
 
 
-static void stdDebugOutput(const String &msg)
+static void stdDebugOutput(const cv::String &msg)
 {
-    std::cout << msg;
+    std::cout << msg.c_str() << std::endl;
 }
 
 
 static NCVDebugOutputHandler *debugOutputHandler = stdDebugOutput;
 
 
-void ncvDebugOutput(const String &msg)
+void ncvDebugOutput(const cv::String &msg)
 {
     debugOutputHandler(msg);
 }
@@ -905,4 +905,4 @@ NCVStatus ncvDrawRects_32u_device(Ncv32u *d_dst,
     return drawRectsWrapperDevice(d_dst, dstStride, dstWidth, dstHeight, d_rects, numRects, color, cuStream);
 }
 
-#endif /* CUDA_DISABLER */
+#endif /* CUDA_DISABLER */
\ No newline at end of file

modules/gpu/src/nvidia/core/NCV.hpp

@@ -53,8 +53,8 @@
 #endif
 
 #include <cuda_runtime.h>
-#include <sstream>
-#include <iostream>
+#include "opencv2/core/cvstd.hpp"
+#include "opencv2/core/utility.hpp"
 
 
 //==============================================================================
@@ -243,10 +243,10 @@ const Ncv32u K_LOG2_WARP_SIZE = 5;
 //==============================================================================
 
 
-NCV_EXPORTS void ncvDebugOutput(const String &msg);
+NCV_EXPORTS void ncvDebugOutput(const cv::String &msg);
 
 
-typedef void NCVDebugOutputHandler(const String &msg);
+typedef void NCVDebugOutputHandler(const cv::String &msg);
 
 
 NCV_EXPORTS void ncvSetDebugOutputHandler(NCVDebugOutputHandler* func);
@@ -257,9 +257,8 @@ NCV_EXPORTS void ncvSetDebugOutputHandler(NCVDebugOutputHandler* func);
     { \
         if (!(pred)) \
         { \
-            std::ostringstream oss; \
-            oss << "NCV Assertion Failed: " << msg << ", file=" << __FILE__ << ", line=" << __LINE__ << std::endl; \
-            ncvDebugOutput(oss.str()); \
+            cv::String str = cv::format("NCV Assertion Failed: %s, file=%s, line=%d", msg, __FILE__, __LINE__); \
+            ncvDebugOutput(str); \
         } \
     } while (0)
 
@@ -273,14 +272,19 @@ NCV_EXPORTS void ncvSetDebugOutputHandler(NCVDebugOutputHandler* func);
 
 
 #define ncvAssertReturn(pred, err) \
-    ncvAssertPrintReturn(pred, "retcode=" << (int)err, err)
+    do \
+    { \
+        cv::String msg = cv::format("retcode=%d", (int)err); \
+        ncvAssertPrintReturn(pred, msg.c_str(), err); \
+    } while (0)
 
 
 #define ncvAssertReturnNcvStat(ncvOp) \
     do \
     { \
         NCVStatus _ncvStat = ncvOp; \
-        ncvAssertPrintReturn(NCV_SUCCESS==_ncvStat, "NcvStat=" << (int)_ncvStat, _ncvStat); \
+        cv::String msg = cv::format("NcvStat=%d", (int)_ncvStat); \
+        ncvAssertPrintReturn(NCV_SUCCESS==_ncvStat, msg.c_str(), _ncvStat); \
     } while (0)
 
 
@@ -288,7 +292,8 @@ NCV_EXPORTS void ncvSetDebugOutputHandler(NCVDebugOutputHandler* func);
     do \
     { \
         cudaError_t res = cudacall; \
-        ncvAssertPrintReturn(cudaSuccess==res, "cudaError_t=" << (int)res, errCode); \
+        cv::String msg = cv::format("cudaError_t=%d", (int)res); \
+        ncvAssertPrintReturn(cudaSuccess==res, msg.c_str(), errCode); \
     } while (0)
 
 
@@ -296,7 +301,8 @@ NCV_EXPORTS void ncvSetDebugOutputHandler(NCVDebugOutputHandler* func);
     do \
     { \
         cudaError_t res = cudaGetLastError(); \
-        ncvAssertPrintReturn(cudaSuccess==res, "cudaError_t=" << (int)res, errCode); \
+        cv::String msg = cv::format("cudaError_t=%d", (int)res); \
+        ncvAssertPrintReturn(cudaSuccess==res, msg.c_str(), errCode); \
     } while (0)
 
 
@@ -805,7 +811,7 @@ public:
     T& at(Ncv32u x, Ncv32u y) const
     {
         NcvBool bOutRange = (x >= this->_width || y >= this->_height);
-        ncvAssertPrintCheck(!bOutRange, "Error addressing matrix at [" << x << ", " << y << "]");
+        ncvAssertPrintCheck(!bOutRange, "Error addressing matrix");
         if (bOutRange)
         {
             return *this->_ptr;

modules/gpu/src/precomp.hpp

@@ -65,6 +65,8 @@
 #include <stdexcept>
 #include <memory>
 
+#include "opencv2/core.hpp"
+#include "opencv2/core/utility.hpp"
 #include "opencv2/gpu.hpp"
 #include "opencv2/imgproc.hpp"
 #include "opencv2/imgproc/imgproc_c.h"

modules/gpu/src/video_writer.cpp

@@ -736,7 +736,7 @@ void NVENCAPI cv::gpu::VideoWriter_GPU::Impl::HandleOnEndFrame(const NVVE_EndFra
 class EncoderCallBackFFMPEG : public cv::gpu::VideoWriter_GPU::EncoderCallBack
 {
 public:
-    EncoderCallBackFFMPEG(const String& fileName, cv::Size frameSize, double fps);
+    EncoderCallBackFFMPEG(const cv::String& fileName, cv::Size frameSize, double fps);
     ~EncoderCallBackFFMPEG();
 
     unsigned char* acquireBitStream(int* bufferSize);
@@ -799,7 +799,7 @@ namespace
     }
 }
 
-EncoderCallBackFFMPEG::EncoderCallBackFFMPEG(const String& fileName, cv::Size frameSize, double fps) :
+EncoderCallBackFFMPEG::EncoderCallBackFFMPEG(const cv::String& fileName, cv::Size frameSize, double fps) :
     stream_(0), isKeyFrame_(false)
 {
     int buf_size = std::max(frameSize.area() * 4, 1024 * 1024);

modules/gpu/test/nvidia/main_nvidia.cpp

@@ -271,7 +271,7 @@ void generateHaarApplicationTests(NCVAutoTestLister &testLister, NCVTestSourcePr
     }
 }
 
-static void devNullOutput(const std::string& msg)
+static void devNullOutput(const cv::String& msg)
 {
     (void)msg;
 }

modules/gpu/test/test_stream.cpp

@@ -46,7 +46,7 @@
 
 using namespace cvtest;
 
-#if CUDA_VERSION >= 5000
+#if CUDART_VERSION >= 5000
 
 struct Async : testing::TestWithParam<cv::gpu::DeviceInfo>
 {

modules/nonfree/src/surf_gpu.cpp

@@ -223,6 +223,9 @@ namespace
         }
 
     private:
+        SURF_GPU_Invoker(const SURF_GPU_Invoker&);
+        SURF_GPU_Invoker& operator =(const SURF_GPU_Invoker&);
+
         SURF_GPU& surf_;
 
         int img_cols, img_rows;

modules/softcascade/src/cuda/icf-sc.cu

@@ -82,8 +82,8 @@ typedef unsigned char uchar;
     }
 
     template<int FACTOR>
-    __global__ void shrink(const uchar* __restrict__ hogluv, const int inPitch,
-                                 uchar* __restrict__ shrank, const int outPitch )
+    __global__ void shrink(const uchar* __restrict__ hogluv, const size_t inPitch,
+                                 uchar* __restrict__ shrank, const size_t outPitch )
     {
         const int y = blockIdx.y * blockDim.y + threadIdx.y;
         const int x = blockIdx.x * blockDim.x + threadIdx.x;
@@ -127,7 +127,7 @@ typedef unsigned char uchar;
         __v = static_cast<uchar>((V + 140.f) * (255.f / (122.f + 140.f )));
     }
 
-    __global__ void bgr2Luv_d(const uchar* rgb, const int rgbPitch, uchar* luvg, const int luvgPitch)
+    __global__ void bgr2Luv_d(const uchar* rgb, const size_t rgbPitch, uchar* luvg, const size_t luvgPitch)
     {
         const int y = blockIdx.y * blockDim.y + threadIdx.y;
         const int x = blockIdx.x * blockDim.x + threadIdx.x;
@@ -256,8 +256,8 @@ typedef unsigned char uchar;
 
     // ToDo: use textures or uncached load instruction.
     __global__ void magToHist(const uchar* __restrict__ mag,
-                              const float* __restrict__ angle, const int angPitch,
-                                    uchar* __restrict__ hog,   const int hogPitch, const int fh)
+                              const float* __restrict__ angle, const size_t angPitch,
+                                    uchar* __restrict__ hog,   const size_t hogPitch, const int fh)
     {
         const int y = blockIdx.y * blockDim.y + threadIdx.y;
         const int x = blockIdx.x * blockDim.x + threadIdx.x;

modules/softcascade/src/detector_cuda.cpp

@@ -274,7 +274,7 @@ struct cv::softcascade::SCascade::Fields
 
     bool check(float mins,float  maxs, int scales)
     {
-        bool updated = ((minScale == mins) || (maxScale == maxs) || (totals = scales));
+        bool updated = ((minScale == mins) || (maxScale == maxs) || (totals == scales));
 
         minScale = mins;
         maxScale = maxScale;