some utility for GPU module internal purposes

diff --git a/modules/gpu/include/opencv2/gpu/gpu.hpp b/modules/gpu/include/opencv2/gpu/gpu.hpp
index 2a745c074ff10df25e5ec4137f420f5f83d396c7..081d4d342f9f5d8e03bb24449db97cf688bc5b3d 100644 (file)
--- a/modules/gpu/include/opencv2/gpu/gpu.hpp
+++ b/modules/gpu/include/opencv2/gpu/gpu.hpp
@@ -52,7 +52,7 @@ namespace cv
 {\r
     namespace gpu\r
     {\r
-        //////////////////////////////// Initialization ////////////////////////\r
+        //////////////////////////////// Initialization & Info ////////////////////////\r
 \r
         //! This is the only function that do not throw exceptions if the library is compiled without Cuda.\r
         CV_EXPORTS int getCudaEnabledDeviceCount();\r
@@ -67,6 +67,11 @@ namespace cv
 \r
         CV_EXPORTS void getGpuMemInfo(size_t& free, size_t& total);\r
 \r
+        //////////////////////////////// Error handling ////////////////////////\r
+        \r
+        CV_EXPORTS void error(const char *error_string, const char *file, const int line, const char *func);\r
+        CV_EXPORTS void nppError( int err, const char *file, const int line, const char *func);        \r
+\r
         //////////////////////////////// GpuMat ////////////////////////////////\r
         class Stream;\r
         class CudaMem;\r

diff --git a/modules/gpu/src/cuda/cuda_shared.hpp b/modules/gpu/src/cuda/cuda_shared.hpp
index c38d70be10efa5c72c6ab5f61cb09e822599bf95..b5049d73e1641bfc23c813ec1f83b4f675dc3443 100644 (file)
--- a/modules/gpu/src/cuda/cuda_shared.hpp
+++ b/modules/gpu/src/cuda/cuda_shared.hpp
@@ -58,8 +58,42 @@ namespace cv
 \r
         static inline int divUp(int total, int grain) { return (total + grain - 1) / grain; }\r
 \r
-        template<class T> \r
-        static inline void uploadConstant(const char* name, const T& value) { cudaSafeCall( cudaMemcpyToSymbol(name, &value, sizeof(T)) ); }\r
+        template<class T> static inline void uploadConstant(const char* name, const T& value) \r
+        { \r
+            cudaSafeCall( cudaMemcpyToSymbol(name, &value, sizeof(T)) ); \r
+        }\r
+\r
+        template<class T> static inline void uploadConstant(const char* name, const T& value, cudaStream_t stream) \r
+        { \r
+            cudaSafeCall( cudaMemcpyToSymbolAsyc(name, &value, sizeof(T), 0, cudaMemcpyHostToDevice, stream) ); \r
+        }        \r
+\r
+        template<class T> static inline void bindTexture(const char* name, const DevMem2D_<T>& img/*, bool normalized = false,\r
+            enum cudaTextureFilterMode filterMode = cudaFilterModePoint, enum cudaTextureAddressMode addrMode = cudaAddressModeClamp*/)\r
+        {            \r
+            //!!!! const_cast is disabled!\r
+            //!!!! Please use constructor of 'class texture'  instead.\r
+\r
+            //textureReference* tex; \r
+            //cudaSafeCall( cudaGetTextureReference((const textureReference**)&tex, name) ); \r
+            //tex->normalized = normalized;\r
+            //tex->filterMode = filterMode;\r
+            //tex->addressMode[0] = addrMode;\r
+            //tex->addressMode[1] = addrMode;\r
+            \r
+            const textureReference* tex; \r
+            cudaSafeCall( cudaGetTextureReference(&tex, name) ); \r
+\r
+            cudaChannelFormatDesc desc = cudaCreateChannelDesc<T>();\r
+            cudaSafeCall( cudaBindTexture2D(0, tex, img.ptr(), &desc, img.cols, img.rows, img.step) );\r
+        }\r
+\r
+        static inline void unbindTexture(const char *name)\r
+        {\r
+            const textureReference* tex; \r
+            cudaSafeCall( cudaGetTextureReference(&tex, name) ); \r
+            cudaSafeCall( cudaUnbindTexture(tex) );\r
+        }        \r
     }\r
 }\r
 \r

diff --git a/modules/gpu/src/cuda/safe_call.hpp b/modules/gpu/src/cuda/safe_call.hpp
index 2cb486092e0ae29772df82b126ab992898e4661c..b92ab4b0aba6750c3a30ddb38fc8c53e54bd3501 100644 (file)
--- a/modules/gpu/src/cuda/safe_call.hpp
+++ b/modules/gpu/src/cuda/safe_call.hpp
@@ -44,7 +44,7 @@
 #define __OPENCV_CUDA_SAFE_CALL_HPP__\r
 \r
 #include "cuda_runtime_api.h"\r
-#include <nppdefs.h>\r
+//#include <nppdefs.h>\r
 \r
 #if defined(__GNUC__)\r
     #define cudaSafeCall(expr)  ___cudaSafeCall(expr, __FILE__, __LINE__, __func__)\r
@@ -58,8 +58,8 @@ namespace cv
 {\r
     namespace gpu\r
     {\r
-        extern "C" void error( const char *error_string, const char *file, const int line, const char *func = "");\r
-        extern "C" void npp_error( int error, const char *file, const int line, const char *func = "");   \r
+        void error( const char *error_string, const char *file, const int line, const char *func = "");\r
+        void nppError( int error, const char *file, const int line, const char *func = "");   \r
 \r
         static inline void ___cudaSafeCall(cudaError_t err, const char *file, const int line, const char *func = "")\r
         {\r
@@ -67,10 +67,10 @@ namespace cv
                 cv::gpu::error(cudaGetErrorString(err), file, line, func);\r
         }\r
 \r
-        static inline void ___nppSafeCall(NppStatus err, const char *file, const int line, const char *func = "")\r
+        static inline void ___nppSafeCall(int err, const char *file, const int line, const char *func = "")\r
         {\r
             if (err < 0)\r
-                cv::gpu::npp_error(err, file, line, func);\r
+                cv::gpu::nppError(err, file, line, func);\r
         }\r
     }\r
 }\r

diff --git a/modules/gpu/src/error.cpp b/modules/gpu/src/error.cpp
index 04efd0e3041e8063b618e7052c1f8c5d007b8d57..432e9302b936bad110395afccc229ad4077c6a18 100644 (file)
--- a/modules/gpu/src/error.cpp
+++ b/modules/gpu/src/error.cpp
@@ -129,12 +129,12 @@ namespace cv
             return interpreter.str();\r
         }\r
 \r
-        extern "C" void npp_error( int err, const char *file, const int line, const char *func)\r
+        void nppError( int err, const char *file, const int line, const char *func)\r
         {                    \r
             cv::error( cv::Exception(CV_GpuNppCallError, getNppErrorString(err), func, file, line) );                \r
         }\r
 \r
-        extern "C" void error(const char *error_string, const char *file, const int line, const char *func)\r
+        void error(const char *error_string, const char *file, const int line, const char *func)\r
         {                       \r
             cv::error( cv::Exception(CV_GpuApiCallError, error_string, func, file, line) );\r
         }\r