cuda_devptrs.hpp
Go to the documentation of this file.
1 /*M///////////////////////////////////////////////////////////////////////////////////////
2 //
3 // IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.
4 //
5 // By downloading, copying, installing or using the software you agree to this license.
6 // If you do not agree to this license, do not download, install,
7 // copy or use the software.
8 //
9 //
10 // License Agreement
11 // For Open Source Computer Vision Library
12 //
13 // Copyright (C) 2000-2008, Intel Corporation, all rights reserved.
14 // Copyright (C) 2009, Willow Garage Inc., all rights reserved.
15 // Third party copyrights are property of their respective owners.
16 //
17 // Redistribution and use in source and binary forms, with or without modification,
18 // are permitted provided that the following conditions are met:
19 //
20 // * Redistribution's of source code must retain the above copyright notice,
21 // this list of conditions and the following disclaimer.
22 //
23 // * Redistribution's in binary form must reproduce the above copyright notice,
24 // this list of conditions and the following disclaimer in the documentation
25 // and/or other materials provided with the distribution.
26 //
27 // * The name of the copyright holders may not be used to endorse or promote products
28 // derived from this software without specific prior written permission.
29 //
30 // This software is provided by the copyright holders and contributors "as is" and
31 // any express or implied warranties, including, but not limited to, the implied
32 // warranties of merchantability and fitness for a particular purpose are disclaimed.
33 // In no event shall the Intel Corporation or contributors be liable for any direct,
34 // indirect, incidental, special, exemplary, or consequential damages
35 // (including, but not limited to, procurement of substitute goods or services;
36 // loss of use, data, or profits; or business interruption) however caused
37 // and on any theory of liability, whether in contract, strict liability,
38 // or tort (including negligence or otherwise) arising in any way out of
39 // the use of this software, even if advised of the possibility of such damage.
40 //
41 //M*/
42 
43 #ifndef __OPENCV_CORE_DEVPTRS_HPP__
44 #define __OPENCV_CORE_DEVPTRS_HPP__
45 
46 #ifdef __cplusplus
47 
48 #ifdef __CUDACC__
49  #define __CV_GPU_HOST_DEVICE__ __host__ __device__ __forceinline__
50 #else
51  #define __CV_GPU_HOST_DEVICE__
52 #endif
53 
54 namespace cv
55 {
56  namespace gpu
57  {
58  // Simple lightweight structures that encapsulates information about an image on device.
59  // It is intended to pass to nvcc-compiled code. GpuMat depends on headers that nvcc can't compile
60 
61  template <bool expr> struct StaticAssert;
62  template <> struct StaticAssert<true> {static __CV_GPU_HOST_DEVICE__ void check(){}};
63 
64  template<typename T> struct DevPtr
65  {
66  typedef T elem_type;
67  typedef int index_type;
68 
69  enum { elem_size = sizeof(elem_type) };
70 
71  T* data;
72 
73  __CV_GPU_HOST_DEVICE__ DevPtr() : data(0) {}
74  __CV_GPU_HOST_DEVICE__ DevPtr(T* data_) : data(data_) {}
75 
76  __CV_GPU_HOST_DEVICE__ size_t elemSize() const { return elem_size; }
77  __CV_GPU_HOST_DEVICE__ operator T*() { return data; }
78  __CV_GPU_HOST_DEVICE__ operator const T*() const { return data; }
79  };
80 
81  template<typename T> struct PtrSz : public DevPtr<T>
82  {
83  __CV_GPU_HOST_DEVICE__ PtrSz() : size(0) {}
84  __CV_GPU_HOST_DEVICE__ PtrSz(T* data_, size_t size_) : DevPtr<T>(data_), size(size_) {}
85 
86  size_t size;
87  };
88 
89  template<typename T> struct PtrStep : public DevPtr<T>
90  {
91  __CV_GPU_HOST_DEVICE__ PtrStep() : step(0) {}
92  __CV_GPU_HOST_DEVICE__ PtrStep(T* data_, size_t step_) : DevPtr<T>(data_), step(step_) {}
93 
95  size_t step;
96 
97  __CV_GPU_HOST_DEVICE__ T* ptr(int y = 0) { return ( T*)( ( char*)DevPtr<T>::data + y * step); }
98  __CV_GPU_HOST_DEVICE__ const T* ptr(int y = 0) const { return (const T*)( (const char*)DevPtr<T>::data + y * step); }
99 
100  __CV_GPU_HOST_DEVICE__ T& operator ()(int y, int x) { return ptr(y)[x]; }
101  __CV_GPU_HOST_DEVICE__ const T& operator ()(int y, int x) const { return ptr(y)[x]; }
102  };
103 
104  template <typename T> struct PtrStepSz : public PtrStep<T>
105  {
106  __CV_GPU_HOST_DEVICE__ PtrStepSz() : cols(0), rows(0) {}
107  __CV_GPU_HOST_DEVICE__ PtrStepSz(int rows_, int cols_, T* data_, size_t step_)
108  : PtrStep<T>(data_, step_), cols(cols_), rows(rows_) {}
109 
110  template <typename U>
111  explicit PtrStepSz(const PtrStepSz<U>& d) : PtrStep<T>((T*)d.data, d.step), cols(d.cols), rows(d.rows){}
112 
113  int cols;
114  int rows;
115  };
116 
120 
124 
125 
126 #if defined __GNUC__
127  #define __CV_GPU_DEPR_BEFORE__
128  #define __CV_GPU_DEPR_AFTER__ __attribute__ ((deprecated))
129 #elif defined(__MSVC__) //|| defined(__CUDACC__)
130  #pragma deprecated(DevMem2D_)
131  #define __CV_GPU_DEPR_BEFORE__ __declspec(deprecated)
132  #define __CV_GPU_DEPR_AFTER__
133 #else
134  #define __CV_GPU_DEPR_BEFORE__
135  #define __CV_GPU_DEPR_AFTER__
136 #endif
137 
138  template <typename T> struct __CV_GPU_DEPR_BEFORE__ DevMem2D_ : public PtrStepSz<T>
139  {
141  DevMem2D_(int rows_, int cols_, T* data_, size_t step_) : PtrStepSz<T>(rows_, cols_, data_, step_) {}
142 
143  template <typename U>
144  explicit __CV_GPU_DEPR_BEFORE__ DevMem2D_(const DevMem2D_<U>& d) : PtrStepSz<T>(d.rows, d.cols, (T*)d.data, d.step) {}
146 
151 
152  template<typename T> struct PtrElemStep_ : public PtrStep<T>
153  {
154  PtrElemStep_(const DevMem2D_<T>& mem) : PtrStep<T>(mem.data, mem.step)
155  {
157 
159  }
160  __CV_GPU_HOST_DEVICE__ T* ptr(int y = 0) { return PtrStep<T>::data + y * PtrStep<T>::step; }
161  __CV_GPU_HOST_DEVICE__ const T* ptr(int y = 0) const { return PtrStep<T>::data + y * PtrStep<T>::step; }
162 
163  __CV_GPU_HOST_DEVICE__ T& operator ()(int y, int x) { return ptr(y)[x]; }
164  __CV_GPU_HOST_DEVICE__ const T& operator ()(int y, int x) const { return ptr(y)[x]; }
165  };
166 
167  template<typename T> struct PtrStep_ : public PtrStep<T>
168  {
169  PtrStep_() {}
170  PtrStep_(const DevMem2D_<T>& mem) : PtrStep<T>(mem.data, mem.step) {}
171  };
172 
176 
177 //#undef __CV_GPU_DEPR_BEFORE__
178 //#undef __CV_GPU_DEPR_AFTER__
179 
180  namespace device
181  {
182  using cv::gpu::PtrSz;
183  using cv::gpu::PtrStep;
184  using cv::gpu::PtrStepSz;
185 
186  using cv::gpu::PtrStepSzb;
187  using cv::gpu::PtrStepSzf;
188  using cv::gpu::PtrStepSzi;
189 
190  using cv::gpu::PtrStepb;
191  using cv::gpu::PtrStepf;
192  using cv::gpu::PtrStepi;
193  }
194  }
195 }
196 
197 #endif // __cplusplus
198 
199 #endif /* __OPENCV_CORE_DEVPTRS_HPP__ */
GLenum GLint GLint y
Definition: core_c.h:613
PtrStep< float > PtrStepf
Definition: cuda_devptrs.hpp:122
__CV_GPU_HOST_DEVICE__ const T * ptr(int y=0) const
Definition: cuda_devptrs.hpp:98
PtrElemStep_< int > PtrElemStepi
Definition: cuda_devptrs.hpp:175
__CV_GPU_HOST_DEVICE__ DevPtr(T *data_)
Definition: cuda_devptrs.hpp:74
__CV_GPU_HOST_DEVICE__ PtrStep()
Definition: cuda_devptrs.hpp:91
int cols
Definition: cuda_devptrs.hpp:113
T * data
Definition: cuda_devptrs.hpp:71
__CV_GPU_HOST_DEVICE__ PtrStepSz()
Definition: cuda_devptrs.hpp:106
__CV_GPU_HOST_DEVICE__ DevPtr()
Definition: cuda_devptrs.hpp:73
GLsizei GLsizei GLenum GLenum const GLvoid * data
Definition: core_c.h:403
int d
Definition: legacy.hpp:3064
const CvMat const CvMat const CvMat CvMat CvMat CvMat CvMat CvSize CvMat CvMat * T
Definition: calib3d.hpp:270
DevMem2Db DevMem2D
Definition: cuda_devptrs.hpp:148
cv::gpu::DevMem2D_ __CV_GPU_DEPR_AFTER__
__CV_GPU_HOST_DEVICE__ size_t elemSize() const
Definition: cuda_devptrs.hpp:76
void int step
Definition: core_c.h:403
PtrStep_()
Definition: cuda_devptrs.hpp:169
Definition: cuda_devptrs.hpp:104
__CV_GPU_HOST_DEVICE__ PtrSz(T *data_, size_t size_)
Definition: cuda_devptrs.hpp:84
__CV_GPU_DEPR_BEFORE__ DevMem2D_(const DevMem2D_< U > &d)
Definition: cuda_devptrs.hpp:144
PtrStep_(const DevMem2D_< T > &mem)
Definition: cuda_devptrs.hpp:170
PtrStepSz< unsigned char > PtrStepSzb
Definition: cuda_devptrs.hpp:117
size_t size
Definition: cuda_devptrs.hpp:86
PtrStep< int > PtrStepi
Definition: cuda_devptrs.hpp:123
DevMem2D_()
Definition: cuda_devptrs.hpp:140
Definition: cuda_devptrs.hpp:138
int index_type
Definition: cuda_devptrs.hpp:67
GLenum GLint x
Definition: core_c.h:632
Definition: cuda_devptrs.hpp:89
__CV_GPU_HOST_DEVICE__ PtrStepSz(int rows_, int cols_, T *data_, size_t step_)
Definition: cuda_devptrs.hpp:107
static __CV_GPU_HOST_DEVICE__ void check()
Definition: cuda_devptrs.hpp:62
size_t step
stride between two consecutive rows in bytes. Step is stored always and everywhere in bytes!!! ...
Definition: cuda_devptrs.hpp:95
DevMem2D_< int > DevMem2Di
Definition: cuda_devptrs.hpp:150
Definition: cuda_devptrs.hpp:81
__CV_GPU_HOST_DEVICE__ const T * ptr(int y=0) const
Definition: cuda_devptrs.hpp:161
__CV_GPU_HOST_DEVICE__ T * ptr(int y=0)
Definition: cuda_devptrs.hpp:160
int rows
Definition: core_c.h:114
DevMem2D_< unsigned char > DevMem2Db
Definition: cuda_devptrs.hpp:147
PtrStepSz< float > PtrStepSzf
Definition: cuda_devptrs.hpp:118
__CV_GPU_HOST_DEVICE__ T * ptr(int y=0)
Definition: cuda_devptrs.hpp:97
int cols
Definition: core_c.h:109
__CV_GPU_HOST_DEVICE__ PtrSz()
Definition: cuda_devptrs.hpp:83
PtrStep< unsigned char > PtrStepb
Definition: cuda_devptrs.hpp:121
__CV_GPU_HOST_DEVICE__ PtrStep(T *data_, size_t step_)
Definition: cuda_devptrs.hpp:92
const char * ptr
Definition: core_c.h:942
int rows
Definition: cuda_devptrs.hpp:114
Definition: cuda_devptrs.hpp:167
Definition: cuda_devptrs.hpp:152
PtrElemStep_< unsigned char > PtrElemStep
Definition: cuda_devptrs.hpp:173
PtrStepSz(const PtrStepSz< U > &d)
Definition: cuda_devptrs.hpp:111
PtrStepSz< int > PtrStepSzi
Definition: cuda_devptrs.hpp:119
PtrElemStep_(const DevMem2D_< T > &mem)
Definition: cuda_devptrs.hpp:154
true
Definition: color.hpp:221
DevMem2D_< float > DevMem2Df
Definition: cuda_devptrs.hpp:149
Definition: cuda_devptrs.hpp:61
int x
Definition: highgui_c.h:186
Definition: cuda_devptrs.hpp:69
T elem_type
Definition: cuda_devptrs.hpp:66
DevMem2D_(int rows_, int cols_, T *data_, size_t step_)
Definition: cuda_devptrs.hpp:141
Definition: cuda_devptrs.hpp:64
PtrElemStep_< float > PtrElemStepf
Definition: cuda_devptrs.hpp:174
GLsizeiptr size
Definition: core_c.h:939