Skip to content

O
opencv
Commit 9cb4292d authored by Vladislav Vinogradov's avatar Vladislav Vinogradov
Browse files
Options

implemented Luv/Lab <-> RGB conversion

parent e446903a
Hide whitespace changes
Inline Side-by-side
Showing with 696 additions and 219 deletions
modules/gpu/include/opencv2/gpu/device/color.hpp
View file @ 9cb4292d
......@@ -216,6 +216,86 @@ namespace cv { namespace gpu { namespace device
OPENCV_GPU_IMPLEMENT_HLS2RGB_TRAITS(hls4_to_bgra, 4, 4, 0)
#undef OPENCV_GPU_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_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)
#undef OPENCV_GPU_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_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)
#undef OPENCV_GPU_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_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)
#undef OPENCV_GPU_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_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)
#undef OPENCV_GPU_IMPLEMENT_Luv2RGB_TRAITS
}}} // namespace cv { namespace gpu { namespace device
#endif // __OPENCV_GPU_BORDER_INTERPOLATE_HPP__
modules/gpu/include/opencv2/gpu/device/detail/color_detail.hpp
View file @ 9cb4292d
This source diff could not be displayed because it is too large. You can view the blob instead.
modules/gpu/perf/perf_imgproc.cpp
View file @ 9cb4292d
......@@ -1511,13 +1511,13 @@ PERF_TEST_P(Sz_Depth_Code, ImgProc_CvtColor, Combine(
CvtColorInfo(3, 3, cv::COLOR_BGR2HLS),
CvtColorInfo(3, 3, cv::COLOR_HLS2BGR),
CvtColorInfo(3, 3, cv::COLOR_BGR2Lab),
CvtColorInfo(3, 3, cv::COLOR_RGB2Lab),
CvtColorInfo(3, 3, cv::COLOR_LBGR2Lab),
CvtColorInfo(3, 3, cv::COLOR_BGR2Luv),
CvtColorInfo(3, 3, cv::COLOR_RGB2Luv),
CvtColorInfo(3, 3, cv::COLOR_LBGR2Luv),
CvtColorInfo(3, 3, cv::COLOR_Lab2BGR),
CvtColorInfo(3, 3, cv::COLOR_Lab2RGB),
CvtColorInfo(3, 3, cv::COLOR_Luv2BGR),
CvtColorInfo(3, 3, cv::COLOR_Lab2LBGR),
CvtColorInfo(3, 3, cv::COLOR_Luv2RGB),
CvtColorInfo(3, 3, cv::COLOR_Luv2LRGB),
CvtColorInfo(1, 3, cv::COLOR_BayerBG2BGR),
CvtColorInfo(1, 3, cv::COLOR_BayerGB2BGR),
CvtColorInfo(1, 3, cv::COLOR_BayerRG2BGR),
......
modules/gpu/src/color.cpp
View file @ 9cb4292d
......@@ -53,7 +53,7 @@ void cv::gpu::gammaCorrection(const GpuMat&, GpuMat&, bool, Stream&) { throw_nog
#else /* !defined (HAVE_CUDA) */
#include <cvt_colot_internal.h>
#include "cvt_color_internal.h"
namespace cv { namespace gpu {
namespace device
......@@ -69,7 +69,7 @@ using namespace ::cv::gpu::device;
namespace
{
typedef void (*gpu_func_t)(const PtrStepSzb& src, const PtrStepSzb& dst, cudaStream_t stream);
typedef void (*gpu_func_t)(PtrStepSzb src, PtrStepSzb dst, cudaStream_t stream);
void bgr_to_rgb(const GpuMat& src, GpuMat& dst, int, Stream& stream)
{
......@@ -1155,154 +1155,420 @@ namespace
funcs[dcn == 4][src.channels() == 4][src.depth()](src, dst, StreamAccessor::getStream(stream));
}
void bgr_to_lab(const GpuMat& src, GpuMat& dst, int dcn, Stream& st)
void bgr_to_lab(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
#if (CUDA_VERSION < 5000)
(void)src;
(void)dst;
(void)dcn;
(void)st;
CV_Error( CV_StsBadFlag, "Unknown/unsupported color conversion code" );
#else
CV_Assert(src.depth() == CV_8U);
CV_Assert(src.channels() == 3);
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{bgr_to_lab_8u, bgr_to_lab_32f},
{bgra_to_lab_8u, bgra_to_lab_32f}
},
{
{bgr_to_lab4_8u, bgr_to_lab4_32f},
{bgra_to_lab4_8u, bgra_to_lab4_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void rgb_to_lab(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{rgb_to_lab_8u, rgb_to_lab_32f},
{rgba_to_lab_8u, rgba_to_lab_32f}
},
{
{rgb_to_lab4_8u, rgb_to_lab4_32f},
{rgba_to_lab4_8u, rgba_to_lab4_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void lbgr_to_lab(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{lbgr_to_lab_8u, lbgr_to_lab_32f},
{lbgra_to_lab_8u, lbgra_to_lab_32f}
},
{
{lbgr_to_lab4_8u, lbgr_to_lab4_32f},
{lbgra_to_lab4_8u, lbgra_to_lab4_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void lrgb_to_lab(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{lrgb_to_lab_8u, lrgb_to_lab_32f},
{lrgba_to_lab_8u, lrgba_to_lab_32f}
},
{
{lrgb_to_lab4_8u, lrgb_to_lab4_32f},
{lrgba_to_lab4_8u, lrgba_to_lab4_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void lab_to_bgr(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{lab_to_bgr_8u, lab_to_bgr_32f},
{lab4_to_bgr_8u, lab4_to_bgr_32f}
},
{
{lab_to_bgra_8u, lab_to_bgra_32f},
{lab4_to_bgra_8u, lab4_to_bgra_32f}
}
};
dcn = src.channels();
if (dcn <= 0) dcn = 3;
dst.create(src.size(), CV_MAKETYPE(src.depth(), dcn));
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void lab_to_rgb(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{lab_to_rgb_8u, lab_to_rgb_32f},
{lab4_to_rgb_8u, lab4_to_rgb_32f}
},
{
{lab_to_rgba_8u, lab_to_rgba_32f},
{lab4_to_rgba_8u, lab4_to_rgba_32f}
}
};
cudaStream_t stream = StreamAccessor::getStream(st);
NppStreamHandler h(stream);
if (dcn <= 0) dcn = 3;
NppiSize oSizeROI;
oSizeROI.width = src.cols;
oSizeROI.height = src.rows;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
nppSafeCall( nppiBGRToLab_8u_C3R(src.ptr<Npp8u>(), static_cast<int>(src.step), dst.ptr<Npp8u>(), static_cast<int>(dst.step), oSizeROI) );
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
if (stream == 0)
cudaSafeCall( cudaDeviceSynchronize() );
#endif
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void rgb_to_lab(const GpuMat& src, GpuMat& dst, int, Stream& stream)
void lab_to_lbgr(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
bgr_to_rgb(src, dst, -1, stream);
bgr_to_lab(dst, dst, -1, stream);
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{lab_to_lbgr_8u, lab_to_lbgr_32f},
{lab4_to_lbgr_8u, lab4_to_lbgr_32f}
},
{
{lab_to_lbgra_8u, lab_to_lbgra_32f},
{lab4_to_lbgra_8u, lab4_to_lbgra_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void lab_to_bgr(const GpuMat& src, GpuMat& dst, int dcn, Stream& st)
void lab_to_lrgb(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
#if (CUDA_VERSION < 5000)
(void)src;
(void)dst;
(void)dcn;
(void)st;
CV_Error( CV_StsBadFlag, "Unknown/unsupported color conversion code" );
#else
CV_Assert(src.depth() == CV_8U);
CV_Assert(src.channels() == 3);
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{lab_to_lrgb_8u, lab_to_lrgb_32f},
{lab4_to_lrgb_8u, lab4_to_lrgb_32f}
},
{
{lab_to_lrgba_8u, lab_to_lrgba_32f},
{lab4_to_lrgba_8u, lab4_to_lrgba_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dcn = src.channels();
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
dst.create(src.size(), CV_MAKETYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void bgr_to_luv(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{bgr_to_luv_8u, bgr_to_luv_32f},
{bgra_to_luv_8u, bgra_to_luv_32f}
},
{
{bgr_to_luv4_8u, bgr_to_luv4_32f},
{bgra_to_luv4_8u, bgra_to_luv4_32f}
}
};
cudaStream_t stream = StreamAccessor::getStream(st);
NppStreamHandler h(stream);
if (dcn <= 0) dcn = 3;
NppiSize oSizeROI;
oSizeROI.width = src.cols;
oSizeROI.height = src.rows;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
nppSafeCall( nppiLabToBGR_8u_C3R(src.ptr<Npp8u>(), static_cast<int>(src.step), dst.ptr<Npp8u>(), static_cast<int>(dst.step), oSizeROI) );
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
if (stream == 0)
cudaSafeCall( cudaDeviceSynchronize() );
#endif
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void lab_to_rgb(const GpuMat& src, GpuMat& dst, int, Stream& stream)
void rgb_to_luv(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
lab_to_bgr(src, dst, -1, stream);
bgr_to_rgb(dst, dst, -1, stream);
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{rgb_to_luv_8u, rgb_to_luv_32f},
{rgba_to_luv_8u, rgba_to_luv_32f}
},
{
{rgb_to_luv4_8u, rgb_to_luv4_32f},
{rgba_to_luv4_8u, rgba_to_luv4_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void rgb_to_luv(const GpuMat& src, GpuMat& dst, int dcn, Stream& st)
void lbgr_to_luv(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
#if (CUDA_VERSION < 5000)
(void)src;
(void)dst;
(void)dcn;
(void)st;
CV_Error( CV_StsBadFlag, "Unknown/unsupported color conversion code" );
#else
CV_Assert(src.depth() == CV_8U);
CV_Assert(src.channels() == 3 || src.channels() == 4);
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{lbgr_to_luv_8u, lbgr_to_luv_32f},
{lbgra_to_luv_8u, lbgra_to_luv_32f}
},
{
{lbgr_to_luv4_8u, lbgr_to_luv4_32f},
{lbgra_to_luv4_8u, lbgra_to_luv4_32f}
}
};
if (dcn <= 0) dcn = 3;
dcn = src.channels();
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
dst.create(src.size(), CV_MAKETYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
cudaStream_t stream = StreamAccessor::getStream(st);
NppStreamHandler h(stream);
void lrgb_to_luv(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{lrgb_to_luv_8u, lrgb_to_luv_32f},
{lrgba_to_luv_8u, lrgba_to_luv_32f}
},
{
{lrgb_to_luv4_8u, lrgb_to_luv4_32f},
{lrgba_to_luv4_8u, lrgba_to_luv4_32f}
}
};
if (dcn <= 0) dcn = 3;
NppiSize oSizeROI;
oSizeROI.width = src.cols;
oSizeROI.height = src.rows;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
if (dcn == 3)
nppSafeCall( nppiRGBToLUV_8u_C3R(src.ptr<Npp8u>(), static_cast<int>(src.step), dst.ptr<Npp8u>(), static_cast<int>(dst.step), oSizeROI) );
else
nppSafeCall( nppiRGBToLUV_8u_AC4R(src.ptr<Npp8u>(), static_cast<int>(src.step), dst.ptr<Npp8u>(), static_cast<int>(dst.step), oSizeROI) );
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
if (stream == 0)
cudaSafeCall( cudaDeviceSynchronize() );
#endif
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void bgr_to_luv(const GpuMat& src, GpuMat& dst, int, Stream& stream)
void luv_to_bgr(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
bgr_to_rgb(src, dst, -1, stream);
rgb_to_luv(dst, dst, -1, stream);
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{luv_to_bgr_8u, luv_to_bgr_32f},
{luv4_to_bgr_8u, luv4_to_bgr_32f}
},
{
{luv_to_bgra_8u, luv_to_bgra_32f},
{luv4_to_bgra_8u, luv4_to_bgra_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void luv_to_rgb(const GpuMat& src, GpuMat& dst, int dcn, Stream& st)
void luv_to_rgb(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
#if (CUDA_VERSION < 5000)
(void)src;
(void)dst;
(void)dcn;
(void)st;
CV_Error( CV_StsBadFlag, "Unknown/unsupported color conversion code" );
#else
CV_Assert(src.depth() == CV_8U);
CV_Assert(src.channels() == 3 || src.channels() == 4);
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{luv_to_rgb_8u, luv_to_rgb_32f},
{luv4_to_rgb_8u, luv4_to_rgb_32f}
},
{
{luv_to_rgba_8u, luv_to_rgba_32f},
{luv4_to_rgba_8u, luv4_to_rgba_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dcn = src.channels();
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
dst.create(src.size(), CV_MAKETYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
cudaStream_t stream = StreamAccessor::getStream(st);
NppStreamHandler h(stream);
void luv_to_lbgr(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{luv_to_lbgr_8u, luv_to_lbgr_32f},
{luv4_to_lbgr_8u, luv4_to_lbgr_32f}
},
{
{luv_to_lbgra_8u, luv_to_lbgra_32f},
{luv4_to_lbgra_8u, luv4_to_lbgra_32f}
}
};
if (dcn <= 0) dcn = 3;
NppiSize oSizeROI;
oSizeROI.width = src.cols;
oSizeROI.height = src.rows;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
if (dcn == 3)
nppSafeCall( nppiLUVToRGB_8u_C3R(src.ptr<Npp8u>(), static_cast<int>(src.step), dst.ptr<Npp8u>(), static_cast<int>(dst.step), oSizeROI) );
else
nppSafeCall( nppiLUVToRGB_8u_AC4R(src.ptr<Npp8u>(), static_cast<int>(src.step), dst.ptr<Npp8u>(), static_cast<int>(dst.step), oSizeROI) );
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
if (stream == 0)
cudaSafeCall( cudaDeviceSynchronize() );
#endif
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void luv_to_bgr(const GpuMat& src, GpuMat& dst, int, Stream& stream)
void luv_to_lrgb(const GpuMat& src, GpuMat& dst, int dcn, Stream& stream)
{
luv_to_rgb(src, dst, -1, stream);
bgr_to_rgb(dst, dst, -1, stream);
using namespace cv::gpu::device;
static const gpu_func_t funcs[2][2][2] =
{
{
{luv_to_lrgb_8u, luv_to_lrgb_32f},
{luv4_to_lrgb_8u, luv4_to_lrgb_32f}
},
{
{luv_to_lrgba_8u, luv_to_lrgba_32f},
{luv4_to_lrgba_8u, luv4_to_lrgba_32f}
}
};
if (dcn <= 0) dcn = 3;
CV_Assert(src.depth() == CV_8U || src.depth() == CV_32F);
CV_Assert(src.channels() == 3 || src.channels() == 4);
CV_Assert(dcn == 3 || dcn == 4);
dst.create(src.size(), CV_MAKE_TYPE(src.depth(), dcn));
funcs[dcn == 4][src.channels() == 4][src.depth() == CV_32F](src, dst, StreamAccessor::getStream(stream));
}
void rgba_to_mbgra(const GpuMat& src, GpuMat& dst, int, Stream& st)
......@@ -1475,15 +1741,15 @@ void cv::gpu::cvtColor(const GpuMat& src, GpuMat& dst, int code, int dcn, Stream
hls_to_bgr_full, // CV_HLS2BGR_FULL = 72
hls_to_rgb_full, // CV_HLS2RGB_FULL = 73
0, // CV_LBGR2Lab = 74
0, // CV_LRGB2Lab = 75
0, // CV_LBGR2Luv = 76
0, // CV_LRGB2Luv = 77
lbgr_to_lab, // CV_LBGR2Lab = 74
lrgb_to_lab, // CV_LRGB2Lab = 75
lbgr_to_luv, // CV_LBGR2Luv = 76
lrgb_to_luv, // CV_LRGB2Luv = 77
0, // CV_Lab2LBGR = 78
0, // CV_Lab2LRGB = 79
0, // CV_Luv2LBGR = 80
0, // CV_Luv2LRGB = 81
lab_to_lbgr, // CV_Lab2LBGR = 78
lab_to_lrgb, // CV_Lab2LRGB = 79
luv_to_lbgr, // CV_Luv2LBGR = 80
luv_to_lrgb, // CV_Luv2LRGB = 81
bgr_to_yuv, // CV_BGR2YUV = 82
rgb_to_yuv, // CV_RGB2YUV = 83
......
modules/gpu/src/cuda/color.cu
View file @ 9cb4292d
......@@ -42,10 +42,10 @@
#if !defined CUDA_DISABLER
#include <internal_shared.hpp>
#include <opencv2/gpu/device/transform.hpp>
#include <opencv2/gpu/device/color.hpp>
#include <cvt_colot_internal.h>
#include "internal_shared.hpp"
#include "opencv2/gpu/device/transform.hpp"
#include "opencv2/gpu/device/color.hpp"
#include "cvt_color_internal.h"
namespace cv { namespace gpu { namespace device
{
......@@ -224,7 +224,7 @@ namespace cv { namespace gpu { namespace device
};
#define OPENCV_GPU_IMPLEMENT_CVTCOLOR(name, traits) \
void name(const PtrStepSzb& src, const PtrStepSzb& dst, cudaStream_t stream) \
void name(PtrStepSzb src, PtrStepSzb dst, cudaStream_t stream) \
{ \
traits::functor_type functor = traits::create_functor(); \
typedef typename traits::functor_type::argument_type src_t; \
......@@ -241,6 +241,10 @@ namespace cv { namespace gpu { namespace device
OPENCV_GPU_IMPLEMENT_CVTCOLOR(name ## _32f, name ## _traits<float>)
#define OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(name) \
OPENCV_GPU_IMPLEMENT_CVTCOLOR(name ## _8u, name ## _traits<uchar>) \
OPENCV_GPU_IMPLEMENT_CVTCOLOR(name ## _32f, name ## _traits<float>)
#define OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(name) \
OPENCV_GPU_IMPLEMENT_CVTCOLOR(name ## _8u, name ## _traits<uchar>) \
OPENCV_GPU_IMPLEMENT_CVTCOLOR(name ## _32f, name ## _traits<float>) \
OPENCV_GPU_IMPLEMENT_CVTCOLOR(name ## _full_8u, name ## _full_traits<uchar>) \
......@@ -339,46 +343,119 @@ namespace cv { namespace gpu { namespace device
OPENCV_GPU_IMPLEMENT_CVTCOLOR_ALL(xyz_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_ALL(xyz4_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgb_to_hsv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgba_to_hsv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgb_to_hsv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgba_to_hsv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgr_to_hsv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgra_to_hsv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgr_to_hsv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgra_to_hsv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hsv_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hsv_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hsv4_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hsv4_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hsv_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hsv_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hsv4_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hsv4_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgb_to_hls)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgba_to_hls)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgb_to_hls4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgba_to_hls4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgr_to_hls)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgra_to_hls)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgr_to_hls4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgra_to_hls4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hls_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hls_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hls4_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hls4_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hls_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hls_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hls4_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(hls4_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(rgb_to_hsv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(rgba_to_hsv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(rgb_to_hsv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(rgba_to_hsv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(bgr_to_hsv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(bgra_to_hsv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(bgr_to_hsv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(bgra_to_hsv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hsv_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hsv_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hsv4_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hsv4_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hsv_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hsv_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hsv4_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hsv4_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(rgb_to_hls)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(rgba_to_hls)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(rgb_to_hls4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(rgba_to_hls4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(bgr_to_hls)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(bgra_to_hls)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(bgr_to_hls4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(bgra_to_hls4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hls_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hls_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hls4_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hls4_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hls_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hls_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hls4_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL(hls4_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgb_to_lab)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgba_to_lab)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgb_to_lab4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgba_to_lab4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgr_to_lab)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgra_to_lab)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgr_to_lab4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgra_to_lab4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lrgb_to_lab)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lrgba_to_lab)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lrgb_to_lab4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lrgba_to_lab4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lbgr_to_lab)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lbgra_to_lab)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lbgr_to_lab4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lbgra_to_lab4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab4_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab4_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab4_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab4_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab_to_lrgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab4_to_lrgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab_to_lrgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab4_to_lrgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab_to_lbgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab4_to_lbgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab_to_lbgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lab4_to_lbgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgb_to_luv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgba_to_luv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgb_to_luv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(rgba_to_luv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgr_to_luv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgra_to_luv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgr_to_luv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(bgra_to_luv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lrgb_to_luv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lrgba_to_luv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lrgb_to_luv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lrgba_to_luv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lbgr_to_luv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lbgra_to_luv)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lbgr_to_luv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(lbgra_to_luv4)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv4_to_rgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv4_to_rgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv4_to_bgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv4_to_bgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv_to_lrgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv4_to_lrgb)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv_to_lrgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv4_to_lrgba)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv_to_lbgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv4_to_lbgr)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv_to_lbgra)
OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F(luv4_to_lbgra)
#undef OPENCV_GPU_IMPLEMENT_CVTCOLOR
#undef OPENCV_GPU_IMPLEMENT_CVTCOLOR_ONE
#undef OPENCV_GPU_IMPLEMENT_CVTCOLOR_ALL
#undef OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F
#undef OPENCV_GPU_IMPLEMENT_CVTCOLOR_8U32F_FULL
}}} // namespace cv { namespace gpu { namespace device
#endif /* CUDA_DISABLER */
\ No newline at end of file
#endif /* CUDA_DISABLER */
modules/gpu/src/cvt_colot_internal.h → modules/gpu/src/cvt_color_internal.h
View file @ 9cb4292d
......@@ -45,15 +45,19 @@
namespace cv { namespace gpu { namespace device
{
#define OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name) \
void name(const PtrStepSzb& src, const PtrStepSzb& dst, cudaStream_t stream);
#define OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name) \
void name(PtrStepSzb src, PtrStepSzb dst, cudaStream_t stream);
#define OPENCV_GPU_DECLARE_CVTCOLOR_ALL(name) \
OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name ## _8u) \
OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name ## _16u) \
OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name ## _32f)
#define OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(name) \
#define OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(name) \
OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name ## _8u) \
OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name ## _32f)
#define OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(name) \
OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name ## _8u) \
OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name ## _32f) \
OPENCV_GPU_DECLARE_CVTCOLOR_ONE(name ## _full_8u) \
......@@ -152,46 +156,119 @@ namespace cv { namespace gpu { namespace device
OPENCV_GPU_DECLARE_CVTCOLOR_ALL(xyz_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_ALL(xyz4_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgb_to_hsv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgba_to_hsv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgb_to_hsv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgba_to_hsv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgr_to_hsv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgra_to_hsv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgr_to_hsv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgra_to_hsv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hsv_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hsv_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hsv4_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hsv4_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hsv_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hsv_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hsv4_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hsv4_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgb_to_hls)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgba_to_hls)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgb_to_hls4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgba_to_hls4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgr_to_hls)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgra_to_hls)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgr_to_hls4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgra_to_hls4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hls_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hls_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hls4_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hls4_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hls_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hls_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hls4_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(hls4_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(rgb_to_hsv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(rgba_to_hsv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(rgb_to_hsv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(rgba_to_hsv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(bgr_to_hsv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(bgra_to_hsv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(bgr_to_hsv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(bgra_to_hsv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hsv_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hsv_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hsv4_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hsv4_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hsv_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hsv_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hsv4_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hsv4_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(rgb_to_hls)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(rgba_to_hls)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(rgb_to_hls4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(rgba_to_hls4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(bgr_to_hls)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(bgra_to_hls)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(bgr_to_hls4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(bgra_to_hls4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hls_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hls_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hls4_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hls4_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hls_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hls_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hls4_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL(hls4_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgb_to_lab)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgba_to_lab)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgb_to_lab4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgba_to_lab4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgr_to_lab)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgra_to_lab)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgr_to_lab4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgra_to_lab4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lrgb_to_lab)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lrgba_to_lab)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lrgb_to_lab4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lrgba_to_lab4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lbgr_to_lab)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lbgra_to_lab)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lbgr_to_lab4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lbgra_to_lab4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab4_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab4_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab4_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab4_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab_to_lrgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab4_to_lrgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab_to_lrgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab4_to_lrgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab_to_lbgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab4_to_lbgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab_to_lbgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lab4_to_lbgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgb_to_luv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgba_to_luv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgb_to_luv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(rgba_to_luv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgr_to_luv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgra_to_luv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgr_to_luv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(bgra_to_luv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lrgb_to_luv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lrgba_to_luv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lrgb_to_luv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lrgba_to_luv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lbgr_to_luv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lbgra_to_luv)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lbgr_to_luv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(lbgra_to_luv4)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv4_to_rgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv4_to_rgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv4_to_bgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv4_to_bgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv_to_lrgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv4_to_lrgb)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv_to_lrgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv4_to_lrgba)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv_to_lbgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv4_to_lbgr)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv_to_lbgra)
OPENCV_GPU_DECLARE_CVTCOLOR_8U32F(luv4_to_lbgra)
#undef OPENCV_GPU_DECLARE_CVTCOLOR_ONE
#undef OPENCV_GPU_DECLARE_CVTCOLOR_ALL
#undef OPENCV_GPU_DECLARE_CVTCOLOR_8U32F
#undef OPENCV_GPU_DECLARE_CVTCOLOR_8U32F_FULL
}}}
#endif
modules/gpu/src/denoising.cpp
View file @ 9cb4292d
......@@ -176,28 +176,11 @@ void cv::gpu::FastNonLocalMeansDenoising::simpleMethod(const GpuMat& src, GpuMat
void cv::gpu::FastNonLocalMeansDenoising::labMethod( const GpuMat& src, GpuMat& dst, float h_luminance, float h_color, int search_window, int block_window, Stream& s)
{
#if (CUDA_VERSION < 5000)
(void)src;
(void)dst;
(void)h_luminance;
(void)h_color;
(void)search_window;
(void)block_window;
(void)s;
CV_Error( CV_GpuApiCallError, "Lab method required CUDA 5.0 and higher" );
#else
CV_Assert(src.type() == CV_8UC3);
lab.create(src.size(), src.type());
cv::gpu::cvtColor(src, lab, CV_BGR2Lab, 0, s);
/*Mat t;
cv::cvtColor(Mat(src), t, CV_BGR2Lab);
lab.upload(t);*/
l.create(src.size(), CV_8U);
ab.create(src.size(), CV_8UC2);
device::imgproc::fnlm_split_channels(lab, l, ab, StreamAccessor::getStream(s));
......@@ -207,11 +190,6 @@ void cv::gpu::FastNonLocalMeansDenoising::labMethod( const GpuMat& src, GpuMat&
device::imgproc::fnlm_merge_channels(l, ab, lab, StreamAccessor::getStream(s));
cv::gpu::cvtColor(lab, dst, CV_Lab2BGR, 0, s);
/*cv::cvtColor(Mat(lab), t, CV_Lab2BGR);
dst.upload(t);*/
#endif
}
......
Markdown is supported
0% or
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment