From bcf8bdb40156ccd4fa77201ae5d7e5ad127c8b67 Mon Sep 17 00:00:00 2001 From: Vladislav Vinogradov Date: Mon, 3 Jun 2013 14:41:23 +0400 Subject: [PATCH] fixed constructors for functional objects (added __host__ modifier) --- .../opencv2/gpu/device/detail/color_detail.hpp | 181 ++++++++--------- .../gpu/include/opencv2/gpu/device/functional.hpp | 214 ++++++++++----------- modules/gpu/include/opencv2/gpu/device/utility.hpp | 4 +- modules/gpu/src/cuda/calib3d.cu | 8 +- modules/gpu/src/cuda/canny.cu | 12 +- modules/gpu/src/cuda/element_operations.cu | 100 +++++----- 6 files changed, 246 insertions(+), 273 deletions(-) diff --git a/modules/gpu/include/opencv2/gpu/device/detail/color_detail.hpp b/modules/gpu/include/opencv2/gpu/device/detail/color_detail.hpp index d02027f..5b42284 100644 --- a/modules/gpu/include/opencv2/gpu/device/detail/color_detail.hpp +++ b/modules/gpu/include/opencv2/gpu/device/detail/color_detail.hpp @@ -120,11 +120,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2RGB() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - - __device__ __forceinline__ RGB2RGB(const RGB2RGB& other_) - :unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ RGB2RGB() {} + __host__ __device__ __forceinline__ RGB2RGB(const RGB2RGB&) {} }; template <> struct RGB2RGB : unary_function @@ -141,8 +138,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2RGB():unary_function(){} - __device__ __forceinline__ RGB2RGB(const RGB2RGB& other_):unary_function(){} + __host__ __device__ __forceinline__ RGB2RGB() {} + __host__ __device__ __forceinline__ RGB2RGB(const RGB2RGB&) {} }; } @@ -203,8 +200,8 @@ namespace cv { namespace gpu { namespace device return RGB2RGB5x5Converter::cvt(src); } - __device__ __forceinline__ RGB2RGB5x5():unary_function(){} - __device__ __forceinline__ RGB2RGB5x5(const RGB2RGB5x5& other_):unary_function(){} + __host__ __device__ __forceinline__ RGB2RGB5x5() {} + __host__ __device__ __forceinline__ RGB2RGB5x5(const RGB2RGB5x5&) {} }; template struct RGB2RGB5x5<4, bidx,green_bits> : unary_function @@ -214,8 +211,8 @@ namespace cv { namespace gpu { namespace device return RGB2RGB5x5Converter::cvt(src); } - __device__ __forceinline__ RGB2RGB5x5():unary_function(){} - __device__ __forceinline__ RGB2RGB5x5(const RGB2RGB5x5& other_):unary_function(){} + __host__ __device__ __forceinline__ RGB2RGB5x5() {} + __host__ __device__ __forceinline__ RGB2RGB5x5(const RGB2RGB5x5&) {} }; } @@ -282,8 +279,8 @@ namespace cv { namespace gpu { namespace device RGB5x52RGBConverter::cvt(src, dst); return dst; } - __device__ __forceinline__ RGB5x52RGB():unary_function(){} - __device__ __forceinline__ RGB5x52RGB(const RGB5x52RGB& other_):unary_function(){} + __host__ __device__ __forceinline__ RGB5x52RGB() {} + __host__ __device__ __forceinline__ RGB5x52RGB(const RGB5x52RGB&) {} }; @@ -295,8 +292,8 @@ namespace cv { namespace gpu { namespace device RGB5x52RGBConverter::cvt(src, dst); return dst; } - __device__ __forceinline__ RGB5x52RGB():unary_function(){} - __device__ __forceinline__ RGB5x52RGB(const RGB5x52RGB& other_):unary_function(){} + __host__ __device__ __forceinline__ RGB5x52RGB() {} + __host__ __device__ __forceinline__ RGB5x52RGB(const RGB5x52RGB&) {} }; } @@ -325,9 +322,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ Gray2RGB():unary_function::vec_type>(){} - __device__ __forceinline__ Gray2RGB(const Gray2RGB& other_) - : unary_function::vec_type>(){} + __host__ __device__ __forceinline__ Gray2RGB() {} + __host__ __device__ __forceinline__ Gray2RGB(const Gray2RGB&) {} }; template <> struct Gray2RGB : unary_function @@ -342,8 +338,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ Gray2RGB():unary_function(){} - __device__ __forceinline__ Gray2RGB(const Gray2RGB& other_):unary_function(){} + __host__ __device__ __forceinline__ Gray2RGB() {} + __host__ __device__ __forceinline__ Gray2RGB(const Gray2RGB&) {} }; } @@ -384,8 +380,8 @@ namespace cv { namespace gpu { namespace device return Gray2RGB5x5Converter::cvt(src); } - __device__ __forceinline__ Gray2RGB5x5():unary_function(){} - __device__ __forceinline__ Gray2RGB5x5(const Gray2RGB5x5& other_):unary_function(){} + __host__ __device__ __forceinline__ Gray2RGB5x5() {} + __host__ __device__ __forceinline__ Gray2RGB5x5(const Gray2RGB5x5&) {} }; } @@ -426,8 +422,8 @@ namespace cv { namespace gpu { namespace device { return RGB5x52GrayConverter::cvt(src); } - __device__ __forceinline__ RGB5x52Gray() : unary_function(){} - __device__ __forceinline__ RGB5x52Gray(const RGB5x52Gray& other_) : unary_function(){} + __host__ __device__ __forceinline__ RGB5x52Gray() {} + __host__ __device__ __forceinline__ RGB5x52Gray(const RGB5x52Gray&) {} }; } @@ -467,9 +463,8 @@ namespace cv { namespace gpu { namespace device { return RGB2GrayConvert(&src.x); } - __device__ __forceinline__ RGB2Gray() : unary_function::vec_type, T>(){} - __device__ __forceinline__ RGB2Gray(const RGB2Gray& other_) - : unary_function::vec_type, T>(){} + __host__ __device__ __forceinline__ RGB2Gray() {} + __host__ __device__ __forceinline__ RGB2Gray(const RGB2Gray&) {} }; template struct RGB2Gray : unary_function @@ -478,8 +473,8 @@ namespace cv { namespace gpu { namespace device { return RGB2GrayConvert(src); } - __device__ __forceinline__ RGB2Gray() : unary_function(){} - __device__ __forceinline__ RGB2Gray(const RGB2Gray& other_) : unary_function(){} + __host__ __device__ __forceinline__ RGB2Gray() {} + __host__ __device__ __forceinline__ RGB2Gray(const RGB2Gray&) {} }; } @@ -529,10 +524,8 @@ namespace cv { namespace gpu { namespace device RGB2YUVConvert(&src.x, dst); return dst; } - __device__ __forceinline__ RGB2YUV() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ RGB2YUV(const RGB2YUV& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ RGB2YUV() {} + __host__ __device__ __forceinline__ RGB2YUV(const RGB2YUV&) {} }; } @@ -609,10 +602,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ YUV2RGB() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ YUV2RGB(const YUV2RGB& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ YUV2RGB() {} + __host__ __device__ __forceinline__ YUV2RGB(const YUV2RGB&) {} }; template struct YUV2RGB : unary_function @@ -621,8 +612,8 @@ namespace cv { namespace gpu { namespace device { return YUV2RGBConvert(src); } - __device__ __forceinline__ YUV2RGB() : unary_function(){} - __device__ __forceinline__ YUV2RGB(const YUV2RGB& other_) : unary_function(){} + __host__ __device__ __forceinline__ YUV2RGB() {} + __host__ __device__ __forceinline__ YUV2RGB(const YUV2RGB&) {} }; } @@ -689,10 +680,8 @@ namespace cv { namespace gpu { namespace device RGB2YCrCbConvert(&src.x, dst); return dst; } - __device__ __forceinline__ RGB2YCrCb() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ RGB2YCrCb(const RGB2YCrCb& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ RGB2YCrCb() {} + __host__ __device__ __forceinline__ RGB2YCrCb(const RGB2YCrCb&) {} }; template struct RGB2YCrCb : unary_function @@ -702,8 +691,8 @@ namespace cv { namespace gpu { namespace device return RGB2YCrCbConvert(src); } - __device__ __forceinline__ RGB2YCrCb() : unary_function(){} - __device__ __forceinline__ RGB2YCrCb(const RGB2YCrCb& other_) : unary_function(){} + __host__ __device__ __forceinline__ RGB2YCrCb() {} + __host__ __device__ __forceinline__ RGB2YCrCb(const RGB2YCrCb&) {} }; } @@ -771,10 +760,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ YCrCb2RGB() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ YCrCb2RGB(const YCrCb2RGB& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ YCrCb2RGB() {} + __host__ __device__ __forceinline__ YCrCb2RGB(const YCrCb2RGB&) {} }; template struct YCrCb2RGB : unary_function @@ -783,8 +770,8 @@ namespace cv { namespace gpu { namespace device { return YCrCb2RGBConvert(src); } - __device__ __forceinline__ YCrCb2RGB() : unary_function(){} - __device__ __forceinline__ YCrCb2RGB(const YCrCb2RGB& other_) : unary_function(){} + __host__ __device__ __forceinline__ YCrCb2RGB() {} + __host__ __device__ __forceinline__ YCrCb2RGB(const YCrCb2RGB&) {} }; } @@ -849,10 +836,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2XYZ() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ RGB2XYZ(const RGB2XYZ& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ RGB2XYZ() {} + __host__ __device__ __forceinline__ RGB2XYZ(const RGB2XYZ&) {} }; template struct RGB2XYZ : unary_function @@ -861,8 +846,8 @@ namespace cv { namespace gpu { namespace device { return RGB2XYZConvert(src); } - __device__ __forceinline__ RGB2XYZ() : unary_function(){} - __device__ __forceinline__ RGB2XYZ(const RGB2XYZ& other_) : unary_function(){} + __host__ __device__ __forceinline__ RGB2XYZ() {} + __host__ __device__ __forceinline__ RGB2XYZ(const RGB2XYZ&) {} }; } @@ -926,10 +911,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ XYZ2RGB() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ XYZ2RGB(const XYZ2RGB& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ XYZ2RGB() {} + __host__ __device__ __forceinline__ XYZ2RGB(const XYZ2RGB&) {} }; template struct XYZ2RGB : unary_function @@ -938,8 +921,8 @@ namespace cv { namespace gpu { namespace device { return XYZ2RGBConvert(src); } - __device__ __forceinline__ XYZ2RGB() : unary_function(){} - __device__ __forceinline__ XYZ2RGB(const XYZ2RGB& other_) : unary_function(){} + __host__ __device__ __forceinline__ XYZ2RGB() {} + __host__ __device__ __forceinline__ XYZ2RGB(const XYZ2RGB&) {} }; } @@ -1066,10 +1049,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2HSV() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ RGB2HSV(const RGB2HSV& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ RGB2HSV() {} + __host__ __device__ __forceinline__ RGB2HSV(const RGB2HSV&) {} }; template struct RGB2HSV : unary_function @@ -1078,8 +1059,8 @@ namespace cv { namespace gpu { namespace device { return RGB2HSVConvert(src); } - __device__ __forceinline__ RGB2HSV():unary_function(){} - __device__ __forceinline__ RGB2HSV(const RGB2HSV& other_):unary_function(){} + __host__ __device__ __forceinline__ RGB2HSV() {} + __host__ __device__ __forceinline__ RGB2HSV(const RGB2HSV&) {} }; } @@ -1208,10 +1189,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ HSV2RGB() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ HSV2RGB(const HSV2RGB& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ HSV2RGB() {} + __host__ __device__ __forceinline__ HSV2RGB(const HSV2RGB&) {} }; template struct HSV2RGB : unary_function @@ -1220,8 +1199,8 @@ namespace cv { namespace gpu { namespace device { return HSV2RGBConvert(src); } - __device__ __forceinline__ HSV2RGB():unary_function(){} - __device__ __forceinline__ HSV2RGB(const HSV2RGB& other_):unary_function(){} + __host__ __device__ __forceinline__ HSV2RGB() {} + __host__ __device__ __forceinline__ HSV2RGB(const HSV2RGB&) {} }; } @@ -1343,10 +1322,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2HLS() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ RGB2HLS(const RGB2HLS& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ RGB2HLS() {} + __host__ __device__ __forceinline__ RGB2HLS(const RGB2HLS&) {} }; template struct RGB2HLS : unary_function @@ -1355,8 +1332,8 @@ namespace cv { namespace gpu { namespace device { return RGB2HLSConvert(src); } - __device__ __forceinline__ RGB2HLS() : unary_function(){} - __device__ __forceinline__ RGB2HLS(const RGB2HLS& other_) : unary_function(){} + __host__ __device__ __forceinline__ RGB2HLS() {} + __host__ __device__ __forceinline__ RGB2HLS(const RGB2HLS&) {} }; } @@ -1485,10 +1462,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ HLS2RGB() - : unary_function::vec_type, typename TypeVec::vec_type>(){} - __device__ __forceinline__ HLS2RGB(const HLS2RGB& other_) - : unary_function::vec_type, typename TypeVec::vec_type>(){} + __host__ __device__ __forceinline__ HLS2RGB() {} + __host__ __device__ __forceinline__ HLS2RGB(const HLS2RGB&) {} }; template struct HLS2RGB : unary_function @@ -1497,8 +1472,8 @@ namespace cv { namespace gpu { namespace device { return HLS2RGBConvert(src); } - __device__ __forceinline__ HLS2RGB() : unary_function(){} - __device__ __forceinline__ HLS2RGB(const HLS2RGB& other_) : unary_function(){} + __host__ __device__ __forceinline__ HLS2RGB() {} + __host__ __device__ __forceinline__ HLS2RGB(const HLS2RGB&) {} }; } @@ -1651,8 +1626,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2Lab() {} - __device__ __forceinline__ RGB2Lab(const RGB2Lab& other_) {} + __host__ __device__ __forceinline__ RGB2Lab() {} + __host__ __device__ __forceinline__ RGB2Lab(const RGB2Lab&) {} }; template struct RGB2Lab @@ -1666,8 +1641,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2Lab() {} - __device__ __forceinline__ RGB2Lab(const RGB2Lab& other_) {} + __host__ __device__ __forceinline__ RGB2Lab() {} + __host__ __device__ __forceinline__ RGB2Lab(const RGB2Lab&) {} }; } @@ -1764,8 +1739,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ Lab2RGB() {} - __device__ __forceinline__ Lab2RGB(const Lab2RGB& other_) {} + __host__ __device__ __forceinline__ Lab2RGB() {} + __host__ __device__ __forceinline__ Lab2RGB(const Lab2RGB&) {} }; template struct Lab2RGB @@ -1779,8 +1754,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ Lab2RGB() {} - __device__ __forceinline__ Lab2RGB(const Lab2RGB& other_) {} + __host__ __device__ __forceinline__ Lab2RGB() {} + __host__ __device__ __forceinline__ Lab2RGB(const Lab2RGB&) {} }; } @@ -1863,8 +1838,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2Luv() {} - __device__ __forceinline__ RGB2Luv(const RGB2Luv& other_) {} + __host__ __device__ __forceinline__ RGB2Luv() {} + __host__ __device__ __forceinline__ RGB2Luv(const RGB2Luv&) {} }; template struct RGB2Luv @@ -1878,8 +1853,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ RGB2Luv() {} - __device__ __forceinline__ RGB2Luv(const RGB2Luv& other_) {} + __host__ __device__ __forceinline__ RGB2Luv() {} + __host__ __device__ __forceinline__ RGB2Luv(const RGB2Luv&) {} }; } @@ -1964,8 +1939,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ Luv2RGB() {} - __device__ __forceinline__ Luv2RGB(const Luv2RGB& other_) {} + __host__ __device__ __forceinline__ Luv2RGB() {} + __host__ __device__ __forceinline__ Luv2RGB(const Luv2RGB&) {} }; template struct Luv2RGB @@ -1979,8 +1954,8 @@ namespace cv { namespace gpu { namespace device return dst; } - __device__ __forceinline__ Luv2RGB() {} - __device__ __forceinline__ Luv2RGB(const Luv2RGB& other_) {} + __host__ __device__ __forceinline__ Luv2RGB() {} + __host__ __device__ __forceinline__ Luv2RGB(const Luv2RGB&) {} }; } diff --git a/modules/gpu/include/opencv2/gpu/device/functional.hpp b/modules/gpu/include/opencv2/gpu/device/functional.hpp index 6064e8e..db26473 100644 --- a/modules/gpu/include/opencv2/gpu/device/functional.hpp +++ b/modules/gpu/include/opencv2/gpu/device/functional.hpp @@ -63,8 +63,8 @@ namespace cv { namespace gpu { namespace device { return a + b; } - __device__ __forceinline__ plus(const plus& other):binary_function(){} - __device__ __forceinline__ plus():binary_function(){} + __host__ __device__ __forceinline__ plus() {} + __host__ __device__ __forceinline__ plus(const plus&) {} }; template struct minus : binary_function @@ -74,8 +74,8 @@ namespace cv { namespace gpu { namespace device { return a - b; } - __device__ __forceinline__ minus(const minus& other):binary_function(){} - __device__ __forceinline__ minus():binary_function(){} + __host__ __device__ __forceinline__ minus() {} + __host__ __device__ __forceinline__ minus(const minus&) {} }; template struct multiplies : binary_function @@ -85,8 +85,8 @@ namespace cv { namespace gpu { namespace device { return a * b; } - __device__ __forceinline__ multiplies(const multiplies& other):binary_function(){} - __device__ __forceinline__ multiplies():binary_function(){} + __host__ __device__ __forceinline__ multiplies() {} + __host__ __device__ __forceinline__ multiplies(const multiplies&) {} }; template struct divides : binary_function @@ -96,8 +96,8 @@ namespace cv { namespace gpu { namespace device { return a / b; } - __device__ __forceinline__ divides(const divides& other):binary_function(){} - __device__ __forceinline__ divides():binary_function(){} + __host__ __device__ __forceinline__ divides() {} + __host__ __device__ __forceinline__ divides(const divides&) {} }; template struct modulus : binary_function @@ -107,8 +107,8 @@ namespace cv { namespace gpu { namespace device { return a % b; } - __device__ __forceinline__ modulus(const modulus& other):binary_function(){} - __device__ __forceinline__ modulus():binary_function(){} + __host__ __device__ __forceinline__ modulus() {} + __host__ __device__ __forceinline__ modulus(const modulus&) {} }; template struct negate : unary_function @@ -117,8 +117,8 @@ namespace cv { namespace gpu { namespace device { return -a; } - __device__ __forceinline__ negate(const negate& other):unary_function(){} - __device__ __forceinline__ negate():unary_function(){} + __host__ __device__ __forceinline__ negate() {} + __host__ __device__ __forceinline__ negate(const negate&) {} }; // Comparison Operations @@ -129,8 +129,8 @@ namespace cv { namespace gpu { namespace device { return a == b; } - __device__ __forceinline__ equal_to(const equal_to& other):binary_function(){} - __device__ __forceinline__ equal_to():binary_function(){} + __host__ __device__ __forceinline__ equal_to() {} + __host__ __device__ __forceinline__ equal_to(const equal_to&) {} }; template struct not_equal_to : binary_function @@ -140,8 +140,8 @@ namespace cv { namespace gpu { namespace device { return a != b; } - __device__ __forceinline__ not_equal_to(const not_equal_to& other):binary_function(){} - __device__ __forceinline__ not_equal_to():binary_function(){} + __host__ __device__ __forceinline__ not_equal_to() {} + __host__ __device__ __forceinline__ not_equal_to(const not_equal_to&) {} }; template struct greater : binary_function @@ -151,8 +151,8 @@ namespace cv { namespace gpu { namespace device { return a > b; } - __device__ __forceinline__ greater(const greater& other):binary_function(){} - __device__ __forceinline__ greater():binary_function(){} + __host__ __device__ __forceinline__ greater() {} + __host__ __device__ __forceinline__ greater(const greater&) {} }; template struct less : binary_function @@ -162,8 +162,8 @@ namespace cv { namespace gpu { namespace device { return a < b; } - __device__ __forceinline__ less(const less& other):binary_function(){} - __device__ __forceinline__ less():binary_function(){} + __host__ __device__ __forceinline__ less() {} + __host__ __device__ __forceinline__ less(const less&) {} }; template struct greater_equal : binary_function @@ -173,8 +173,8 @@ namespace cv { namespace gpu { namespace device { return a >= b; } - __device__ __forceinline__ greater_equal(const greater_equal& other):binary_function(){} - __device__ __forceinline__ greater_equal():binary_function(){} + __host__ __device__ __forceinline__ greater_equal() {} + __host__ __device__ __forceinline__ greater_equal(const greater_equal&) {} }; template struct less_equal : binary_function @@ -184,8 +184,8 @@ namespace cv { namespace gpu { namespace device { return a <= b; } - __device__ __forceinline__ less_equal(const less_equal& other):binary_function(){} - __device__ __forceinline__ less_equal():binary_function(){} + __host__ __device__ __forceinline__ less_equal() {} + __host__ __device__ __forceinline__ less_equal(const less_equal&) {} }; // Logical Operations @@ -196,8 +196,8 @@ namespace cv { namespace gpu { namespace device { return a && b; } - __device__ __forceinline__ logical_and(const logical_and& other):binary_function(){} - __device__ __forceinline__ logical_and():binary_function(){} + __host__ __device__ __forceinline__ logical_and() {} + __host__ __device__ __forceinline__ logical_and(const logical_and&) {} }; template struct logical_or : binary_function @@ -207,8 +207,8 @@ namespace cv { namespace gpu { namespace device { return a || b; } - __device__ __forceinline__ logical_or(const logical_or& other):binary_function(){} - __device__ __forceinline__ logical_or():binary_function(){} + __host__ __device__ __forceinline__ logical_or() {} + __host__ __device__ __forceinline__ logical_or(const logical_or&) {} }; template struct logical_not : unary_function @@ -217,8 +217,8 @@ namespace cv { namespace gpu { namespace device { return !a; } - __device__ __forceinline__ logical_not(const logical_not& other):unary_function(){} - __device__ __forceinline__ logical_not():unary_function(){} + __host__ __device__ __forceinline__ logical_not() {} + __host__ __device__ __forceinline__ logical_not(const logical_not&) {} }; // Bitwise Operations @@ -229,8 +229,8 @@ namespace cv { namespace gpu { namespace device { return a & b; } - __device__ __forceinline__ bit_and(const bit_and& other):binary_function(){} - __device__ __forceinline__ bit_and():binary_function(){} + __host__ __device__ __forceinline__ bit_and() {} + __host__ __device__ __forceinline__ bit_and(const bit_and&) {} }; template struct bit_or : binary_function @@ -240,8 +240,8 @@ namespace cv { namespace gpu { namespace device { return a | b; } - __device__ __forceinline__ bit_or(const bit_or& other):binary_function(){} - __device__ __forceinline__ bit_or():binary_function(){} + __host__ __device__ __forceinline__ bit_or() {} + __host__ __device__ __forceinline__ bit_or(const bit_or&) {} }; template struct bit_xor : binary_function @@ -251,8 +251,8 @@ namespace cv { namespace gpu { namespace device { return a ^ b; } - __device__ __forceinline__ bit_xor(const bit_xor& other):binary_function(){} - __device__ __forceinline__ bit_xor():binary_function(){} + __host__ __device__ __forceinline__ bit_xor() {} + __host__ __device__ __forceinline__ bit_xor(const bit_xor&) {} }; template struct bit_not : unary_function @@ -261,8 +261,8 @@ namespace cv { namespace gpu { namespace device { return ~v; } - __device__ __forceinline__ bit_not(const bit_not& other):unary_function(){} - __device__ __forceinline__ bit_not():unary_function(){} + __host__ __device__ __forceinline__ bit_not() {} + __host__ __device__ __forceinline__ bit_not(const bit_not&) {} }; // Generalized Identity Operations @@ -272,8 +272,8 @@ namespace cv { namespace gpu { namespace device { return x; } - __device__ __forceinline__ identity(const identity& other):unary_function(){} - __device__ __forceinline__ identity():unary_function(){} + __host__ __device__ __forceinline__ identity() {} + __host__ __device__ __forceinline__ identity(const identity&) {} }; template struct project1st : binary_function @@ -282,8 +282,8 @@ namespace cv { namespace gpu { namespace device { return lhs; } - __device__ __forceinline__ project1st(const project1st& other):binary_function(){} - __device__ __forceinline__ project1st():binary_function(){} + __host__ __device__ __forceinline__ project1st() {} + __host__ __device__ __forceinline__ project1st(const project1st&) {} }; template struct project2nd : binary_function @@ -292,8 +292,8 @@ namespace cv { namespace gpu { namespace device { return rhs; } - __device__ __forceinline__ project2nd(const project2nd& other):binary_function(){} - __device__ __forceinline__ project2nd():binary_function(){} + __host__ __device__ __forceinline__ project2nd() {} + __host__ __device__ __forceinline__ project2nd(const project2nd&) {} }; // Min/Max Operations @@ -302,8 +302,8 @@ namespace cv { namespace gpu { namespace device template <> struct name : binary_function \ { \ __device__ __forceinline__ type operator()(type lhs, type rhs) const {return op(lhs, rhs);} \ - __device__ __forceinline__ name() {}\ - __device__ __forceinline__ name(const name&) {}\ + __host__ __device__ __forceinline__ name() {}\ + __host__ __device__ __forceinline__ name(const name&) {}\ }; template struct maximum : binary_function @@ -312,8 +312,8 @@ namespace cv { namespace gpu { namespace device { return max(lhs, rhs); } - __device__ __forceinline__ maximum() {} - __device__ __forceinline__ maximum(const maximum&) {} + __host__ __device__ __forceinline__ maximum() {} + __host__ __device__ __forceinline__ maximum(const maximum&) {} }; OPENCV_GPU_IMPLEMENT_MINMAX(maximum, uchar, ::max) @@ -332,8 +332,8 @@ namespace cv { namespace gpu { namespace device { return min(lhs, rhs); } - __device__ __forceinline__ minimum() {} - __device__ __forceinline__ minimum(const minimum&) {} + __host__ __device__ __forceinline__ minimum() {} + __host__ __device__ __forceinline__ minimum(const minimum&) {} }; OPENCV_GPU_IMPLEMENT_MINMAX(minimum, uchar, ::min) @@ -349,7 +349,6 @@ namespace cv { namespace gpu { namespace device #undef OPENCV_GPU_IMPLEMENT_MINMAX // Math functions -///bound========================================= template struct abs_func : unary_function { @@ -358,8 +357,8 @@ namespace cv { namespace gpu { namespace device return abs(x); } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -368,8 +367,8 @@ namespace cv { namespace gpu { namespace device return x; } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -378,8 +377,8 @@ namespace cv { namespace gpu { namespace device return ::abs((int)x); } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -388,8 +387,8 @@ namespace cv { namespace gpu { namespace device return ::abs((int)x); } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -398,8 +397,8 @@ namespace cv { namespace gpu { namespace device return x; } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -408,8 +407,8 @@ namespace cv { namespace gpu { namespace device return ::abs((int)x); } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -418,8 +417,8 @@ namespace cv { namespace gpu { namespace device return x; } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -428,8 +427,8 @@ namespace cv { namespace gpu { namespace device return ::abs(x); } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -438,8 +437,8 @@ namespace cv { namespace gpu { namespace device return ::fabsf(x); } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; template <> struct abs_func : unary_function { @@ -448,8 +447,8 @@ namespace cv { namespace gpu { namespace device return ::fabs(x); } - __device__ __forceinline__ abs_func() {} - __device__ __forceinline__ abs_func(const abs_func&) {} + __host__ __device__ __forceinline__ abs_func() {} + __host__ __device__ __forceinline__ abs_func(const abs_func&) {} }; #define OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(name, func) \ @@ -459,8 +458,8 @@ namespace cv { namespace gpu { namespace device { \ return func ## f(v); \ } \ - __device__ __forceinline__ name ## _func() {} \ - __device__ __forceinline__ name ## _func(const name ## _func&) {} \ + __host__ __device__ __forceinline__ name ## _func() {} \ + __host__ __device__ __forceinline__ name ## _func(const name ## _func&) {} \ }; \ template <> struct name ## _func : unary_function \ { \ @@ -468,8 +467,8 @@ namespace cv { namespace gpu { namespace device { \ return func(v); \ } \ - __device__ __forceinline__ name ## _func() {} \ - __device__ __forceinline__ name ## _func(const name ## _func&) {} \ + __host__ __device__ __forceinline__ name ## _func() {} \ + __host__ __device__ __forceinline__ name ## _func(const name ## _func&) {} \ }; #define OPENCV_GPU_IMPLEMENT_BIN_FUNCTOR(name, func) \ @@ -479,6 +478,8 @@ namespace cv { namespace gpu { namespace device { \ return func ## f(v1, v2); \ } \ + __host__ __device__ __forceinline__ name ## _func() {} \ + __host__ __device__ __forceinline__ name ## _func(const name ## _func&) {} \ }; \ template <> struct name ## _func : binary_function \ { \ @@ -486,6 +487,8 @@ namespace cv { namespace gpu { namespace device { \ return func(v1, v2); \ } \ + __host__ __device__ __forceinline__ name ## _func() {} \ + __host__ __device__ __forceinline__ name ## _func(const name ## _func&) {} \ }; OPENCV_GPU_IMPLEMENT_UN_FUNCTOR(sqrt, ::sqrt) @@ -522,8 +525,8 @@ namespace cv { namespace gpu { namespace device { return src1 * src1 + src2 * src2; } - __device__ __forceinline__ hypot_sqr_func(const hypot_sqr_func& other) : binary_function(){} - __device__ __forceinline__ hypot_sqr_func() : binary_function(){} + __host__ __device__ __forceinline__ hypot_sqr_func() {} + __host__ __device__ __forceinline__ hypot_sqr_func(const hypot_sqr_func&) {} }; // Saturate Cast Functor @@ -533,8 +536,8 @@ namespace cv { namespace gpu { namespace device { return saturate_cast(v); } - __device__ __forceinline__ saturate_cast_func(const saturate_cast_func& other):unary_function(){} - __device__ __forceinline__ saturate_cast_func():unary_function(){} + __host__ __device__ __forceinline__ saturate_cast_func() {} + __host__ __device__ __forceinline__ saturate_cast_func(const saturate_cast_func&) {} }; // Threshold Functors @@ -547,10 +550,9 @@ namespace cv { namespace gpu { namespace device return (src > thresh) * maxVal; } - __device__ __forceinline__ thresh_binary_func(const thresh_binary_func& other) - : unary_function(), thresh(other.thresh), maxVal(other.maxVal){} - - __device__ __forceinline__ thresh_binary_func():unary_function(){} + __host__ __device__ __forceinline__ thresh_binary_func() {} + __host__ __device__ __forceinline__ thresh_binary_func(const thresh_binary_func& other) + : thresh(other.thresh), maxVal(other.maxVal) {} const T thresh; const T maxVal; @@ -565,10 +567,9 @@ namespace cv { namespace gpu { namespace device return (src <= thresh) * maxVal; } - __device__ __forceinline__ thresh_binary_inv_func(const thresh_binary_inv_func& other) - : unary_function(), thresh(other.thresh), maxVal(other.maxVal){} - - __device__ __forceinline__ thresh_binary_inv_func():unary_function(){} + __host__ __device__ __forceinline__ thresh_binary_inv_func() {} + __host__ __device__ __forceinline__ thresh_binary_inv_func(const thresh_binary_inv_func& other) + : thresh(other.thresh), maxVal(other.maxVal) {} const T thresh; const T maxVal; @@ -583,10 +584,9 @@ namespace cv { namespace gpu { namespace device return minimum()(src, thresh); } - __device__ __forceinline__ thresh_trunc_func(const thresh_trunc_func& other) - : unary_function(), thresh(other.thresh){} - - __device__ __forceinline__ thresh_trunc_func():unary_function(){} + __host__ __device__ __forceinline__ thresh_trunc_func() {} + __host__ __device__ __forceinline__ thresh_trunc_func(const thresh_trunc_func& other) + : thresh(other.thresh) {} const T thresh; }; @@ -599,10 +599,10 @@ namespace cv { namespace gpu { namespace device { return (src > thresh) * src; } - __device__ __forceinline__ thresh_to_zero_func(const thresh_to_zero_func& other) - : unary_function(), thresh(other.thresh){} - __device__ __forceinline__ thresh_to_zero_func():unary_function(){} + __host__ __device__ __forceinline__ thresh_to_zero_func() {} + __host__ __device__ __forceinline__ thresh_to_zero_func(const thresh_to_zero_func& other) + : thresh(other.thresh) {} const T thresh; }; @@ -615,14 +615,14 @@ namespace cv { namespace gpu { namespace device { return (src <= thresh) * src; } - __device__ __forceinline__ thresh_to_zero_inv_func(const thresh_to_zero_inv_func& other) - : unary_function(), thresh(other.thresh){} - __device__ __forceinline__ thresh_to_zero_inv_func():unary_function(){} + __host__ __device__ __forceinline__ thresh_to_zero_inv_func() {} + __host__ __device__ __forceinline__ thresh_to_zero_inv_func(const thresh_to_zero_inv_func& other) + : thresh(other.thresh) {} const T thresh; }; -//bound!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!! ============> + // Function Object Adaptors template struct unary_negate : unary_function { @@ -633,8 +633,8 @@ namespace cv { namespace gpu { namespace device return !pred(x); } - __device__ __forceinline__ unary_negate(const unary_negate& other) : unary_function(){} - __device__ __forceinline__ unary_negate() : unary_function(){} + __host__ __device__ __forceinline__ unary_negate() {} + __host__ __device__ __forceinline__ unary_negate(const unary_negate& other) : pred(other.pred) {} const Predicate pred; }; @@ -653,11 +653,9 @@ namespace cv { namespace gpu { namespace device { return !pred(x,y); } - __device__ __forceinline__ binary_negate(const binary_negate& other) - : binary_function(){} - __device__ __forceinline__ binary_negate() : - binary_function(){} + __host__ __device__ __forceinline__ binary_negate() {} + __host__ __device__ __forceinline__ binary_negate(const binary_negate& other) : pred(other.pred) {} const Predicate pred; }; @@ -676,8 +674,8 @@ namespace cv { namespace gpu { namespace device return op(arg1, a); } - __device__ __forceinline__ binder1st(const binder1st& other) : - unary_function(){} + __host__ __device__ __forceinline__ binder1st() {} + __host__ __device__ __forceinline__ binder1st(const binder1st& other) : op(other.op), arg1(other.arg1) {} const Op op; const typename Op::first_argument_type arg1; @@ -697,8 +695,8 @@ namespace cv { namespace gpu { namespace device return op(a, arg2); } - __device__ __forceinline__ binder2nd(const binder2nd& other) : - unary_function(), op(other.op), arg2(other.arg2){} + __host__ __device__ __forceinline__ binder2nd() {} + __host__ __device__ __forceinline__ binder2nd(const binder2nd& other) : op(other.op), arg2(other.arg2) {} const Op op; const typename Op::second_argument_type arg2; diff --git a/modules/gpu/include/opencv2/gpu/device/utility.hpp b/modules/gpu/include/opencv2/gpu/device/utility.hpp index 83eaaa2..85e81ac 100644 --- a/modules/gpu/include/opencv2/gpu/device/utility.hpp +++ b/modules/gpu/include/opencv2/gpu/device/utility.hpp @@ -124,8 +124,8 @@ namespace cv { namespace gpu { namespace device struct WithOutMask { - __device__ __forceinline__ WithOutMask(){} - __device__ __forceinline__ WithOutMask(const WithOutMask& mask){} + __host__ __device__ __forceinline__ WithOutMask(){} + __host__ __device__ __forceinline__ WithOutMask(const WithOutMask&){} __device__ __forceinline__ void next() const { diff --git a/modules/gpu/src/cuda/calib3d.cu b/modules/gpu/src/cuda/calib3d.cu index 0fd482c..f29471f 100644 --- a/modules/gpu/src/cuda/calib3d.cu +++ b/modules/gpu/src/cuda/calib3d.cu @@ -67,8 +67,8 @@ namespace cv { namespace gpu { namespace device crot1.x * p.x + crot1.y * p.y + crot1.z * p.z + ctransl.y, crot2.x * p.x + crot2.y * p.y + crot2.z * p.z + ctransl.z); } - __device__ __forceinline__ TransformOp() {} - __device__ __forceinline__ TransformOp(const TransformOp&) {} + __host__ __device__ __forceinline__ TransformOp() {} + __host__ __device__ __forceinline__ TransformOp(const TransformOp&) {} }; void call(const PtrStepSz src, const float* rot, @@ -106,8 +106,8 @@ namespace cv { namespace gpu { namespace device (cproj0.x * t.x + cproj0.y * t.y) / t.z + cproj0.z, (cproj1.x * t.x + cproj1.y * t.y) / t.z + cproj1.z); } - __device__ __forceinline__ ProjectOp() {} - __device__ __forceinline__ ProjectOp(const ProjectOp&) {} + __host__ __device__ __forceinline__ ProjectOp() {} + __host__ __device__ __forceinline__ ProjectOp(const ProjectOp&) {} }; void call(const PtrStepSz src, const float* rot, diff --git a/modules/gpu/src/cuda/canny.cu b/modules/gpu/src/cuda/canny.cu index 1afcddc..aab922f 100644 --- a/modules/gpu/src/cuda/canny.cu +++ b/modules/gpu/src/cuda/canny.cu @@ -62,8 +62,8 @@ namespace canny return ::abs(x) + ::abs(y); } - __device__ __forceinline__ L1() {} - __device__ __forceinline__ L1(const L1&) {} + __host__ __device__ __forceinline__ L1() {} + __host__ __device__ __forceinline__ L1(const L1&) {} }; struct L2 : binary_function { @@ -72,8 +72,8 @@ namespace canny return ::sqrtf(x * x + y * y); } - __device__ __forceinline__ L2() {} - __device__ __forceinline__ L2(const L2&) {} + __host__ __device__ __forceinline__ L2() {} + __host__ __device__ __forceinline__ L2(const L2&) {} }; } @@ -470,8 +470,8 @@ namespace canny return (uchar)(-(e >> 1)); } - __device__ __forceinline__ GetEdges() {} - __device__ __forceinline__ GetEdges(const GetEdges&) {} + __host__ __device__ __forceinline__ GetEdges() {} + __host__ __device__ __forceinline__ GetEdges(const GetEdges&) {} }; } diff --git a/modules/gpu/src/cuda/element_operations.cu b/modules/gpu/src/cuda/element_operations.cu index e9397e5..876d4ad 100644 --- a/modules/gpu/src/cuda/element_operations.cu +++ b/modules/gpu/src/cuda/element_operations.cu @@ -162,8 +162,8 @@ namespace arithm return vadd4(a, b); } - __device__ __forceinline__ VAdd4() {} - __device__ __forceinline__ VAdd4(const VAdd4& other) {} + __host__ __device__ __forceinline__ VAdd4() {} + __host__ __device__ __forceinline__ VAdd4(const VAdd4&) {} }; //////////////////////////////////// @@ -175,8 +175,8 @@ namespace arithm return vadd2(a, b); } - __device__ __forceinline__ VAdd2() {} - __device__ __forceinline__ VAdd2(const VAdd2& other) {} + __host__ __device__ __forceinline__ VAdd2() {} + __host__ __device__ __forceinline__ VAdd2(const VAdd2&) {} }; //////////////////////////////////// @@ -188,8 +188,8 @@ namespace arithm return saturate_cast(a + b); } - __device__ __forceinline__ AddMat() {} - __device__ __forceinline__ AddMat(const AddMat& other) {} + __host__ __device__ __forceinline__ AddMat() {} + __host__ __device__ __forceinline__ AddMat(const AddMat&) {} }; } @@ -397,8 +397,8 @@ namespace arithm return vsub4(a, b); } - __device__ __forceinline__ VSub4() {} - __device__ __forceinline__ VSub4(const VSub4& other) {} + __host__ __device__ __forceinline__ VSub4() {} + __host__ __device__ __forceinline__ VSub4(const VSub4&) {} }; //////////////////////////////////// @@ -410,8 +410,8 @@ namespace arithm return vsub2(a, b); } - __device__ __forceinline__ VSub2() {} - __device__ __forceinline__ VSub2(const VSub2& other) {} + __host__ __device__ __forceinline__ VSub2() {} + __host__ __device__ __forceinline__ VSub2(const VSub2&) {} }; //////////////////////////////////// @@ -423,8 +423,8 @@ namespace arithm return saturate_cast(a - b); } - __device__ __forceinline__ SubMat() {} - __device__ __forceinline__ SubMat(const SubMat& other) {} + __host__ __device__ __forceinline__ SubMat() {} + __host__ __device__ __forceinline__ SubMat(const SubMat&) {} }; } @@ -617,8 +617,8 @@ namespace arithm return res; } - __device__ __forceinline__ Mul_8uc4_32f() {} - __device__ __forceinline__ Mul_8uc4_32f(const Mul_8uc4_32f& other) {} + __host__ __device__ __forceinline__ Mul_8uc4_32f() {} + __host__ __device__ __forceinline__ Mul_8uc4_32f(const Mul_8uc4_32f&) {} }; struct Mul_16sc4_32f : binary_function @@ -629,8 +629,8 @@ namespace arithm saturate_cast(a.z * b), saturate_cast(a.w * b)); } - __device__ __forceinline__ Mul_16sc4_32f() {} - __device__ __forceinline__ Mul_16sc4_32f(const Mul_16sc4_32f& other) {} + __host__ __device__ __forceinline__ Mul_16sc4_32f() {} + __host__ __device__ __forceinline__ Mul_16sc4_32f(const Mul_16sc4_32f&) {} }; template struct Mul : binary_function @@ -640,8 +640,8 @@ namespace arithm return saturate_cast(a * b); } - __device__ __forceinline__ Mul() {} - __device__ __forceinline__ Mul(const Mul& other) {} + __host__ __device__ __forceinline__ Mul() {} + __host__ __device__ __forceinline__ Mul(const Mul&) {} }; template struct MulScale : binary_function @@ -888,8 +888,8 @@ namespace arithm return b != 0 ? saturate_cast(a / b) : 0; } - __device__ __forceinline__ Div() {} - __device__ __forceinline__ Div(const Div& other) {} + __host__ __device__ __forceinline__ Div() {} + __host__ __device__ __forceinline__ Div(const Div&) {} }; template struct Div : binary_function { @@ -898,8 +898,8 @@ namespace arithm return b != 0 ? static_cast(a) / b : 0; } - __device__ __forceinline__ Div() {} - __device__ __forceinline__ Div(const Div& other) {} + __host__ __device__ __forceinline__ Div() {} + __host__ __device__ __forceinline__ Div(const Div&) {} }; template struct Div : binary_function { @@ -908,8 +908,8 @@ namespace arithm return b != 0 ? static_cast(a) / b : 0; } - __device__ __forceinline__ Div() {} - __device__ __forceinline__ Div(const Div& other) {} + __host__ __device__ __forceinline__ Div() {} + __host__ __device__ __forceinline__ Div(const Div&) {} }; template struct DivScale : binary_function @@ -1196,8 +1196,8 @@ namespace arithm return vabsdiff4(a, b); } - __device__ __forceinline__ VAbsDiff4() {} - __device__ __forceinline__ VAbsDiff4(const VAbsDiff4& other) {} + __host__ __device__ __forceinline__ VAbsDiff4() {} + __host__ __device__ __forceinline__ VAbsDiff4(const VAbsDiff4&) {} }; //////////////////////////////////// @@ -1209,8 +1209,8 @@ namespace arithm return vabsdiff2(a, b); } - __device__ __forceinline__ VAbsDiff2() {} - __device__ __forceinline__ VAbsDiff2(const VAbsDiff2& other) {} + __host__ __device__ __forceinline__ VAbsDiff2() {} + __host__ __device__ __forceinline__ VAbsDiff2(const VAbsDiff2&) {} }; //////////////////////////////////// @@ -1235,8 +1235,8 @@ namespace arithm return saturate_cast(_abs(a - b)); } - __device__ __forceinline__ AbsDiffMat() {} - __device__ __forceinline__ AbsDiffMat(const AbsDiffMat& other) {} + __host__ __device__ __forceinline__ AbsDiffMat() {} + __host__ __device__ __forceinline__ AbsDiffMat(const AbsDiffMat&) {} }; } @@ -1370,8 +1370,8 @@ namespace arithm return saturate_cast(x * x); } - __device__ __forceinline__ Sqr() {} - __device__ __forceinline__ Sqr(const Sqr& other) {} + __host__ __device__ __forceinline__ Sqr() {} + __host__ __device__ __forceinline__ Sqr(const Sqr&) {} }; } @@ -1466,8 +1466,8 @@ namespace arithm return saturate_cast(f(x)); } - __device__ __forceinline__ Exp() {} - __device__ __forceinline__ Exp(const Exp& other) {} + __host__ __device__ __forceinline__ Exp() {} + __host__ __device__ __forceinline__ Exp(const Exp&) {} }; } @@ -1507,8 +1507,8 @@ namespace arithm return vcmpeq4(a, b); } - __device__ __forceinline__ VCmpEq4() {} - __device__ __forceinline__ VCmpEq4(const VCmpEq4& other) {} + __host__ __device__ __forceinline__ VCmpEq4() {} + __host__ __device__ __forceinline__ VCmpEq4(const VCmpEq4&) {} }; struct VCmpNe4 : binary_function { @@ -1517,8 +1517,8 @@ namespace arithm return vcmpne4(a, b); } - __device__ __forceinline__ VCmpNe4() {} - __device__ __forceinline__ VCmpNe4(const VCmpNe4& other) {} + __host__ __device__ __forceinline__ VCmpNe4() {} + __host__ __device__ __forceinline__ VCmpNe4(const VCmpNe4&) {} }; struct VCmpLt4 : binary_function { @@ -1527,8 +1527,8 @@ namespace arithm return vcmplt4(a, b); } - __device__ __forceinline__ VCmpLt4() {} - __device__ __forceinline__ VCmpLt4(const VCmpLt4& other) {} + __host__ __device__ __forceinline__ VCmpLt4() {} + __host__ __device__ __forceinline__ VCmpLt4(const VCmpLt4&) {} }; struct VCmpLe4 : binary_function { @@ -1537,8 +1537,8 @@ namespace arithm return vcmple4(a, b); } - __device__ __forceinline__ VCmpLe4() {} - __device__ __forceinline__ VCmpLe4(const VCmpLe4& other) {} + __host__ __device__ __forceinline__ VCmpLe4() {} + __host__ __device__ __forceinline__ VCmpLe4(const VCmpLe4&) {} }; //////////////////////////////////// @@ -2008,8 +2008,8 @@ namespace arithm return vmin4(a, b); } - __device__ __forceinline__ VMin4() {} - __device__ __forceinline__ VMin4(const VMin4& other) {} + __host__ __device__ __forceinline__ VMin4() {} + __host__ __device__ __forceinline__ VMin4(const VMin4&) {} }; //////////////////////////////////// @@ -2021,8 +2021,8 @@ namespace arithm return vmin2(a, b); } - __device__ __forceinline__ VMin2() {} - __device__ __forceinline__ VMin2(const VMin2& other) {} + __host__ __device__ __forceinline__ VMin2() {} + __host__ __device__ __forceinline__ VMin2(const VMin2&) {} }; } @@ -2100,8 +2100,8 @@ namespace arithm return vmax4(a, b); } - __device__ __forceinline__ VMax4() {} - __device__ __forceinline__ VMax4(const VMax4& other) {} + __host__ __device__ __forceinline__ VMax4() {} + __host__ __device__ __forceinline__ VMax4(const VMax4&) {} }; //////////////////////////////////// @@ -2113,8 +2113,8 @@ namespace arithm return vmax2(a, b); } - __device__ __forceinline__ VMax2() {} - __device__ __forceinline__ VMax2(const VMax2& other) {} + __host__ __device__ __forceinline__ VMax2() {} + __host__ __device__ __forceinline__ VMax2(const VMax2&) {} }; } -- 2.7.4