1
0
mirror of https://github.com/opencv/opencv.git synced 2026-07-31 08:13:04 +04:00

fixed compilation errors

This commit is contained in:
Vladislav Vinogradov
2013-04-17 15:24:53 +04:00
parent 8fdab9f631
commit 54e7c76d99
41 changed files with 205 additions and 234 deletions
+5 -23
View File
@@ -41,25 +41,7 @@
//M*/
#include "opencv2/core/cuda_devptrs.hpp"
namespace cv { namespace softcascade { namespace internal {
void error(const char *error_string, const char *file, const int line, const char *func);
}}}
#if defined(__GNUC__)
#define cudaSafeCall(expr) ___cudaSafeCall(expr, __FILE__, __LINE__, __func__)
#else /* defined(__CUDACC__) || defined(__MSVC__) */
#define cudaSafeCall(expr) ___cudaSafeCall(expr, __FILE__, __LINE__)
#endif
static inline void ___cudaSafeCall(cudaError_t err, const char *file, const int line, const char *func = "")
{
if (cudaSuccess != err) cv::softcascade::internal::error(cudaGetErrorString(err), file, line, func);
}
__host__ __device__ __forceinline__ int divUp(int total, int grain)
{
return (total + grain - 1) / grain;
}
#include "opencv2/core/cuda/common.hpp"
namespace cv { namespace softcascade { namespace cudev
{
@@ -393,7 +375,7 @@ namespace cv { namespace softcascade { namespace cudev
{
const dim3 block(32, 8);
const dim3 grid(divUp(integral.cols, block.x), 1);
const dim3 grid(cv::gpu::cudev::divUp(integral.cols, block.x), 1);
shfl_integral_vertical<<<grid, block, 0, stream>>>(integral);
cudaSafeCall( cudaGetLastError() );
@@ -478,7 +460,7 @@ namespace cv { namespace softcascade { namespace cudev
{
const dim3 block(32, 8);
const dim3 grid(divUp(integral.cols, block.x), 1);
const dim3 grid(cv::gpu::cudev::divUp(integral.cols, block.x), 1);
shfl_integral_vertical<<<grid, block, 0, stream>>>((cv::gpu::PtrStepSz<unsigned int>)buffer, integral);
cudaSafeCall( cudaGetLastError() );
@@ -518,8 +500,8 @@ namespace cv { namespace softcascade { namespace cudev
void transform(const cv::gpu::PtrStepSz<uchar3>& bgr, cv::gpu::PtrStepSzb gray)
{
const dim3 block(32, 8);
const dim3 grid(divUp(bgr.cols, block.x), divUp(bgr.rows, block.y));
const dim3 grid(cv::gpu::cudev::divUp(bgr.cols, block.x), cv::gpu::cudev::divUp(bgr.rows, block.y));
device_transform<<<grid, block>>>(bgr, gray);
cudaSafeCall(cudaDeviceSynchronize());
}
}}}
}}}
+1 -18
View File
@@ -43,24 +43,7 @@
#include <cuda_invoker.hpp>
#include <float.h>
#include <stdio.h>
namespace cv { namespace softcascade { namespace internal {
void error(const char *error_string, const char *file, const int line, const char *func);
}}}
#if defined(__GNUC__)
#define cudaSafeCall(expr) ___cudaSafeCall(expr, __FILE__, __LINE__, __func__)
#else /* defined(__CUDACC__) || defined(__MSVC__) */
#define cudaSafeCall(expr) ___cudaSafeCall(expr, __FILE__, __LINE__)
#endif
static inline void ___cudaSafeCall(cudaError_t err, const char *file, const int line, const char *func = "")
{
if (cudaSuccess != err) cv::softcascade::internal::error(cudaGetErrorString(err), file, line, func);
}
#ifndef CV_PI
#define CV_PI 3.1415926535897932384626433832795
#endif
#include "opencv2/core/cuda/common.hpp"
namespace cv { namespace softcascade { namespace cudev {
+1 -1
View File
@@ -501,7 +501,7 @@ void integral(const cv::gpu::GpuMat& src, cv::gpu::GpuMat& sum, cv::gpu::GpuMat&
else
res.copyTo(inner);
}
else {CV_Error(CV_GpuNotSupported, ": CC 3.x required.");}
else {CV_Error(cv::Error::GpuNotSupported, ": CC 3.x required.");}
}
}