generic-library/opencv
1
0
Fork 0
Code Issues Pull Requests Releases Wiki Activity

replaced GPU -> CUDA

Browse Source
This commit is contained in:
Vladislav Vinogradov
2013-07-24 13:55:18 +04:00
parent 3c34b24f1f
commit fd88654b45
172 changed files with 2434 additions and 2436 deletions
Download Patch File Download Diff File Expand all files Collapse all files

10
modules/core/include/opencv2/core/cuda.hpp
View File

@@ -41,8 +41,8 @@
//
//M*/
#ifndef __OPENCV_CORE_GPU_HPP__
#define __OPENCV_CORE_GPU_HPP__
#ifndef __OPENCV_CORE_CUDA_HPP__
#define __OPENCV_CORE_CUDA_HPP__
#ifndef __cplusplus
# error cuda.hpp header must be compiled as C++
@@ -453,7 +453,7 @@ enum FeatureSet
//! checks whether current device supports the given feature
CV_EXPORTS bool deviceSupports(FeatureSet feature_set);
//! information about what GPU archs this OpenCV GPU module was compiled for
//! information about what GPU archs this OpenCV CUDA module was compiled for
class CV_EXPORTS TargetArchs
{
public:
@@ -654,7 +654,7 @@ public:
//! checks whether device supports the given feature
bool supports(FeatureSet feature_set) const;
//! checks whether the GPU module can be run on the given device
//! checks whether the CUDA module can be run on the given device
bool isCompatible() const;
private:
@@ -675,4 +675,4 @@ template <> CV_EXPORTS void Ptr<cv::cuda::Event::Impl>::delete_obj();
#include "opencv2/core/cuda.inl.hpp"
#endif /* __OPENCV_CORE_GPU_HPP__ */
#endif /* __OPENCV_CORE_CUDA_HPP__ */

6
modules/core/include/opencv2/core/cuda.inl.hpp
View File

@@ -41,8 +41,8 @@
//
//M*/
#ifndef __OPENCV_CORE_GPUINL_HPP__
#define __OPENCV_CORE_GPUINL_HPP__
#ifndef __OPENCV_CORE_CUDAINL_HPP__
#define __OPENCV_CORE_CUDAINL_HPP__
#include "opencv2/core/cuda.hpp"
@@ -602,4 +602,4 @@ Mat::Mat(const cuda::GpuMat& m)
}
#endif // __OPENCV_CORE_GPUINL_HPP__
#endif // __OPENCV_CORE_CUDAINL_HPP__

