From 55a722fc9dafca76b633a0f8d22254c237d638d3 Mon Sep 17 00:00:00 2001
From: Anatoly Baksheev <no@email>
Date: Wed, 10 Nov 2010 17:05:36 +0000
Subject: [PATCH] some utility for GPU module internal purposes
---
modules/gpu/include/opencv2/gpu/gpu.hpp | 7 ++++-
modules/gpu/src/cuda/cuda_shared.hpp | 38 +++++++++++++++++++++++--
modules/gpu/src/cuda/safe_call.hpp | 10 +++----
modules/gpu/src/error.cpp | 4 +--
4 files changed, 49 insertions(+), 10 deletions(-)
diff --git a/modules/gpu/include/opencv2/gpu/gpu.hpp b/modules/gpu/include/opencv2/gpu/gpu.hpp
index 2a745c074f..081d4d342f 100644
--- a/modules/gpu/include/opencv2/gpu/gpu.hpp
+++ b/modules/gpu/include/opencv2/gpu/gpu.hpp
@@ -52,7 +52,7 @@ namespace cv
{
namespace gpu
{
- //////////////////////////////// Initialization ////////////////////////
+ //////////////////////////////// Initialization & Info ////////////////////////
//! This is the only function that do not throw exceptions if the library is compiled without Cuda.
CV_EXPORTS int getCudaEnabledDeviceCount();
@@ -67,6 +67,11 @@ namespace cv
CV_EXPORTS void getGpuMemInfo(size_t& free, size_t& total);
+ //////////////////////////////// Error handling ////////////////////////
+
+ CV_EXPORTS void error(const char *error_string, const char *file, const int line, const char *func);
+ CV_EXPORTS void nppError( int err, const char *file, const int line, const char *func);
+
//////////////////////////////// GpuMat ////////////////////////////////
class Stream;
class CudaMem;
diff --git a/modules/gpu/src/cuda/cuda_shared.hpp b/modules/gpu/src/cuda/cuda_shared.hpp
index c38d70be10..b5049d73e1 100644
--- a/modules/gpu/src/cuda/cuda_shared.hpp
+++ b/modules/gpu/src/cuda/cuda_shared.hpp
@@ -58,8 +58,42 @@ namespace cv
static inline int divUp(int total, int grain) { return (total + grain - 1) / grain; }
- template<class T>
- static inline void uploadConstant(const char* name, const T& value) { cudaSafeCall( cudaMemcpyToSymbol(name, &value, sizeof(T)) ); }
+ template<class T> static inline void uploadConstant(const char* name, const T& value)
+ {
+ cudaSafeCall( cudaMemcpyToSymbol(name, &value, sizeof(T)) );
+ }
+
+ template<class T> static inline void uploadConstant(const char* name, const T& value, cudaStream_t stream)
+ {
+ cudaSafeCall( cudaMemcpyToSymbolAsyc(name, &value, sizeof(T), 0, cudaMemcpyHostToDevice, stream) );
+ }
+
+ template<class T> static inline void bindTexture(const char* name, const DevMem2D_<T>& img/*, bool normalized = false,
+ enum cudaTextureFilterMode filterMode = cudaFilterModePoint, enum cudaTextureAddressMode addrMode = cudaAddressModeClamp*/)
+ {
+ //!!!! const_cast is disabled!
+ //!!!! Please use constructor of 'class texture' instead.
+
+ //textureReference* tex;
+ //cudaSafeCall( cudaGetTextureReference((const textureReference**)&tex, name) );
+ //tex->normalized = normalized;
+ //tex->filterMode = filterMode;
+ //tex->addressMode[0] = addrMode;
+ //tex->addressMode[1] = addrMode;
+
+ const textureReference* tex;
+ cudaSafeCall( cudaGetTextureReference(&tex, name) );
+
+ cudaChannelFormatDesc desc = cudaCreateChannelDesc<T>();
+ cudaSafeCall( cudaBindTexture2D(0, tex, img.ptr(), &desc, img.cols, img.rows, img.step) );
+ }
+
+ static inline void unbindTexture(const char *name)
+ {
+ const textureReference* tex;
+ cudaSafeCall( cudaGetTextureReference(&tex, name) );
+ cudaSafeCall( cudaUnbindTexture(tex) );
+ }
}
}
diff --git a/modules/gpu/src/cuda/safe_call.hpp b/modules/gpu/src/cuda/safe_call.hpp
index 2cb486092e..b92ab4b0ab 100644
--- 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__
#include "cuda_runtime_api.h"
-#include <nppdefs.h>
+//#include <nppdefs.h>
#if defined(__GNUC__)
#define cudaSafeCall(expr) ___cudaSafeCall(expr, __FILE__, __LINE__, __func__)
@@ -58,8 +58,8 @@ namespace cv
{
namespace gpu
{
- extern "C" void error( const char *error_string, const char *file, const int line, const char *func = "");
- extern "C" void npp_error( int error, const char *file, const int line, const char *func = "");
+ void error( const char *error_string, const char *file, const int line, const char *func = "");
+ void nppError( int error, const char *file, const int line, const char *func = "");
static inline void ___cudaSafeCall(cudaError_t err, const char *file, const int line, const char *func = "")
{
@@ -67,10 +67,10 @@ namespace cv
cv::gpu::error(cudaGetErrorString(err), file, line, func);
}
- static inline void ___nppSafeCall(NppStatus err, const char *file, const int line, const char *func = "")
+ static inline void ___nppSafeCall(int err, const char *file, const int line, const char *func = "")
{
if (err < 0)
- cv::gpu::npp_error(err, file, line, func);
+ cv::gpu::nppError(err, file, line, func);
}
}
}
diff --git a/modules/gpu/src/error.cpp b/modules/gpu/src/error.cpp
index 04efd0e304..432e9302b9 100644
--- a/modules/gpu/src/error.cpp
+++ b/modules/gpu/src/error.cpp
@@ -129,12 +129,12 @@ namespace cv
return interpreter.str();
}
- extern "C" void npp_error( int err, const char *file, const int line, const char *func)
+ void nppError( int err, const char *file, const int line, const char *func)
{
cv::error( cv::Exception(CV_GpuNppCallError, getNppErrorString(err), func, file, line) );
}
- extern "C" void error(const char *error_string, const char *file, const int line, const char *func)
+ void error(const char *error_string, const char *file, const int line, const char *func)
{
cv::error( cv::Exception(CV_GpuApiCallError, error_string, func, file, line) );
}