43 #ifndef __OPENCV_GPU_VEC_TRAITS_HPP__
44 #define __OPENCV_GPU_VEC_TRAITS_HPP__
48 namespace cv {
namespace gpu {
namespace device
50 template<
typename T,
int N>
struct TypeVec;
54 uchar a0, a1, a2, a3, a4, a5, a6, a7;
58 uchar8
val = {a0, a1, a2, a3, a4, a5, a6, a7};
63 schar a0, a1, a2, a3, a4, a5, a6, a7;
67 char8
val = {a0, a1, a2, a3, a4, a5, a6, a7};
72 ushort a0, a1, a2, a3, a4, a5, a6, a7;
76 ushort8
val = {a0, a1, a2, a3, a4, a5, a6, a7};
81 short a0, a1, a2, a3, a4, a5, a6, a7;
83 static __host__ __device__ __forceinline__ short8 make_short8(
short a0,
short a1,
short a2,
short a3,
short a4,
short a5,
short a6,
short a7)
85 short8
val = {a0, a1, a2, a3, a4, a5, a6, a7};
90 uint a0, a1, a2, a3, a4, a5, a6, a7;
94 uint8
val = {a0, a1, a2, a3, a4, a5, a6, a7};
99 int a0, a1, a2, a3, a4, a5, a6, a7;
101 static __host__ __device__ __forceinline__ int8 make_int8(
int a0,
int a1,
int a2,
int a3,
int a4,
int a5,
int a6,
int a7)
103 int8
val = {a0, a1, a2, a3, a4, a5, a6, a7};
108 float a0, a1, a2, a3, a4, a5, a6, a7;
110 static __host__ __device__ __forceinline__ float8 make_float8(
float a0,
float a1,
float a2,
float a3,
float a4,
float a5,
float a6,
float a7)
112 float8
val = {a0, a1, a2, a3, a4, a5, a6, a7};
119 static __host__ __device__ __forceinline__
double8 make_double8(
double a0,
double a1,
double a2,
double a3,
double a4,
double a5,
double a6,
double a7)
121 double8 val = {a0, a1, a2, a3, a4, a5, a6, a7};
125 #define OPENCV_GPU_IMPLEMENT_TYPE_VEC(type) \
126 template<> struct TypeVec<type, 1> { typedef type vec_type; }; \
127 template<> struct TypeVec<type ## 1, 1> { typedef type ## 1 vec_type; }; \
128 template<> struct TypeVec<type, 2> { typedef type ## 2 vec_type; }; \
129 template<> struct TypeVec<type ## 2, 2> { typedef type ## 2 vec_type; }; \
130 template<> struct TypeVec<type, 3> { typedef type ## 3 vec_type; }; \
131 template<> struct TypeVec<type ## 3, 3> { typedef type ## 3 vec_type; }; \
132 template<> struct TypeVec<type, 4> { typedef type ## 4 vec_type; }; \
133 template<> struct TypeVec<type ## 4, 4> { typedef type ## 4 vec_type; }; \
134 template<> struct TypeVec<type, 8> { typedef type ## 8 vec_type; }; \
135 template<> struct TypeVec<type ## 8, 8> { typedef type ## 8 vec_type; };
146 #undef OPENCV_GPU_IMPLEMENT_TYPE_VEC
162 #define OPENCV_GPU_IMPLEMENT_VEC_TRAITS(type) \
163 template<> struct VecTraits<type> \
165 typedef type elem_type; \
167 static __device__ __host__ __forceinline__ type all(type v) {return v;} \
168 static __device__ __host__ __forceinline__ type make(type x) {return x;} \
169 static __device__ __host__ __forceinline__ type make(const type* v) {return *v;} \
171 template<> struct VecTraits<type ## 1> \
173 typedef type elem_type; \
175 static __device__ __host__ __forceinline__ type ## 1 all(type v) {return make_ ## type ## 1(v);} \
176 static __device__ __host__ __forceinline__ type ## 1 make(type x) {return make_ ## type ## 1(x);} \
177 static __device__ __host__ __forceinline__ type ## 1 make(const type* v) {return make_ ## type ## 1(*v);} \
179 template<> struct VecTraits<type ## 2> \
181 typedef type elem_type; \
183 static __device__ __host__ __forceinline__ type ## 2 all(type v) {return make_ ## type ## 2(v, v);} \
184 static __device__ __host__ __forceinline__ type ## 2 make(type x, type y) {return make_ ## type ## 2(x, y);} \
185 static __device__ __host__ __forceinline__ type ## 2 make(const type* v) {return make_ ## type ## 2(v[0], v[1]);} \
187 template<> struct VecTraits<type ## 3> \
189 typedef type elem_type; \
191 static __device__ __host__ __forceinline__ type ## 3 all(type v) {return make_ ## type ## 3(v, v, v);} \
192 static __device__ __host__ __forceinline__ type ## 3 make(type x, type y, type z) {return make_ ## type ## 3(x, y, z);} \
193 static __device__ __host__ __forceinline__ type ## 3 make(const type* v) {return make_ ## type ## 3(v[0], v[1], v[2]);} \
195 template<> struct VecTraits<type ## 4> \
197 typedef type elem_type; \
199 static __device__ __host__ __forceinline__ type ## 4 all(type v) {return make_ ## type ## 4(v, v, v, v);} \
200 static __device__ __host__ __forceinline__ type ## 4 make(type x, type y, type z, type w) {return make_ ## type ## 4(x, y, z, w);} \
201 static __device__ __host__ __forceinline__ type ## 4 make(const type* v) {return make_ ## type ## 4(v[0], v[1], v[2], v[3]);} \
203 template<> struct VecTraits<type ## 8> \
205 typedef type elem_type; \
207 static __device__ __host__ __forceinline__ type ## 8 all(type v) {return make_ ## type ## 8(v, v, v, v, v, v, v, v);} \
208 static __device__ __host__ __forceinline__ type ## 8 make(type a0, type a1, type a2, type a3, type a4, type a5, type a6, type a7) {return make_ ## type ## 8(a0, a1, a2, a3, a4, a5, a6, a7);} \
209 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]);} \
220 #undef OPENCV_GPU_IMPLEMENT_VEC_TRAITS
224 typedef char elem_type;
226 static __device__ __host__ __forceinline__
char all(
char v) {
return v;}
227 static __device__ __host__ __forceinline__
char make(
char x) {
return x;}
228 static __device__ __host__ __forceinline__
char make(
const char* x) {
return *
x;}
242 static __device__ __host__ __forceinline__ char1
all(
schar v) {
return make_char1(v);}
243 static __device__ __host__ __forceinline__ char1
make(
schar x) {
return make_char1(x);}
244 static __device__ __host__ __forceinline__ char1
make(
const schar*
v) {
return make_char1(v[0]);}
250 static __device__ __host__ __forceinline__ char2
all(
schar v) {
return make_char2(v, v);}
251 static __device__ __host__ __forceinline__ char2
make(
schar x,
schar y) {
return make_char2(x, y);}
252 static __device__ __host__ __forceinline__ char2
make(
const schar*
v) {
return make_char2(v[0], v[1]);}
258 static __device__ __host__ __forceinline__ char3
all(
schar v) {
return make_char3(v, v, v);}
260 static __device__ __host__ __forceinline__ char3
make(
const schar*
v) {
return make_char3(v[0], v[1], v[2]);}
266 static __device__ __host__ __forceinline__ char4
all(
schar v) {
return make_char4(v, v, v, v);}
268 static __device__ __host__ __forceinline__ char4
make(
const schar*
v) {
return make_char4(v[0], v[1], v[2], v[3]);}
274 static __device__ __host__ __forceinline__ char8
all(
schar v) {
return make_char8(v, v, v, v, v, v, v, v);}
275 static __device__ __host__ __forceinline__ char8
make(
schar a0,
schar a1,
schar a2,
schar a3,
schar a4,
schar a5,
schar a6,
schar a7) {
return make_char8(a0, a1, a2, a3, a4, a5, a6, a7);}
276 static __device__ __host__ __forceinline__ char8
make(
const schar*
v) {
return make_char8(v[0], v[1], v[2], v[3], v[4], v[5], v[6], v[7]);}
280 #endif // __OPENCV_GPU_VEC_TRAITS_HPP__
uchar3 vec_type
Definition: vec_traits.hpp:156
short float uchar uchar uchar uchar uchar ushort int uchar ushort int float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float float int int int float int int int float int CV_CUDEV_IMPLEMENT_VEC_BINARY_OP char CV_CUDEV_IMPLEMENT_VEC_BINARY_OP ushort CV_CUDEV_IMPLEMENT_VEC_BINARY_OP short CV_CUDEV_IMPLEMENT_VEC_BINARY_OP int CV_CUDEV_IMPLEMENT_VEC_BINARY_OP uint CV_CUDEV_IMPLEMENT_VEC_BINARY_OP float CV_CUDEV_IMPLEMENT_VEC_BINARY_OP double char
Definition: vec_math.hpp:426
GLenum GLint GLint y
Definition: core_c.h:613
schar vec_type
Definition: vec_traits.hpp:148
schar elem_type
Definition: vec_traits.hpp:240
signed char schar
Definition: types_c.h:174
schar elem_type
Definition: vec_traits.hpp:264
static __device__ __host__ __forceinline__ char4 all(schar v)
Definition: vec_traits.hpp:266
static __device__ __host__ __forceinline__ schar make(schar x)
Definition: vec_traits.hpp:235
static __device__ __host__ __forceinline__ char3 all(schar v)
Definition: vec_traits.hpp:258
static __device__ __host__ __forceinline__ char8 make(schar a0, schar a1, schar a2, schar a3, schar a4, schar a5, schar a6, schar a7)
Definition: vec_traits.hpp:275
static __device__ __host__ __forceinline__ schar make(const schar *x)
Definition: vec_traits.hpp:236
uchar vec_type
Definition: vec_traits.hpp:154
double a5
Definition: vec_traits.hpp:117
static __device__ __host__ __forceinline__ char1 make(const schar *v)
Definition: vec_traits.hpp:244
double a4
Definition: vec_traits.hpp:117
static __device__ __host__ __forceinline__ char2 make(schar x, schar y)
Definition: vec_traits.hpp:251
char2 vec_type
Definition: vec_traits.hpp:149
uchar8 vec_type
Definition: vec_traits.hpp:158
double a2
Definition: vec_traits.hpp:117
schar elem_type
Definition: vec_traits.hpp:272
struct __align__(8) uchar8
Definition: vec_traits.hpp:52
signed char schar
Definition: common.hpp:102
unsigned char uchar
Definition: common.hpp:100
double a0
Definition: vec_traits.hpp:117
double a6
Definition: vec_traits.hpp:117
schar elem_type
Definition: vec_traits.hpp:256
double a3
Definition: vec_traits.hpp:117
static __device__ __host__ __forceinline__ char2 make(const schar *v)
Definition: vec_traits.hpp:252
static __device__ __host__ __forceinline__ char8 make(const schar *v)
Definition: vec_traits.hpp:276
static __device__ __host__ __forceinline__ char1 make(schar x)
Definition: vec_traits.hpp:243
GLenum GLint x
Definition: core_c.h:632
static __device__ __host__ __forceinline__ char4 make(schar x, schar y, schar z, schar w)
Definition: vec_traits.hpp:267
unsigned int uint
Definition: common.hpp:104
static __device__ __host__ __forceinline__ schar all(schar v)
Definition: vec_traits.hpp:234
schar elem_type
Definition: vec_traits.hpp:248
unsigned short ushort
Definition: common.hpp:101
static __device__ __host__ __forceinline__ char4 make(const schar *v)
Definition: vec_traits.hpp:268
char4 vec_type
Definition: vec_traits.hpp:151
unsigned short ushort
Definition: types_c.h:171
static __device__ __host__ __forceinline__ char2 all(schar v)
Definition: vec_traits.hpp:250
char8 vec_type
Definition: vec_traits.hpp:152
uchar2 vec_type
Definition: vec_traits.hpp:155
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) template<> struct TypeVec< schar
Definition: vec_traits.hpp:50
unsigned char uchar
Definition: types_c.h:170
schar elem_type
Definition: vec_traits.hpp:232
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) template<> struct VecTraits< char >
Definition: vec_traits.hpp:212
GLubyte GLubyte GLubyte GLubyte w
double a7
Definition: vec_traits.hpp:117
static __device__ __host__ __forceinline__ char8 all(schar v)
Definition: vec_traits.hpp:274
int x
Definition: highgui_c.h:186
double a1
Definition: vec_traits.hpp:117
static __device__ __host__ __forceinline__ char1 all(schar v)
Definition: vec_traits.hpp:242
uchar4 vec_type
Definition: vec_traits.hpp:157
static __device__ __host__ __forceinline__ char3 make(const schar *v)
Definition: vec_traits.hpp:260
Definition: vec_traits.hpp:160
char3 vec_type
Definition: vec_traits.hpp:150
Definition: vec_traits.hpp:115
static __device__ __host__ __forceinline__ char3 make(schar x, schar y, schar z)
Definition: vec_traits.hpp:259