6
modules/core/include/opencv2/core/cuda/block.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_DEVICE_BLOCK_HPP__
#define __OPENCV_GPU_DEVICE_BLOCK_HPP__
#ifndef __OPENCV_CUDA_DEVICE_BLOCK_HPP__
#define __OPENCV_CUDA_DEVICE_BLOCK_HPP__
namespace cv { namespace cuda { namespace device
{
@@ -200,4 +200,4 @@ namespace cv { namespace cuda { namespace device
};
}}}
#endif /* __OPENCV_GPU_DEVICE_BLOCK_HPP__ */
#endif /* __OPENCV_CUDA_DEVICE_BLOCK_HPP__ */

6
modules/core/include/opencv2/core/cuda/border_interpolate.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_BORDER_INTERPOLATE_HPP__
#define __OPENCV_GPU_BORDER_INTERPOLATE_HPP__
#ifndef __OPENCV_CUDA_BORDER_INTERPOLATE_HPP__
#define __OPENCV_CUDA_BORDER_INTERPOLATE_HPP__
#include "saturate_cast.hpp"
#include "vec_traits.hpp"
@@ -711,4 +711,4 @@ namespace cv { namespace cuda { namespace device
};
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_BORDER_INTERPOLATE_HPP__
#endif // __OPENCV_CUDA_BORDER_INTERPOLATE_HPP__

402
modules/core/include/opencv2/core/cuda/color.hpp
View File

@@ -40,262 +40,262 @@
//
//M*/
#ifndef __OPENCV_GPU_COLOR_HPP__
#define __OPENCV_GPU_COLOR_HPP__
#ifndef __OPENCV_CUDA_COLOR_HPP__
#define __OPENCV_CUDA_COLOR_HPP__
#include "detail/color_detail.hpp"
namespace cv { namespace cuda { namespace device
{
// All OPENCV_GPU_IMPLEMENT_*_TRAITS(ColorSpace1_to_ColorSpace2, ...) macros implements
// All OPENCV_CUDA_IMPLEMENT_*_TRAITS(ColorSpace1_to_ColorSpace2, ...) macros implements
// template <typename T> class ColorSpace1_to_ColorSpace2_traits
// {
// typedef ... functor_type;
// static __host__ __device__ functor_type create_functor();
// };
OPENCV_GPU_IMPLEMENT_RGB2RGB_TRAITS(bgr_to_rgb, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2RGB_TRAITS(bgr_to_bgra, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_RGB2RGB_TRAITS(bgr_to_rgba, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2RGB_TRAITS(bgra_to_bgr, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2RGB_TRAITS(bgra_to_rgb, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2RGB_TRAITS(bgra_to_rgba, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2RGB_TRAITS(bgr_to_rgb, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2RGB_TRAITS(bgr_to_bgra, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2RGB_TRAITS(bgr_to_rgba, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2RGB_TRAITS(bgra_to_bgr, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2RGB_TRAITS(bgra_to_rgb, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2RGB_TRAITS(bgra_to_rgba, 4, 4, 2)
#undef OPENCV_GPU_IMPLEMENT_RGB2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2RGB_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(bgr_to_bgr555, 3, 0, 5)
OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(bgr_to_bgr565, 3, 0, 6)
OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(rgb_to_bgr555, 3, 2, 5)
OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(rgb_to_bgr565, 3, 2, 6)
OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(bgra_to_bgr555, 4, 0, 5)
OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(bgra_to_bgr565, 4, 0, 6)
OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(rgba_to_bgr555, 4, 2, 5)
OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(rgba_to_bgr565, 4, 2, 6)
OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(bgr_to_bgr555, 3, 0, 5)
OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(bgr_to_bgr565, 3, 0, 6)
OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(rgb_to_bgr555, 3, 2, 5)
OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(rgb_to_bgr565, 3, 2, 6)
OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(bgra_to_bgr555, 4, 0, 5)
OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(bgra_to_bgr565, 4, 0, 6)
OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(rgba_to_bgr555, 4, 2, 5)
OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(rgba_to_bgr565, 4, 2, 6)
#undef OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS
OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(bgr555_to_rgb, 3, 2, 5)
OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(bgr565_to_rgb, 3, 2, 6)
OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(bgr555_to_bgr, 3, 0, 5)
OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(bgr565_to_bgr, 3, 0, 6)
OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(bgr555_to_rgba, 4, 2, 5)
OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(bgr565_to_rgba, 4, 2, 6)
OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(bgr555_to_bgra, 4, 0, 5)
OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(bgr565_to_bgra, 4, 0, 6)
OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(bgr555_to_rgb, 3, 2, 5)
OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(bgr565_to_rgb, 3, 2, 6)
OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(bgr555_to_bgr, 3, 0, 5)
OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(bgr565_to_bgr, 3, 0, 6)
OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(bgr555_to_rgba, 4, 2, 5)
OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(bgr565_to_rgba, 4, 2, 6)
OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(bgr555_to_bgra, 4, 0, 5)
OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(bgr565_to_bgra, 4, 0, 6)
#undef OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS
OPENCV_GPU_IMPLEMENT_GRAY2RGB_TRAITS(gray_to_bgr, 3)
OPENCV_GPU_IMPLEMENT_GRAY2RGB_TRAITS(gray_to_bgra, 4)
OPENCV_CUDA_IMPLEMENT_GRAY2RGB_TRAITS(gray_to_bgr, 3)
OPENCV_CUDA_IMPLEMENT_GRAY2RGB_TRAITS(gray_to_bgra, 4)
#undef OPENCV_GPU_IMPLEMENT_GRAY2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_GRAY2RGB_TRAITS
OPENCV_GPU_IMPLEMENT_GRAY2RGB5x5_TRAITS(gray_to_bgr555, 5)
OPENCV_GPU_IMPLEMENT_GRAY2RGB5x5_TRAITS(gray_to_bgr565, 6)
OPENCV_CUDA_IMPLEMENT_GRAY2RGB5x5_TRAITS(gray_to_bgr555, 5)
OPENCV_CUDA_IMPLEMENT_GRAY2RGB5x5_TRAITS(gray_to_bgr565, 6)
#undef OPENCV_GPU_IMPLEMENT_GRAY2RGB5x5_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_GRAY2RGB5x5_TRAITS
OPENCV_GPU_IMPLEMENT_RGB5x52GRAY_TRAITS(bgr555_to_gray, 5)
OPENCV_GPU_IMPLEMENT_RGB5x52GRAY_TRAITS(bgr565_to_gray, 6)
OPENCV_CUDA_IMPLEMENT_RGB5x52GRAY_TRAITS(bgr555_to_gray, 5)
OPENCV_CUDA_IMPLEMENT_RGB5x52GRAY_TRAITS(bgr565_to_gray, 6)
#undef OPENCV_GPU_IMPLEMENT_RGB5x52GRAY_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB5x52GRAY_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2GRAY_TRAITS(rgb_to_gray, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2GRAY_TRAITS(bgr_to_gray, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2GRAY_TRAITS(rgba_to_gray, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2GRAY_TRAITS(bgra_to_gray, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2GRAY_TRAITS(rgb_to_gray, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2GRAY_TRAITS(bgr_to_gray, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2GRAY_TRAITS(rgba_to_gray, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2GRAY_TRAITS(bgra_to_gray, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_RGB2GRAY_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2GRAY_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(rgb_to_yuv, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(rgba_to_yuv, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(rgb_to_yuv4, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(rgba_to_yuv4, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(bgr_to_yuv, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(bgra_to_yuv, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(bgr_to_yuv4, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(bgra_to_yuv4, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(rgb_to_yuv, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(rgba_to_yuv, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(rgb_to_yuv4, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(rgba_to_yuv4, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(bgr_to_yuv, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(bgra_to_yuv, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(bgr_to_yuv4, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(bgra_to_yuv4, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS
OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(yuv_to_rgb, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(yuv_to_rgba, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(yuv4_to_rgb, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(yuv4_to_rgba, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(yuv_to_bgr, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(yuv_to_bgra, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(yuv4_to_bgr, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(yuv4_to_bgra, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(yuv_to_rgb, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(yuv_to_rgba, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(yuv4_to_rgb, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(yuv4_to_rgba, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(yuv_to_bgr, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(yuv_to_bgra, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(yuv4_to_bgr, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(yuv4_to_bgra, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(rgb_to_YCrCb, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(rgba_to_YCrCb, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(rgb_to_YCrCb4, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(rgba_to_YCrCb4, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(bgr_to_YCrCb, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(bgra_to_YCrCb, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(bgr_to_YCrCb4, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(bgra_to_YCrCb4, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(rgb_to_YCrCb, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(rgba_to_YCrCb, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(rgb_to_YCrCb4, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(rgba_to_YCrCb4, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(bgr_to_YCrCb, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(bgra_to_YCrCb, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(bgr_to_YCrCb4, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(bgra_to_YCrCb4, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS
OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb_to_rgb, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb_to_rgba, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb4_to_rgb, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb4_to_rgba, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb_to_bgr, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb_to_bgra, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb4_to_bgr, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb4_to_bgra, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb_to_rgb, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb_to_rgba, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb4_to_rgb, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb4_to_rgba, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb_to_bgr, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb_to_bgra, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb4_to_bgr, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(YCrCb4_to_bgra, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(rgb_to_xyz, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(rgba_to_xyz, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(rgb_to_xyz4, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(rgba_to_xyz4, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(bgr_to_xyz, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(bgra_to_xyz, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(bgr_to_xyz4, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(bgra_to_xyz4, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(rgb_to_xyz, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(rgba_to_xyz, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(rgb_to_xyz4, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(rgba_to_xyz4, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(bgr_to_xyz, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(bgra_to_xyz, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(bgr_to_xyz4, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(bgra_to_xyz4, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS
OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(xyz_to_rgb, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(xyz4_to_rgb, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(xyz_to_rgba, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(xyz4_to_rgba, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(xyz_to_bgr, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(xyz4_to_bgr, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(xyz_to_bgra, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(xyz4_to_bgra, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(xyz_to_rgb, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(xyz4_to_rgb, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(xyz_to_rgba, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(xyz4_to_rgba, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(xyz_to_bgr, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(xyz4_to_bgr, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(xyz_to_bgra, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(xyz4_to_bgra, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(rgb_to_hsv, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(rgba_to_hsv, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(rgb_to_hsv4, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(rgba_to_hsv4, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(bgr_to_hsv, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(bgra_to_hsv, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(bgr_to_hsv4, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(bgra_to_hsv4, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(rgb_to_hsv, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(rgba_to_hsv, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(rgb_to_hsv4, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(rgba_to_hsv4, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(bgr_to_hsv, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(bgra_to_hsv, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(bgr_to_hsv4, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(bgra_to_hsv4, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS
OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(hsv_to_rgb, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(hsv_to_rgba, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(hsv4_to_rgb, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(hsv4_to_rgba, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(hsv_to_bgr, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(hsv_to_bgra, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(hsv4_to_bgr, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(hsv4_to_bgra, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(hsv_to_rgb, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(hsv_to_rgba, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(hsv4_to_rgb, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(hsv4_to_rgba, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(hsv_to_bgr, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(hsv_to_bgra, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(hsv4_to_bgr, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(hsv4_to_bgra, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(rgb_to_hls, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(rgba_to_hls, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(rgb_to_hls4, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(rgba_to_hls4, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(bgr_to_hls, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(bgra_to_hls, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(bgr_to_hls4, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(bgra_to_hls4, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(rgb_to_hls, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(rgba_to_hls, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(rgb_to_hls4, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(rgba_to_hls4, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(bgr_to_hls, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(bgra_to_hls, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(bgr_to_hls4, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(bgra_to_hls4, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls_to_rgb, 3, 3, 2)
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls_to_rgba, 3, 4, 2)
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_rgb, 4, 3, 2)
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_rgba, 4, 4, 2)
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls_to_bgr, 3, 3, 0)
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls_to_bgra, 3, 4, 0)
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_bgr, 4, 3, 0)
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_bgra, 4, 4, 0)
OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(hls_to_rgb, 3, 3, 2)
OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(hls_to_rgba, 3, 4, 2)
OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_rgb, 4, 3, 2)
OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_rgba, 4, 4, 2)
OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(hls_to_bgr, 3, 3, 0)
OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(hls_to_bgra, 3, 4, 0)
OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_bgr, 4, 3, 0)
OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_bgra, 4, 4, 0)
#undef OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(rgb_to_lab, 3, 3, true, 2)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(rgba_to_lab, 4, 3, true, 2)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(rgb_to_lab4, 3, 4, true, 2)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(rgba_to_lab4, 4, 4, true, 2)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(bgr_to_lab, 3, 3, true, 0)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(bgra_to_lab, 4, 3, true, 0)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(bgr_to_lab4, 3, 4, true, 0)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(bgra_to_lab4, 4, 4, true, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(rgb_to_lab, 3, 3, true, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(rgba_to_lab, 4, 3, true, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(rgb_to_lab4, 3, 4, true, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(rgba_to_lab4, 4, 4, true, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(bgr_to_lab, 3, 3, true, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(bgra_to_lab, 4, 3, true, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(bgr_to_lab4, 3, 4, true, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(bgra_to_lab4, 4, 4, true, 0)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(lrgb_to_lab, 3, 3, false, 2)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(lrgba_to_lab, 4, 3, false, 2)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(lrgb_to_lab4, 3, 4, false, 2)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(lrgba_to_lab4, 4, 4, false, 2)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(lbgr_to_lab, 3, 3, false, 0)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(lbgra_to_lab, 4, 3, false, 0)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(lbgr_to_lab4, 3, 4, false, 0)
OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(lbgra_to_lab4, 4, 4, false, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(lrgb_to_lab, 3, 3, false, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(lrgba_to_lab, 4, 3, false, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(lrgb_to_lab4, 3, 4, false, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(lrgba_to_lab4, 4, 4, false, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(lbgr_to_lab, 3, 3, false, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(lbgra_to_lab, 4, 3, false, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(lbgr_to_lab4, 3, 4, false, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(lbgra_to_lab4, 4, 4, false, 0)
#undef OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab_to_rgb, 3, 3, true, 2)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_rgb, 4, 3, true, 2)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab_to_rgba, 3, 4, true, 2)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_rgba, 4, 4, true, 2)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab_to_bgr, 3, 3, true, 0)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_bgr, 4, 3, true, 0)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab_to_bgra, 3, 4, true, 0)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_bgra, 4, 4, true, 0)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab_to_rgb, 3, 3, true, 2)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_rgb, 4, 3, true, 2)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab_to_rgba, 3, 4, true, 2)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_rgba, 4, 4, true, 2)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab_to_bgr, 3, 3, true, 0)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_bgr, 4, 3, true, 0)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab_to_bgra, 3, 4, true, 0)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_bgra, 4, 4, true, 0)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab_to_lrgb, 3, 3, false, 2)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_lrgb, 4, 3, false, 2)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab_to_lrgba, 3, 4, false, 2)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_lrgba, 4, 4, false, 2)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab_to_lbgr, 3, 3, false, 0)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_lbgr, 4, 3, false, 0)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab_to_lbgra, 3, 4, false, 0)
OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_lbgra, 4, 4, false, 0)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab_to_lrgb, 3, 3, false, 2)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_lrgb, 4, 3, false, 2)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab_to_lrgba, 3, 4, false, 2)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_lrgba, 4, 4, false, 2)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab_to_lbgr, 3, 3, false, 0)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_lbgr, 4, 3, false, 0)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab_to_lbgra, 3, 4, false, 0)
OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(lab4_to_lbgra, 4, 4, false, 0)
#undef OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(rgb_to_luv, 3, 3, true, 2)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(rgba_to_luv, 4, 3, true, 2)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(rgb_to_luv4, 3, 4, true, 2)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(rgba_to_luv4, 4, 4, true, 2)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(bgr_to_luv, 3, 3, true, 0)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(bgra_to_luv, 4, 3, true, 0)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(bgr_to_luv4, 3, 4, true, 0)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(bgra_to_luv4, 4, 4, true, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(rgb_to_luv, 3, 3, true, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(rgba_to_luv, 4, 3, true, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(rgb_to_luv4, 3, 4, true, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(rgba_to_luv4, 4, 4, true, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(bgr_to_luv, 3, 3, true, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(bgra_to_luv, 4, 3, true, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(bgr_to_luv4, 3, 4, true, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(bgra_to_luv4, 4, 4, true, 0)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(lrgb_to_luv, 3, 3, false, 2)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(lrgba_to_luv, 4, 3, false, 2)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(lrgb_to_luv4, 3, 4, false, 2)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(lrgba_to_luv4, 4, 4, false, 2)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(lbgr_to_luv, 3, 3, false, 0)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(lbgra_to_luv, 4, 3, false, 0)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(lbgr_to_luv4, 3, 4, false, 0)
OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(lbgra_to_luv4, 4, 4, false, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(lrgb_to_luv, 3, 3, false, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(lrgba_to_luv, 4, 3, false, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(lrgb_to_luv4, 3, 4, false, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(lrgba_to_luv4, 4, 4, false, 2)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(lbgr_to_luv, 3, 3, false, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(lbgra_to_luv, 4, 3, false, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(lbgr_to_luv4, 3, 4, false, 0)
OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(lbgra_to_luv4, 4, 4, false, 0)
#undef OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv_to_rgb, 3, 3, true, 2)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_rgb, 4, 3, true, 2)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv_to_rgba, 3, 4, true, 2)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_rgba, 4, 4, true, 2)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv_to_bgr, 3, 3, true, 0)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_bgr, 4, 3, true, 0)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv_to_bgra, 3, 4, true, 0)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_bgra, 4, 4, true, 0)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv_to_rgb, 3, 3, true, 2)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_rgb, 4, 3, true, 2)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv_to_rgba, 3, 4, true, 2)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_rgba, 4, 4, true, 2)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv_to_bgr, 3, 3, true, 0)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_bgr, 4, 3, true, 0)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv_to_bgra, 3, 4, true, 0)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_bgra, 4, 4, true, 0)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv_to_lrgb, 3, 3, false, 2)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_lrgb, 4, 3, false, 2)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv_to_lrgba, 3, 4, false, 2)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_lrgba, 4, 4, false, 2)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv_to_lbgr, 3, 3, false, 0)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_lbgr, 4, 3, false, 0)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv_to_lbgra, 3, 4, false, 0)
OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_lbgra, 4, 4, false, 0)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv_to_lrgb, 3, 3, false, 2)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_lrgb, 4, 3, false, 2)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv_to_lrgba, 3, 4, false, 2)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_lrgba, 4, 4, false, 2)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv_to_lbgr, 3, 3, false, 0)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_lbgr, 4, 3, false, 0)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv_to_lbgra, 3, 4, false, 0)
OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(luv4_to_lbgra, 4, 4, false, 0)
#undef OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_BORDER_INTERPOLATE_HPP__
#endif // __OPENCV_CUDA_BORDER_INTERPOLATE_HPP__

6
modules/core/include/opencv2/core/cuda/common.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_COMMON_HPP__
#define __OPENCV_GPU_COMMON_HPP__
#ifndef __OPENCV_CUDA_COMMON_HPP__
#define __OPENCV_CUDA_COMMON_HPP__
#include <cuda_runtime.h>
#include "opencv2/core/cuda_types.hpp"
@@ -104,4 +104,4 @@ namespace cv { namespace cuda
#endif // __OPENCV_GPU_COMMON_HPP__
#endif // __OPENCV_CUDA_COMMON_HPP__

42
modules/core/include/opencv2/core/cuda/datamov_utils.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_DATAMOV_UTILS_HPP__
#define __OPENCV_GPU_DATAMOV_UTILS_HPP__
#ifndef __OPENCV_CUDA_DATAMOV_UTILS_HPP__
#define __OPENCV_CUDA_DATAMOV_UTILS_HPP__
#include "common.hpp"
@@ -59,47 +59,47 @@ namespace cv { namespace cuda { namespace device
#if defined(_WIN64) || defined(__LP64__)
// 64-bit register modifier for inlined asm
#define OPENCV_GPU_ASM_PTR "l"
#define OPENCV_CUDA_ASM_PTR "l"
#else
// 32-bit register modifier for inlined asm
#define OPENCV_GPU_ASM_PTR "r"
#define OPENCV_CUDA_ASM_PTR "r"
#endif
template<class T> struct ForceGlob;
#define OPENCV_GPU_DEFINE_FORCE_GLOB(base_type, ptx_type, reg_mod) \
#define OPENCV_CUDA_DEFINE_FORCE_GLOB(base_type, ptx_type, reg_mod) \
template <> struct ForceGlob<base_type> \
{ \
__device__ __forceinline__ static void Load(const base_type* ptr, int offset, base_type& val) \
{ \
asm("ld.global."#ptx_type" %0, [%1];" : "="#reg_mod(val) : OPENCV_GPU_ASM_PTR(ptr + offset)); \
asm("ld.global."#ptx_type" %0, [%1];" : "="#reg_mod(val) : OPENCV_CUDA_ASM_PTR(ptr + offset)); \
} \
};
#define OPENCV_GPU_DEFINE_FORCE_GLOB_B(base_type, ptx_type) \
#define OPENCV_CUDA_DEFINE_FORCE_GLOB_B(base_type, ptx_type) \
template <> struct ForceGlob<base_type> \
{ \
__device__ __forceinline__ static void Load(const base_type* ptr, int offset, base_type& val) \
{ \
asm("ld.global."#ptx_type" %0, [%1];" : "=r"(*reinterpret_cast<uint*>(&val)) : OPENCV_GPU_ASM_PTR(ptr + offset)); \
asm("ld.global."#ptx_type" %0, [%1];" : "=r"(*reinterpret_cast<uint*>(&val)) : OPENCV_CUDA_ASM_PTR(ptr + offset)); \
} \
};
OPENCV_GPU_DEFINE_FORCE_GLOB_B(uchar, u8)
OPENCV_GPU_DEFINE_FORCE_GLOB_B(schar, s8)
OPENCV_GPU_DEFINE_FORCE_GLOB_B(char, b8)
OPENCV_GPU_DEFINE_FORCE_GLOB (ushort, u16, h)
OPENCV_GPU_DEFINE_FORCE_GLOB (short, s16, h)
OPENCV_GPU_DEFINE_FORCE_GLOB (uint, u32, r)
OPENCV_GPU_DEFINE_FORCE_GLOB (int, s32, r)
OPENCV_GPU_DEFINE_FORCE_GLOB (float, f32, f)
OPENCV_GPU_DEFINE_FORCE_GLOB (double, f64, d)
OPENCV_CUDA_DEFINE_FORCE_GLOB_B(uchar, u8)
OPENCV_CUDA_DEFINE_FORCE_GLOB_B(schar, s8)
OPENCV_CUDA_DEFINE_FORCE_GLOB_B(char, b8)
OPENCV_CUDA_DEFINE_FORCE_GLOB (ushort, u16, h)
OPENCV_CUDA_DEFINE_FORCE_GLOB (short, s16, h)
OPENCV_CUDA_DEFINE_FORCE_GLOB (uint, u32, r)
OPENCV_CUDA_DEFINE_FORCE_GLOB (int, s32, r)
OPENCV_CUDA_DEFINE_FORCE_GLOB (float, f32, f)
OPENCV_CUDA_DEFINE_FORCE_GLOB (double, f64, d)
#undef OPENCV_GPU_DEFINE_FORCE_GLOB
#undef OPENCV_GPU_DEFINE_FORCE_GLOB_B
#undef OPENCV_GPU_ASM_PTR
#undef OPENCV_CUDA_DEFINE_FORCE_GLOB
#undef OPENCV_CUDA_DEFINE_FORCE_GLOB_B
#undef OPENCV_CUDA_ASM_PTR
#endif // __CUDA_ARCH__ >= 200
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_DATAMOV_UTILS_HPP__
#endif // __OPENCV_CUDA_DATAMOV_UTILS_HPP__

48
modules/core/include/opencv2/core/cuda/detail/color_detail.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_COLOR_DETAIL_HPP__
#define __OPENCV_GPU_COLOR_DETAIL_HPP__
#ifndef __OPENCV_CUDA_COLOR_DETAIL_HPP__
#define __OPENCV_CUDA_COLOR_DETAIL_HPP__
#include "../common.hpp"
#include "../vec_traits.hpp"
@@ -143,7 +143,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2RGB_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2RGB_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2RGB<T, scn, dcn, bidx> functor_type; \
@@ -216,7 +216,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2RGB5x5_TRAITS(name, scn, bidx, green_bits) \
#define OPENCV_CUDA_IMPLEMENT_RGB2RGB5x5_TRAITS(name, scn, bidx, green_bits) \
struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2RGB5x5<scn, bidx, green_bits> functor_type; \
@@ -297,7 +297,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB5x52RGB_TRAITS(name, dcn, bidx, green_bits) \
#define OPENCV_CUDA_IMPLEMENT_RGB5x52RGB_TRAITS(name, dcn, bidx, green_bits) \
struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB5x52RGB<dcn, bidx, green_bits> functor_type; \
@@ -343,7 +343,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_GRAY2RGB_TRAITS(name, dcn) \
#define OPENCV_CUDA_IMPLEMENT_GRAY2RGB_TRAITS(name, dcn) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::Gray2RGB<T, dcn> functor_type; \
@@ -385,7 +385,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_GRAY2RGB5x5_TRAITS(name, green_bits) \
#define OPENCV_CUDA_IMPLEMENT_GRAY2RGB5x5_TRAITS(name, green_bits) \
struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::Gray2RGB5x5<green_bits> functor_type; \
@@ -427,7 +427,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB5x52GRAY_TRAITS(name, green_bits) \
#define OPENCV_CUDA_IMPLEMENT_RGB5x52GRAY_TRAITS(name, green_bits) \
struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB5x52Gray<green_bits> functor_type; \
@@ -478,7 +478,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2GRAY_TRAITS(name, scn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2GRAY_TRAITS(name, scn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2Gray<T, scn, bidx> functor_type; \
@@ -529,7 +529,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2YUV_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2YUV_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2YUV<T, scn, dcn, bidx> functor_type; \
@@ -617,7 +617,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_YUV2RGB_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_YUV2RGB_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::YUV2RGB<T, scn, dcn, bidx> functor_type; \
@@ -696,7 +696,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2YCrCb_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2YCrCb_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2YCrCb<T, scn, dcn, bidx> functor_type; \
@@ -775,7 +775,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_YCrCb2RGB_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_YCrCb2RGB_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::YCrCb2RGB<T, scn, dcn, bidx> functor_type; \
@@ -851,7 +851,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2XYZ_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2XYZ_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2XYZ<T, scn, dcn, bidx> functor_type; \
@@ -926,7 +926,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_XYZ2RGB_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_XYZ2RGB_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::XYZ2RGB<T, scn, dcn, bidx> functor_type; \
@@ -1064,7 +1064,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2HSV_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2HSV_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2HSV<T, scn, dcn, bidx, 180> functor_type; \
@@ -1204,7 +1204,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_HSV2RGB_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_HSV2RGB_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::HSV2RGB<T, scn, dcn, bidx, 180> functor_type; \
@@ -1337,7 +1337,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2HLS_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2HLS_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2HLS<T, scn, dcn, bidx, 180> functor_type; \
@@ -1477,7 +1477,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(name, scn, dcn, bidx) \
#define OPENCV_CUDA_IMPLEMENT_HLS2RGB_TRAITS(name, scn, dcn, bidx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::HLS2RGB<T, scn, dcn, bidx, 180> functor_type; \
@@ -1646,7 +1646,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2Lab_TRAITS(name, scn, dcn, srgb, blueIdx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2Lab_TRAITS(name, scn, dcn, srgb, blueIdx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2Lab<T, scn, dcn, srgb, blueIdx> functor_type; \
@@ -1759,7 +1759,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_Lab2RGB_TRAITS(name, scn, dcn, srgb, blueIdx) \
#define OPENCV_CUDA_IMPLEMENT_Lab2RGB_TRAITS(name, scn, dcn, srgb, blueIdx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::Lab2RGB<T, scn, dcn, srgb, blueIdx> functor_type; \
@@ -1858,7 +1858,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_RGB2Luv_TRAITS(name, scn, dcn, srgb, blueIdx) \
#define OPENCV_CUDA_IMPLEMENT_RGB2Luv_TRAITS(name, scn, dcn, srgb, blueIdx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::RGB2Luv<T, scn, dcn, srgb, blueIdx> functor_type; \
@@ -1959,7 +1959,7 @@ namespace cv { namespace cuda { namespace device
};
}
#define OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS(name, scn, dcn, srgb, blueIdx) \
#define OPENCV_CUDA_IMPLEMENT_Luv2RGB_TRAITS(name, scn, dcn, srgb, blueIdx) \
template <typename T> struct name ## _traits \
{ \
typedef ::cv::cuda::device::color_detail::Luv2RGB<T, scn, dcn, srgb, blueIdx> functor_type; \
@@ -1973,4 +1973,4 @@ namespace cv { namespace cuda { namespace device
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_COLOR_DETAIL_HPP__
#endif // __OPENCV_CUDA_COLOR_DETAIL_HPP__

6
modules/core/include/opencv2/core/cuda/detail/reduce.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_REDUCE_DETAIL_HPP__
#define __OPENCV_GPU_REDUCE_DETAIL_HPP__
#ifndef __OPENCV_CUDA_REDUCE_DETAIL_HPP__
#define __OPENCV_CUDA_REDUCE_DETAIL_HPP__
#include <thrust/tuple.h>
#include "../warp.hpp"
@@ -358,4 +358,4 @@ namespace cv { namespace cuda { namespace device
}
}}}
#endif // __OPENCV_GPU_REDUCE_DETAIL_HPP__
#endif // __OPENCV_CUDA_REDUCE_DETAIL_HPP__

6
modules/core/include/opencv2/core/cuda/detail/reduce_key_val.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_PRED_VAL_REDUCE_DETAIL_HPP__
#define __OPENCV_GPU_PRED_VAL_REDUCE_DETAIL_HPP__
#ifndef __OPENCV_CUDA_PRED_VAL_REDUCE_DETAIL_HPP__
#define __OPENCV_CUDA_PRED_VAL_REDUCE_DETAIL_HPP__
#include <thrust/tuple.h>
#include "../warp.hpp"
@@ -495,4 +495,4 @@ namespace cv { namespace cuda { namespace device
}
}}}
#endif // __OPENCV_GPU_PRED_VAL_REDUCE_DETAIL_HPP__
#endif // __OPENCV_CUDA_PRED_VAL_REDUCE_DETAIL_HPP__

6
modules/core/include/opencv2/core/cuda/detail/transform_detail.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_TRANSFORM_DETAIL_HPP__
#define __OPENCV_GPU_TRANSFORM_DETAIL_HPP__
#ifndef __OPENCV_CUDA_TRANSFORM_DETAIL_HPP__
#define __OPENCV_CUDA_TRANSFORM_DETAIL_HPP__
#include "../common.hpp"
#include "../vec_traits.hpp"
@@ -392,4 +392,4 @@ namespace cv { namespace cuda { namespace device
} // namespace transform_detail
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_TRANSFORM_DETAIL_HPP__
#endif // __OPENCV_CUDA_TRANSFORM_DETAIL_HPP__

6
modules/core/include/opencv2/core/cuda/detail/type_traits_detail.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_TYPE_TRAITS_DETAIL_HPP__
#define __OPENCV_GPU_TYPE_TRAITS_DETAIL_HPP__
#ifndef __OPENCV_CUDA_TYPE_TRAITS_DETAIL_HPP__
#define __OPENCV_CUDA_TYPE_TRAITS_DETAIL_HPP__
#include "../common.hpp"
#include "../vec_traits.hpp"
@@ -184,4 +184,4 @@ namespace cv { namespace cuda { namespace device
} // namespace type_traits_detail
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_TYPE_TRAITS_DETAIL_HPP__
#endif // __OPENCV_CUDA_TYPE_TRAITS_DETAIL_HPP__

6
modules/core/include/opencv2/core/cuda/detail/vec_distance_detail.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_VEC_DISTANCE_DETAIL_HPP__
#define __OPENCV_GPU_VEC_DISTANCE_DETAIL_HPP__
#ifndef __OPENCV_CUDA_VEC_DISTANCE_DETAIL_HPP__
#define __OPENCV_CUDA_VEC_DISTANCE_DETAIL_HPP__
#include "../datamov_utils.hpp"
@@ -114,4 +114,4 @@ namespace cv { namespace cuda { namespace device
} // namespace vec_distance_detail
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_VEC_DISTANCE_DETAIL_HPP__
#endif // __OPENCV_CUDA_VEC_DISTANCE_DETAIL_HPP__

6
modules/core/include/opencv2/core/cuda/dynamic_smem.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_DYNAMIC_SMEM_HPP__
#define __OPENCV_GPU_DYNAMIC_SMEM_HPP__
#ifndef __OPENCV_CUDA_DYNAMIC_SMEM_HPP__
#define __OPENCV_CUDA_DYNAMIC_SMEM_HPP__
namespace cv { namespace cuda { namespace device
{
@@ -77,4 +77,4 @@ namespace cv { namespace cuda { namespace device
};
}}}
#endif // __OPENCV_GPU_DYNAMIC_SMEM_HPP__
#endif // __OPENCV_CUDA_DYNAMIC_SMEM_HPP__

6
modules/core/include/opencv2/core/cuda/emulation.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef OPENCV_GPU_EMULATION_HPP_
#define OPENCV_GPU_EMULATION_HPP_
#ifndef OPENCV_CUDA_EMULATION_HPP_
#define OPENCV_CUDA_EMULATION_HPP_
#include "common.hpp"
#include "warp_reduce.hpp"
@@ -258,4 +258,4 @@ namespace cv { namespace cuda { namespace device
}; //struct Emulation
}}} // namespace cv { namespace cuda { namespace cudev
#endif /* OPENCV_GPU_EMULATION_HPP_ */
#endif /* OPENCV_CUDA_EMULATION_HPP_ */

6
modules/core/include/opencv2/core/cuda/filters.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_FILTERS_HPP__
#define __OPENCV_GPU_FILTERS_HPP__
#ifndef __OPENCV_CUDA_FILTERS_HPP__
#define __OPENCV_CUDA_FILTERS_HPP__
#include "saturate_cast.hpp"
#include "vec_traits.hpp"
@@ -275,4 +275,4 @@ namespace cv { namespace cuda { namespace device
};
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_FILTERS_HPP__
#endif // __OPENCV_CUDA_FILTERS_HPP__

6
modules/core/include/opencv2/core/cuda/funcattrib.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_DEVICE_FUNCATTRIB_HPP_
#define __OPENCV_GPU_DEVICE_FUNCATTRIB_HPP_
#ifndef __OPENCV_CUDA_DEVICE_FUNCATTRIB_HPP_
#define __OPENCV_CUDA_DEVICE_FUNCATTRIB_HPP_
#include <cstdio>
@@ -68,4 +68,4 @@ namespace cv { namespace cuda { namespace device
}
}}} // namespace cv { namespace cuda { namespace cudev
#endif /* __OPENCV_GPU_DEVICE_FUNCATTRIB_HPP_ */
#endif /* __OPENCV_CUDA_DEVICE_FUNCATTRIB_HPP_ */

102
modules/core/include/opencv2/core/cuda/functional.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_FUNCTIONAL_HPP__
#define __OPENCV_GPU_FUNCTIONAL_HPP__
#ifndef __OPENCV_CUDA_FUNCTIONAL_HPP__
#define __OPENCV_CUDA_FUNCTIONAL_HPP__
#include <functional>
#include "saturate_cast.hpp"
@@ -298,7 +298,7 @@ namespace cv { namespace cuda { namespace device
// Min/Max Operations
#define OPENCV_GPU_IMPLEMENT_MINMAX(name, type, op) \
#define OPENCV_CUDA_IMPLEMENT_MINMAX(name, type, op) \
template <> struct name<type> : binary_function<type, type, type> \
{ \
__device__ __forceinline__ type operator()(type lhs, type rhs) const {return op(lhs, rhs);} \
@@ -316,15 +316,15 @@ namespace cv { namespace cuda { namespace device
__host__ __device__ __forceinline__ maximum(const maximum&) {}
};
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, uchar, ::max)
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, schar, ::max)
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, char, ::max)
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, ushort, ::max)
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, short, ::max)
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, int, ::max)
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, uint, ::max)
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, float, ::fmax)
OPENCV_GPU_IMPLEMENT_MINMAX(maximum, double, ::fmax)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, uchar, ::max)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, schar, ::max)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, char, ::max)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, ushort, ::max)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, short, ::max)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, int, ::max)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, uint, ::max)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, float, ::fmax)
OPENCV_CUDA_IMPLEMENT_MINMAX(maximum, double, ::fmax)
template <typename T> struct minimum : binary_function<T, T, T>
{
@@ -336,17 +336,17 @@ namespace cv { namespace cuda { namespace device
__host__ __device__ __forceinline__ minimum(const minimum&) {}
};
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, uchar, ::min)
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, schar, ::min)
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, char, ::min)
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, ushort, ::min)
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, short, ::min)
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, int, ::min)
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, uint, ::min)
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, float, ::fmin)
OPENCV_GPU_IMPLEMENT_MINMAX(minimum, double, ::fmin)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, uchar, ::min)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, schar, ::min)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, char, ::min)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, ushort, ::min)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, short, ::min)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, int, ::min)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, uint, ::min)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, float, ::fmin)
OPENCV_CUDA_IMPLEMENT_MINMAX(minimum, double, ::fmin)
#undef OPENCV_GPU_IMPLEMENT_MINMAX
#undef OPENCV_CUDA_IMPLEMENT_MINMAX
// Math functions
@@ -451,7 +451,7 @@ namespace cv { namespace cuda { namespace device
__host__ __device__ __forceinline__ abs_func(const abs_func&) {}
};
#define OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(name, func) \
#define OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(name, func) \
template <typename T> struct name ## _func : unary_function<T, float> \
{ \
__device__ __forceinline__ float operator ()(typename TypeTraits<T>::ParameterType v) const \
@@ -471,7 +471,7 @@ namespace cv { namespace cuda { namespace device
__host__ __device__ __forceinline__ name ## _func(const name ## _func&) {} \
};
#define OPENCV_GPU_IMPLEMENT_BIN_FUNCTOR(name, func) \
#define OPENCV_CUDA_IMPLEMENT_BIN_FUNCTOR(name, func) \
template <typename T> struct name ## _func : binary_function<T, T, float> \
{ \
__device__ __forceinline__ float operator ()(typename TypeTraits<T>::ParameterType v1, typename TypeTraits<T>::ParameterType v2) const \
@@ -491,33 +491,33 @@ namespace cv { namespace cuda { namespace device
__host__ __device__ __forceinline__ name ## _func(const name ## _func&) {} \
};
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(sqrt, ::sqrt)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(exp, ::exp)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(exp2, ::exp2)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(exp10, ::exp10)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(log, ::log)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(log2, ::log2)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(log10, ::log10)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(sin, ::sin)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(cos, ::cos)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(tan, ::tan)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(asin, ::asin)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(acos, ::acos)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(atan, ::atan)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(sinh, ::sinh)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(cosh, ::cosh)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(tanh, ::tanh)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(asinh, ::asinh)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(acosh, ::acosh)
OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(atanh, ::atanh)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(sqrt, ::sqrt)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(exp, ::exp)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(exp2, ::exp2)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(exp10, ::exp10)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(log, ::log)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(log2, ::log2)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(log10, ::log10)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(sin, ::sin)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(cos, ::cos)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(tan, ::tan)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(asin, ::asin)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(acos, ::acos)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(atan, ::atan)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(sinh, ::sinh)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(cosh, ::cosh)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(tanh, ::tanh)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(asinh, ::asinh)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(acosh, ::acosh)
OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR(atanh, ::atanh)
OPENCV_GPU_IMPLEMENT_BIN_FUNCTOR(hypot, ::hypot)
OPENCV_GPU_IMPLEMENT_BIN_FUNCTOR(atan2, ::atan2)
OPENCV_GPU_IMPLEMENT_BIN_FUNCTOR(pow, ::pow)
OPENCV_CUDA_IMPLEMENT_BIN_FUNCTOR(hypot, ::hypot)
OPENCV_CUDA_IMPLEMENT_BIN_FUNCTOR(atan2, ::atan2)
OPENCV_CUDA_IMPLEMENT_BIN_FUNCTOR(pow, ::pow)
#undef OPENCV_GPU_IMPLEMENT_UN_FUNCTOR
#undef OPENCV_GPU_IMPLEMENT_UN_FUNCTOR_NO_DOUBLE
#undef OPENCV_GPU_IMPLEMENT_BIN_FUNCTOR
#undef OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR
#undef OPENCV_CUDA_IMPLEMENT_UN_FUNCTOR_NO_DOUBLE
#undef OPENCV_CUDA_IMPLEMENT_BIN_FUNCTOR
template<typename T> struct hypot_sqr_func : binary_function<T, T, float>
{
@@ -782,8 +782,8 @@ namespace cv { namespace cuda { namespace device
template <typename Func> struct TransformFunctorTraits : DefaultTransformFunctorTraits<Func> {};
#define OPENCV_GPU_TRANSFORM_FUNCTOR_TRAITS(type) \
#define OPENCV_CUDA_TRANSFORM_FUNCTOR_TRAITS(type) \
template <> struct TransformFunctorTraits< type > : DefaultTransformFunctorTraits< type >
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_FUNCTIONAL_HPP__
#endif // __OPENCV_CUDA_FUNCTIONAL_HPP__

6
modules/core/include/opencv2/core/cuda/limits.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_LIMITS_GPU_HPP__
#define __OPENCV_GPU_LIMITS_GPU_HPP__
#ifndef __OPENCV_CUDA_LIMITS_HPP__
#define __OPENCV_CUDA_LIMITS_HPP__
#include <limits.h>
#include <float.h>
@@ -119,4 +119,4 @@ template <> struct numeric_limits<double>
}}} // namespace cv { namespace cuda { namespace cudev {
#endif // __OPENCV_GPU_LIMITS_GPU_HPP__
#endif // __OPENCV_CUDA_LIMITS_HPP__

6
modules/core/include/opencv2/core/cuda/reduce.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_REDUCE_HPP__
#define __OPENCV_GPU_REDUCE_HPP__
#ifndef __OPENCV_CUDA_REDUCE_HPP__
#define __OPENCV_CUDA_REDUCE_HPP__
#include <thrust/tuple.h>
#include "detail/reduce.hpp"
@@ -194,4 +194,4 @@ namespace cv { namespace cuda { namespace device
}
}}}
#endif // __OPENCV_GPU_UTILITY_HPP__
#endif // __OPENCV_CUDA_UTILITY_HPP__

6
modules/core/include/opencv2/core/cuda/saturate_cast.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_SATURATE_CAST_HPP__
#define __OPENCV_GPU_SATURATE_CAST_HPP__
#ifndef __OPENCV_CUDA_SATURATE_CAST_HPP__
#define __OPENCV_CUDA_SATURATE_CAST_HPP__
#include "common.hpp"
@@ -281,4 +281,4 @@ namespace cv { namespace cuda { namespace device
}
}}}
#endif /* __OPENCV_GPU_SATURATE_CAST_HPP__ */
#endif /* __OPENCV_CUDA_SATURATE_CAST_HPP__ */

22
modules/core/include/opencv2/core/cuda/scan.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_SCAN_HPP__
#define __OPENCV_GPU_SCAN_HPP__
#ifndef __OPENCV_CUDA_SCAN_HPP__
#define __OPENCV_CUDA_SCAN_HPP__
#include "opencv2/core/cuda/common.hpp"
#include "opencv2/core/cuda/utility.hpp"
@@ -178,7 +178,7 @@ namespace cv { namespace cuda { namespace device
// scan on shuffl functions
#pragma unroll
for (int i = 1; i <= (OPENCV_GPU_WARP_SIZE / 2); i *= 2)
for (int i = 1; i <= (OPENCV_CUDA_WARP_SIZE / 2); i *= 2)
{
const T n = cv::cuda::device::shfl_up(idata, i);
if (laneId >= i)
@@ -187,9 +187,9 @@ namespace cv { namespace cuda { namespace device
return idata;
#else
unsigned int pos = 2 * tid - (tid & (OPENCV_GPU_WARP_SIZE - 1));
unsigned int pos = 2 * tid - (tid & (OPENCV_CUDA_WARP_SIZE - 1));
s_Data[pos] = 0;
pos += OPENCV_GPU_WARP_SIZE;
pos += OPENCV_CUDA_WARP_SIZE;
s_Data[pos] = idata;
s_Data[pos] += s_Data[pos - 1];
@@ -211,7 +211,7 @@ namespace cv { namespace cuda { namespace device
template <int tiNumScanThreads, typename T>
__device__ T blockScanInclusive(T idata, volatile T* s_Data, unsigned int tid)
{
if (tiNumScanThreads > OPENCV_GPU_WARP_SIZE)
if (tiNumScanThreads > OPENCV_CUDA_WARP_SIZE)
{
//Bottom-level inclusive warp scan
T warpResult = warpScanInclusive(idata, s_Data, tid);
@@ -219,15 +219,15 @@ namespace cv { namespace cuda { namespace device
//Save top elements of each warp for exclusive warp scan
//sync to wait for warp scans to complete (because s_Data is being overwritten)
__syncthreads();
if ((tid & (OPENCV_GPU_WARP_SIZE - 1)) == (OPENCV_GPU_WARP_SIZE - 1))
if ((tid & (OPENCV_CUDA_WARP_SIZE - 1)) == (OPENCV_CUDA_WARP_SIZE - 1))
{
s_Data[tid >> OPENCV_GPU_LOG_WARP_SIZE] = warpResult;
s_Data[tid >> OPENCV_CUDA_LOG_WARP_SIZE] = warpResult;
}
//wait for warp scans to complete
__syncthreads();
if (tid < (tiNumScanThreads / OPENCV_GPU_WARP_SIZE) )
if (tid < (tiNumScanThreads / OPENCV_CUDA_WARP_SIZE) )
{
//grab top warp elements
T val = s_Data[tid];
@@ -238,7 +238,7 @@ namespace cv { namespace cuda { namespace device
//return updated warp scans with exclusive scan results
__syncthreads();
return warpResult + s_Data[tid >> OPENCV_GPU_LOG_WARP_SIZE];
return warpResult + s_Data[tid >> OPENCV_CUDA_LOG_WARP_SIZE];
}
else
{
@@ -247,4 +247,4 @@ namespace cv { namespace cuda { namespace device
}
}}}
#endif // __OPENCV_GPU_SCAN_HPP__
#endif // __OPENCV_CUDA_SCAN_HPP__

10
modules/core/include/opencv2/core/cuda/simd_functions.hpp
View File

@@ -70,16 +70,16 @@
* POSSIBILITY OF SUCH DAMAGE.
*/
#ifndef __OPENCV_GPU_SIMD_FUNCTIONS_HPP__
#define __OPENCV_GPU_SIMD_FUNCTIONS_HPP__
#ifndef __OPENCV_CUDA_SIMD_FUNCTIONS_HPP__
#define __OPENCV_CUDA_SIMD_FUNCTIONS_HPP__
#include "common.hpp"
/*
This header file contains inline functions that implement intra-word SIMD
operations, that are hardware accelerated on sm_3x (Kepler) GPUs. Efficient
operations, that are hardware accelerated on sm_3x (Kepler) CUDAs. Efficient
emulation code paths are provided for earlier architectures (sm_1x, sm_2x)
to make the code portable across all GPUs supported by CUDA. The following
to make the code portable across all CUDAs supported by CUDA. The following
functions are currently implemented:
vadd2(a,b) per-halfword unsigned addition, with wrap-around: a + b
@@ -906,4 +906,4 @@ namespace cv { namespace cuda { namespace device
}
}}}
#endif // __OPENCV_GPU_SIMD_FUNCTIONS_HPP__
#endif // __OPENCV_CUDA_SIMD_FUNCTIONS_HPP__

6
modules/core/include/opencv2/core/cuda/transform.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_TRANSFORM_HPP__
#define __OPENCV_GPU_TRANSFORM_HPP__
#ifndef __OPENCV_CUDA_TRANSFORM_HPP__
#define __OPENCV_CUDA_TRANSFORM_HPP__
#include "common.hpp"
#include "utility.hpp"
@@ -64,4 +64,4 @@ namespace cv { namespace cuda { namespace device
}
}}}
#endif // __OPENCV_GPU_TRANSFORM_HPP__
#endif // __OPENCV_CUDA_TRANSFORM_HPP__

6
modules/core/include/opencv2/core/cuda/type_traits.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_TYPE_TRAITS_HPP__
#define __OPENCV_GPU_TYPE_TRAITS_HPP__
#ifndef __OPENCV_CUDA_TYPE_TRAITS_HPP__
#define __OPENCV_CUDA_TYPE_TRAITS_HPP__
#include "detail/type_traits_detail.hpp"
@@ -79,4 +79,4 @@ namespace cv { namespace cuda { namespace device
};
}}}
#endif // __OPENCV_GPU_TYPE_TRAITS_HPP__
#endif // __OPENCV_CUDA_TYPE_TRAITS_HPP__

14
modules/core/include/opencv2/core/cuda/utility.hpp
View File

@@ -40,18 +40,18 @@
//
//M*/
#ifndef __OPENCV_GPU_UTILITY_HPP__
#define __OPENCV_GPU_UTILITY_HPP__
#ifndef __OPENCV_CUDA_UTILITY_HPP__
#define __OPENCV_CUDA_UTILITY_HPP__
#include "saturate_cast.hpp"
#include "datamov_utils.hpp"
namespace cv { namespace cuda { namespace device
{
#define OPENCV_GPU_LOG_WARP_SIZE (5)
#define OPENCV_GPU_WARP_SIZE (1 << OPENCV_GPU_LOG_WARP_SIZE)
#define OPENCV_GPU_LOG_MEM_BANKS ((__CUDA_ARCH__ >= 200) ? 5 : 4) // 32 banks on fermi, 16 on tesla
#define OPENCV_GPU_MEM_BANKS (1 << OPENCV_GPU_LOG_MEM_BANKS)
#define OPENCV_CUDA_LOG_WARP_SIZE (5)
#define OPENCV_CUDA_WARP_SIZE (1 << OPENCV_CUDA_LOG_WARP_SIZE)
#define OPENCV_CUDA_LOG_MEM_BANKS ((__CUDA_ARCH__ >= 200) ? 5 : 4) // 32 banks on fermi, 16 on tesla
#define OPENCV_CUDA_MEM_BANKS (1 << OPENCV_CUDA_LOG_MEM_BANKS)
///////////////////////////////////////////////////////////////////////////////
// swap
@@ -210,4 +210,4 @@ namespace cv { namespace cuda { namespace device
}
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_UTILITY_HPP__
#endif // __OPENCV_CUDA_UTILITY_HPP__

6
modules/core/include/opencv2/core/cuda/vec_distance.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_VEC_DISTANCE_HPP__
#define __OPENCV_GPU_VEC_DISTANCE_HPP__
#ifndef __OPENCV_CUDA_VEC_DISTANCE_HPP__
#define __OPENCV_CUDA_VEC_DISTANCE_HPP__
#include "reduce.hpp"
#include "functional.hpp"
@@ -221,4 +221,4 @@ namespace cv { namespace cuda { namespace device
};
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_VEC_DISTANCE_HPP__
#endif // __OPENCV_CUDA_VEC_DISTANCE_HPP__

6
modules/core/include/opencv2/core/cuda/vec_math.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_VECMATH_HPP__
#define __OPENCV_GPU_VECMATH_HPP__
#ifndef __OPENCV_CUDA_VECMATH_HPP__
#define __OPENCV_CUDA_VECMATH_HPP__
#include "vec_traits.hpp"
#include "saturate_cast.hpp"
@@ -919,4 +919,4 @@ CV_CUDEV_IMPLEMENT_SCALAR_BINARY_FUNC(atan2, ::atan2, double, double, double)
}}} // namespace cv { namespace cuda { namespace device
#endif // __OPENCV_GPU_VECMATH_HPP__
#endif // __OPENCV_CUDA_VECMATH_HPP__

44
modules/core/include/opencv2/core/cuda/vec_traits.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_VEC_TRAITS_HPP__
#define __OPENCV_GPU_VEC_TRAITS_HPP__
#ifndef __OPENCV_CUDA_VEC_TRAITS_HPP__
#define __OPENCV_CUDA_VEC_TRAITS_HPP__
#include "common.hpp"
@@ -122,7 +122,7 @@ namespace cv { namespace cuda { namespace device
return val;
}
#define OPENCV_GPU_IMPLEMENT_TYPE_VEC(type) \
#define OPENCV_CUDA_IMPLEMENT_TYPE_VEC(type) \
template<> struct TypeVec<type, 1> { typedef type vec_type; }; \
template<> struct TypeVec<type ## 1, 1> { typedef type ## 1 vec_type; }; \
template<> struct TypeVec<type, 2> { typedef type ## 2 vec_type; }; \
@@ -134,16 +134,16 @@ namespace cv { namespace cuda { namespace device
template<> struct TypeVec<type, 8> { typedef type ## 8 vec_type; }; \
template<> struct TypeVec<type ## 8, 8> { typedef type ## 8 vec_type; };
OPENCV_GPU_IMPLEMENT_TYPE_VEC(uchar)
OPENCV_GPU_IMPLEMENT_TYPE_VEC(char)
OPENCV_GPU_IMPLEMENT_TYPE_VEC(ushort)
OPENCV_GPU_IMPLEMENT_TYPE_VEC(short)
OPENCV_GPU_IMPLEMENT_TYPE_VEC(int)
OPENCV_GPU_IMPLEMENT_TYPE_VEC(uint)
OPENCV_GPU_IMPLEMENT_TYPE_VEC(float)
OPENCV_GPU_IMPLEMENT_TYPE_VEC(double)
OPENCV_CUDA_IMPLEMENT_TYPE_VEC(uchar)
OPENCV_CUDA_IMPLEMENT_TYPE_VEC(char)
OPENCV_CUDA_IMPLEMENT_TYPE_VEC(ushort)
OPENCV_CUDA_IMPLEMENT_TYPE_VEC(short)
OPENCV_CUDA_IMPLEMENT_TYPE_VEC(int)
OPENCV_CUDA_IMPLEMENT_TYPE_VEC(uint)
OPENCV_CUDA_IMPLEMENT_TYPE_VEC(float)
OPENCV_CUDA_IMPLEMENT_TYPE_VEC(double)
#undef OPENCV_GPU_IMPLEMENT_TYPE_VEC
#undef OPENCV_CUDA_IMPLEMENT_TYPE_VEC
template<> struct TypeVec<schar, 1> { typedef schar vec_type; };
template<> struct TypeVec<schar, 2> { typedef char2 vec_type; };
@@ -159,7 +159,7 @@ namespace cv { namespace cuda { namespace device
template<typename T> struct VecTraits;
#define OPENCV_GPU_IMPLEMENT_VEC_TRAITS(type) \
#define OPENCV_CUDA_IMPLEMENT_VEC_TRAITS(type) \
template<> struct VecTraits<type> \
{ \
typedef type elem_type; \
@@ -209,15 +209,15 @@ namespace cv { namespace cuda { namespace device
static __device__ __host__ __forceinline__ type ## 8 make(const type* v) {return make_ ## type ## 8(v[0], v[1], v[2], v[3], v[4], v[5], v[6], v[7]);} \
};
OPENCV_GPU_IMPLEMENT_VEC_TRAITS(uchar)
OPENCV_GPU_IMPLEMENT_VEC_TRAITS(ushort)
OPENCV_GPU_IMPLEMENT_VEC_TRAITS(short)
OPENCV_GPU_IMPLEMENT_VEC_TRAITS(int)
OPENCV_GPU_IMPLEMENT_VEC_TRAITS(uint)
OPENCV_GPU_IMPLEMENT_VEC_TRAITS(float)
OPENCV_GPU_IMPLEMENT_VEC_TRAITS(double)
OPENCV_CUDA_IMPLEMENT_VEC_TRAITS(uchar)
OPENCV_CUDA_IMPLEMENT_VEC_TRAITS(ushort)
OPENCV_CUDA_IMPLEMENT_VEC_TRAITS(short)
OPENCV_CUDA_IMPLEMENT_VEC_TRAITS(int)
OPENCV_CUDA_IMPLEMENT_VEC_TRAITS(uint)
OPENCV_CUDA_IMPLEMENT_VEC_TRAITS(float)
OPENCV_CUDA_IMPLEMENT_VEC_TRAITS(double)
#undef OPENCV_GPU_IMPLEMENT_VEC_TRAITS
#undef OPENCV_CUDA_IMPLEMENT_VEC_TRAITS
template<> struct VecTraits<char>
{
@@ -277,4 +277,4 @@ namespace cv { namespace cuda { namespace device
};
}}} // namespace cv { namespace cuda { namespace cudev
#endif // __OPENCV_GPU_VEC_TRAITS_HPP__
#endif // __OPENCV_CUDA_VEC_TRAITS_HPP__

6
modules/core/include/opencv2/core/cuda/warp.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_DEVICE_WARP_HPP__
#define __OPENCV_GPU_DEVICE_WARP_HPP__
#ifndef __OPENCV_CUDA_DEVICE_WARP_HPP__
#define __OPENCV_CUDA_DEVICE_WARP_HPP__
namespace cv { namespace cuda { namespace device
{
@@ -128,4 +128,4 @@ namespace cv { namespace cuda { namespace device
};
}}} // namespace cv { namespace cuda { namespace cudev
#endif /* __OPENCV_GPU_DEVICE_WARP_HPP__ */
#endif /* __OPENCV_CUDA_DEVICE_WARP_HPP__ */

6
modules/core/include/opencv2/core/cuda/warp_reduce.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef OPENCV_GPU_WARP_REDUCE_HPP__
#define OPENCV_GPU_WARP_REDUCE_HPP__
#ifndef OPENCV_CUDA_WARP_REDUCE_HPP__
#define OPENCV_CUDA_WARP_REDUCE_HPP__
namespace cv { namespace cuda { namespace device
{
@@ -65,4 +65,4 @@ namespace cv { namespace cuda { namespace device
}
}}} // namespace cv { namespace cuda { namespace cudev {
#endif /* OPENCV_GPU_WARP_REDUCE_HPP__ */
#endif /* OPENCV_CUDA_WARP_REDUCE_HPP__ */

6
modules/core/include/opencv2/core/cuda/warp_shuffle.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_GPU_WARP_SHUFFLE_HPP__
#define __OPENCV_GPU_WARP_SHUFFLE_HPP__
#ifndef __OPENCV_CUDA_WARP_SHUFFLE_HPP__
#define __OPENCV_CUDA_WARP_SHUFFLE_HPP__
namespace cv { namespace cuda { namespace device
{
@@ -142,4 +142,4 @@ namespace cv { namespace cuda { namespace device
}
}}}
#endif // __OPENCV_GPU_WARP_SHUFFLE_HPP__
#endif // __OPENCV_CUDA_WARP_SHUFFLE_HPP__

8
modules/core/include/opencv2/core/cuda_stream_accessor.hpp
View File

@@ -40,8 +40,8 @@
//
//M*/
#ifndef __OPENCV_CORE_GPU_STREAM_ACCESSOR_HPP__
#define __OPENCV_CORE_GPU_STREAM_ACCESSOR_HPP__
#ifndef __OPENCV_CORE_CUDA_STREAM_ACCESSOR_HPP__
#define __OPENCV_CORE_CUDA_STREAM_ACCESSOR_HPP__
#ifndef __cplusplus
# error cuda_stream_accessor.hpp header must be compiled as C++
@@ -49,7 +49,7 @@
// This is only header file that depends on Cuda. All other headers are independent.
// So if you use OpenCV binaries you do noot need to install Cuda Toolkit.
// But of you wanna use GPU by yourself, may get cuda stream instance using the class below.
// But of you wanna use CUDA by yourself, may get cuda stream instance using the class below.
// In this case you have to install Cuda Toolkit.
#include <cuda_runtime.h>
@@ -74,4 +74,4 @@ namespace cv
}
}
#endif /* __OPENCV_CORE_GPU_STREAM_ACCESSOR_HPP__ */
#endif /* __OPENCV_CORE_CUDA_STREAM_ACCESSOR_HPP__ */

40
modules/core/include/opencv2/core/cuda_types.hpp
View File

@@ -40,17 +40,17 @@
//
//M*/
#ifndef __OPENCV_CORE_GPU_TYPES_HPP__
#define __OPENCV_CORE_GPU_TYPES_HPP__
#ifndef __OPENCV_CORE_CUDA_TYPES_HPP__
#define __OPENCV_CORE_CUDA_TYPES_HPP__
#ifndef __cplusplus
# error cuda_types.hpp header must be compiled as C++
#endif
#ifdef __CUDACC__
#define __CV_GPU_HOST_DEVICE__ __host__ __device__ __forceinline__
#define __CV_CUDA_HOST_DEVICE__ __host__ __device__ __forceinline__
#else
#define __CV_GPU_HOST_DEVICE__
#define __CV_CUDA_HOST_DEVICE__
#endif
namespace cv
@@ -69,41 +69,41 @@ namespace cv
T* data;
__CV_GPU_HOST_DEVICE__ DevPtr() : data(0) {}
__CV_GPU_HOST_DEVICE__ DevPtr(T* data_) : data(data_) {}
__CV_CUDA_HOST_DEVICE__ DevPtr() : data(0) {}
__CV_CUDA_HOST_DEVICE__ DevPtr(T* data_) : data(data_) {}
__CV_GPU_HOST_DEVICE__ size_t elemSize() const { return elem_size; }
__CV_GPU_HOST_DEVICE__ operator T*() { return data; }
__CV_GPU_HOST_DEVICE__ operator const T*() const { return data; }
__CV_CUDA_HOST_DEVICE__ size_t elemSize() const { return elem_size; }
__CV_CUDA_HOST_DEVICE__ operator T*() { return data; }
__CV_CUDA_HOST_DEVICE__ operator const T*() const { return data; }
};
template <typename T> struct PtrSz : public DevPtr<T>
{
__CV_GPU_HOST_DEVICE__ PtrSz() : size(0) {}
__CV_GPU_HOST_DEVICE__ PtrSz(T* data_, size_t size_) : DevPtr<T>(data_), size(size_) {}
__CV_CUDA_HOST_DEVICE__ PtrSz() : size(0) {}
__CV_CUDA_HOST_DEVICE__ PtrSz(T* data_, size_t size_) : DevPtr<T>(data_), size(size_) {}
size_t size;
};
template <typename T> struct PtrStep : public DevPtr<T>
{
__CV_GPU_HOST_DEVICE__ PtrStep() : step(0) {}
__CV_GPU_HOST_DEVICE__ PtrStep(T* data_, size_t step_) : DevPtr<T>(data_), step(step_) {}
__CV_CUDA_HOST_DEVICE__ PtrStep() : step(0) {}
__CV_CUDA_HOST_DEVICE__ PtrStep(T* data_, size_t step_) : DevPtr<T>(data_), step(step_) {}
//! stride between two consecutive rows in bytes. Step is stored always and everywhere in bytes!!!
size_t step;
__CV_GPU_HOST_DEVICE__ T* ptr(int y = 0) { return ( T*)( ( char*)DevPtr<T>::data + y * step); }
__CV_GPU_HOST_DEVICE__ const T* ptr(int y = 0) const { return (const T*)( (const char*)DevPtr<T>::data + y * step); }
__CV_CUDA_HOST_DEVICE__ T* ptr(int y = 0) { return ( T*)( ( char*)DevPtr<T>::data + y * step); }
__CV_CUDA_HOST_DEVICE__ const T* ptr(int y = 0) const { return (const T*)( (const char*)DevPtr<T>::data + y * step); }
__CV_GPU_HOST_DEVICE__ T& operator ()(int y, int x) { return ptr(y)[x]; }
__CV_GPU_HOST_DEVICE__ const T& operator ()(int y, int x) const { return ptr(y)[x]; }
__CV_CUDA_HOST_DEVICE__ T& operator ()(int y, int x) { return ptr(y)[x]; }
__CV_CUDA_HOST_DEVICE__ const T& operator ()(int y, int x) const { return ptr(y)[x]; }
};
template <typename T> struct PtrStepSz : public PtrStep<T>
{
__CV_GPU_HOST_DEVICE__ PtrStepSz() : cols(0), rows(0) {}
__CV_GPU_HOST_DEVICE__ PtrStepSz(int rows_, int cols_, T* data_, size_t step_)
__CV_CUDA_HOST_DEVICE__ PtrStepSz() : cols(0), rows(0) {}
__CV_CUDA_HOST_DEVICE__ PtrStepSz(int rows_, int cols_, T* data_, size_t step_)
: PtrStep<T>(data_, step_), cols(cols_), rows(rows_) {}
template <typename U>
@@ -123,4 +123,4 @@ namespace cv
}
}
#endif /* __OPENCV_CORE_GPU_TYPES_HPP__ */
#endif /* __OPENCV_CORE_CUDA_TYPES_HPP__ */

10
modules/core/include/opencv2/core/private.cuda.hpp
View File

@@ -41,8 +41,8 @@
//
//M*/
#ifndef __OPENCV_CORE_PRIVATE_GPU_HPP__
#define __OPENCV_CORE_PRIVATE_GPU_HPP__
#ifndef __OPENCV_CORE_PRIVATE_CUDA_HPP__
#define __OPENCV_CORE_PRIVATE_CUDA_HPP__
#ifndef __OPENCV_BUILD
# error this is a private header which should not be used from outside of the OpenCV library
@@ -71,7 +71,7 @@
# endif
# if defined(CUDA_ARCH_BIN_OR_PTX_10)
# error "OpenCV GPU module doesn't support NVIDIA compute capability 1.0"
# error "OpenCV CUDA module doesn't support NVIDIA compute capability 1.0"
# endif
#endif
@@ -82,7 +82,7 @@ namespace cv { namespace cuda {
#ifndef HAVE_CUDA
static inline void throw_no_cuda() { CV_Error(cv::Error::GpuNotSupported, "The library is compiled without GPU support"); }
static inline void throw_no_cuda() { CV_Error(cv::Error::GpuNotSupported, "The library is compiled without CUDA support"); }
#else // HAVE_CUDA
@@ -140,4 +140,4 @@ namespace cv { namespace cuda
#endif // HAVE_CUDA
#endif // __OPENCV_CORE_GPU_PRIVATE_HPP__
#endif // __OPENCV_CORE_CUDA_PRIVATE_HPP__