Home
last modified time | relevance | path

Searched refs:Ncv32u (Results 1 – 25 of 44) sorted by relevance

12

/external/opencv3/modules/cudalegacy/include/opencv2/cudalegacy/
DNPP_staging.hpp117 Ncv32u nStep; ///< pitch
142 Ncv32u nStep,
143 Ncv32u *hpSize);
179 Ncv32u nSrcStep,
182 Ncv32u nDstStep,
213 Ncv32u nSrcStep,
216 Ncv32u nDstStep,
235 Ncv32u nSrcStep,
236 Ncv32u *hpSize);
262 Ncv32u nSrcStep,
[all …]
DNCVHaarObjectDetection.hpp84 …__host__ NCVStatus setRect(Ncv32u rectX, Ncv32u rectY, Ncv32u rectWidth, Ncv32u rectHeight, Ncv32u… in setRect()
100 …__device__ __host__ void getRect(Ncv32u *rectX, Ncv32u *rectY, Ncv32u *rectWidth, Ncv32u *rectHeig… in getRect()
127 Ncv32u desc;
132 Ncv32u numFeatures, Ncv32u offsetFeatures) in create()
166 __device__ __host__ Ncv32u getNumFeatures(void) in getNumFeatures()
171 __device__ __host__ Ncv32u getFeaturesOffset(void) in getFeaturesOffset()
187 __host__ NCVStatus create(Ncv32u offsetHaarClassifierNode) in create()
205 __device__ __host__ Ncv32u getNextNodeOffset(void) in getNextNodeOffset()
212 typedef Ncv32u __attribute__((__may_alias__)) Ncv32u_a;
214 typedef Ncv32u Ncv32u_a;
[all …]
DNCV.hpp128 typedef unsigned int Ncv32u; typedef
162 Ncv32u x; ///< x-coordinate of upper left corner.
163 Ncv32u y; ///< y-coordinate of upper left corner.
164 Ncv32u width; ///< Rectangle width.
165 Ncv32u height; ///< Rectangle height.
167 __host__ __device__ NcvRect32u(Ncv32u x_, Ncv32u y_, Ncv32u width_, Ncv32u height_) in NcvRect32u()
183 Ncv32u width; ///< Rectangle width.
184 Ncv32u height; ///< Rectangle height.
186 … __host__ __device__ NcvSize32u(Ncv32u width_, Ncv32u height_) : width(width_), height(height_) {} in NcvSize32u()
202 Ncv32u x; ///< Point X.
[all …]
/external/opencv3/modules/cudalegacy/src/cuda/
DNCVHaarObjectDetection.cu92 __device__ Ncv32u warpScanInclusive(Ncv32u idata, volatile Ncv32u *s_Data) in warpScanInclusive()
101 const Ncv32u n = cv::cuda::device::shfl_up(idata, i); in warpScanInclusive()
108 Ncv32u pos = 2 * threadIdx.x - (threadIdx.x & (K_WARP_SIZE - 1)); in warpScanInclusive()
123 __device__ __forceinline__ Ncv32u warpScanExclusive(Ncv32u idata, volatile Ncv32u *s_Data) in warpScanExclusive()
128 template <Ncv32u tiNumScanThreads>
129 __device__ Ncv32u scan1Inclusive(Ncv32u idata, volatile Ncv32u *s_Data) in scan1Inclusive()
134 Ncv32u warpResult = warpScanInclusive(idata, s_Data); in scan1Inclusive()
150 Ncv32u val = s_Data[threadIdx.x]; in scan1Inclusive()
173 const Ncv32u MAX_GRID_DIM = 65535;
176 const Ncv32u NUM_THREADS_ANCHORSPARALLEL = 64;
[all …]
DNPP_staging.cu55 texture<Ncv32u, 1, cudaReadModeElementType> tex32u;
113 Ncv32u pos = 2 * threadIdx.x - (threadIdx.x & (K_WARP_SIZE - 1)); in warpScanInclusive()
129 Ncv32u pos = 2 * threadIdx.x - (threadIdx.x & (K_WARP_SIZE - 1)); in warpScanInclusive()
151 template <class T, Ncv32u tiNumScanThreads>
196 const Ncv32u NUM_SCAN_THREADS = 256;
197 const Ncv32u LOG2_NUM_SCAN_THREADS = 8;
226 inline __device__ T readElem(T *d_src, Ncv32u texOffs, Ncv32u srcStride, Ncv32u curElemOffs);
230 inline __device__ Ncv8u readElem<Ncv8u>(Ncv8u *d_src, Ncv32u texOffs, Ncv32u srcStride, Ncv32u curE… in readElem()
237 inline __device__ Ncv32u readElem<Ncv32u>(Ncv32u *d_src, Ncv32u texOffs, Ncv32u srcStride, Ncv32u c… in readElem()
244 inline __device__ Ncv32f readElem<Ncv32f>(Ncv32f *d_src, Ncv32u texOffs, Ncv32u srcStride, Ncv32u c… in readElem()
[all …]
DNCV.cu55 const Ncv32u NUMTHREADS_DRAWRECTS = 32;
56 const Ncv32u NUMTHREADS_DRAWRECTS_LOG2 = 5;
61 Ncv32u dstStride, in drawRects()
62 Ncv32u dstWidth, in drawRects()
63 Ncv32u dstHeight, in drawRects()
65 Ncv32u numRects, in drawRects()
68 Ncv32u blockId = blockIdx.y * 65535 + blockIdx.x; in drawRects()
78 Ncv32u pt0x, pt0y; in drawRects()
81 Ncv32u numChunks = (curRect.height + NUMTHREADS_DRAWRECTS - 1) >> NUMTHREADS_DRAWRECTS_LOG2; in drawRects()
88 for (Ncv32u chunkId = 0; chunkId < numChunks; chunkId++) in drawRects()
[all …]
DNCVPixelOperations.hpp53 template<> inline __host__ __device__ Ncv32u _pixMaxVal<Ncv32u>() {return UINT_MAX;} in _pixMaxVal()
63 template<> inline __host__ __device__ Ncv32u _pixMinVal<Ncv32u>() {return 0;} in _pixMinVal()
77 template<> struct TConvVec2Base<uint1> {typedef Ncv32u TBase;};
78 template<> struct TConvVec2Base<uint3> {typedef Ncv32u TBase;};
79 template<> struct TConvVec2Base<uint4> {typedef Ncv32u TBase;};
89 template<typename TBase, Ncv32u NC> struct TConvBase2Vec;
96 template<> struct TConvBase2Vec<Ncv32u, 1> {typedef uint1 TVec;};
97 template<> struct TConvBase2Vec<Ncv32u, 3> {typedef uint3 TVec;};
98 template<> struct TConvBase2Vec<Ncv32u, 4> {typedef uint4 TVec;};
109 …me Tin> inline __host__ __device__ void _TDemoteClampZ(Tin &a, Ncv32u &out) {out = (Ncv32u)CLAMP(a… in _TDemoteClampZ()
[all …]
DNCVPyramid.cu54 template<typename T, Ncv32u CN> struct __average4_CN {static __host__ __device__ T _average4_CN(con…
149 template<typename Tin, typename Tout, Ncv32u CN> struct __lerp_CN {static __host__ __device__ Tout …
185 Ncv32u srcPitch, in kernelDownsampleX2()
187 Ncv32u dstPitch, in kernelDownsampleX2()
190 Ncv32u i = blockIdx.y * blockDim.y + threadIdx.y; in kernelDownsampleX2()
191 Ncv32u j = blockIdx.x * blockDim.x + threadIdx.x; in kernelDownsampleX2()
217 … kernelDownsampleX2<<<gDim, bDim, 0, stream>>>((T*)src.data, static_cast<Ncv32u>(src.step), in kernelDownsampleX2_gpu()
218 (T*)dst.data, static_cast<Ncv32u>(dst.step), NcvSize32u(dst.cols, dst.rows)); in kernelDownsampleX2_gpu()
253 Ncv32u srcTopPitch, in kernelInterpolateFrom1()
256 Ncv32u dstPitch, in kernelInterpolateFrom1()
[all …]
/external/opencv3/modules/cudalegacy/test/
Dmain_nvidia.cpp56 Ncv32u maxWidth, Ncv32u maxHeight) in generateIntegralTests()
60 Ncv32u i = (Ncv32u)_i; in generateIntegralTests()
67 Ncv32u i = (Ncv32u)_i; in generateIntegralTests()
77 Ncv32u maxWidth, Ncv32u maxHeight) in generateSquaredIntegralTests()
81 Ncv32u i = (Ncv32u)_i; in generateSquaredIntegralTests()
88 Ncv32u i = (Ncv32u)_i; in generateSquaredIntegralTests()
98 Ncv32u maxWidth, Ncv32u maxHeight) in generateRectStdDevTests()
104 Ncv32u i = (Ncv32u)_i; in generateRectStdDevTests()
119 for (Ncv32u i=2; i<10; ++i) in generateResizeTests()
127 for (Ncv32u i=2; i<10; ++i) in generateResizeTests()
[all …]
DTestHypothesesGrow.h54 TestHypothesesGrow(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
55 Ncv32u rectWidth, Ncv32u rectHeight, Ncv32f rectScale,
56 Ncv32u maxLenSrc, Ncv32u lenSrc, Ncv32u maxLenDst, Ncv32u lenDst);
68 NCVTestSourceProvider<Ncv32u> &src;
69 Ncv32u rectWidth;
70 Ncv32u rectHeight;
72 Ncv32u maxLenSrc;
73 Ncv32u lenSrc;
74 Ncv32u maxLenDst;
75 Ncv32u lenDst;
DNCVTestSourceProvider.hpp57 NCVTestSourceProvider(Ncv32u seed, T rangeLow, T rangeHigh, Ncv32u maxWidth, Ncv32u maxHeight) in NCVTestSourceProvider()
70 …allocatorCPU.reset(new NCVMemNativeAllocator(NCVMemoryTypeHostPinned, static_cast<Ncv32u>(devProp.… in NCVTestSourceProvider()
79 for (Ncv32u i=0; i<maxHeight; i++) in NCVTestSourceProvider()
81 for (Ncv32u j=0; j<data.get()->stride(); j++) in NCVTestSourceProvider()
105 …allocatorCPU.reset(new NCVMemNativeAllocator(NCVMemoryTypeHostPinned, static_cast<Ncv32u>(devProp.… in NCVTestSourceProvider()
128 for (Ncv32u i=0; i<dst.height(); i++) in fill()
130 Ncv32u srcLine = i % this->dataHeight; in fill()
132 Ncv32u srcFullChunks = dst.width() / this->dataWidth; in fill()
133 for (Ncv32u j=0; j<srcFullChunks; j++) in fill()
140 Ncv32u srcLastChunk = dst.width() % this->dataWidth; in fill()
[all …]
DTestCompact.cpp45 TestCompact::TestCompact(std::string testName_, NCVTestSourceProvider<Ncv32u> &src_, in TestCompact()
46 … Ncv32u length_, Ncv32u badElem_, Ncv32u badElemPercentage_) in TestCompact()
77 NCVVectorAlloc<Ncv32u> h_vecSrc(*this->allocatorCPU.get(), this->length); in process()
79 NCVVectorAlloc<Ncv32u> d_vecSrc(*this->allocatorGPU.get(), this->length); in process()
82 NCVVectorAlloc<Ncv32u> h_vecDst(*this->allocatorCPU.get(), this->length); in process()
84 NCVVectorAlloc<Ncv32u> d_vecDst(*this->allocatorGPU.get(), this->length); in process()
86 NCVVectorAlloc<Ncv32u> h_vecDst_d(*this->allocatorCPU.get(), this->length); in process()
92 for (Ncv32u i=0; i<this->length; i++) in process()
94 Ncv32u tmp = (h_vecSrc.ptr()[i]) & 0xFF; in process()
103 NCVVectorAlloc<Ncv32u> h_dstLen(*this->allocatorCPU.get(), 1); in process()
[all …]
DTestDrawRects.cpp48 NCVTestSourceProvider<Ncv32u> &src32u_, in TestDrawRects()
49 Ncv32u width_, Ncv32u height_, Ncv32u numRects_, T color_) in TestDrawRects()
107 NCVVectorReuse<Ncv32u> h_rects_as32u(h_rects.getSegment()); in process()
110 for (Ncv32u i=0; i<this->numRects; i++) in process()
112 h_rects.ptr()[i].x = (Ncv32u)(((1.0 * h_rects.ptr()[i].x) / RAND_MAX) * (this->width-2)); in process()
113 h_rects.ptr()[i].y = (Ncv32u)(((1.0 * h_rects.ptr()[i].y) / RAND_MAX) * (this->height-2)); in process()
114 …h_rects.ptr()[i].width = (Ncv32u)(((1.0 * h_rects.ptr()[i].width) / RAND_MAX) * (this->width+10 - … in process()
115 …h_rects.ptr()[i].height = (Ncv32u)(((1.0 * h_rects.ptr()[i].height) / RAND_MAX) * (this->height+10… in process()
121 if (sizeof(T) == sizeof(Ncv32u)) in process()
123 …ncvStat = ncvDrawRects_32u_device((Ncv32u *)d_img.ptr(), d_img.stride(), this->width, this->height, in process()
[all …]
DTestHypothesesGrow.cpp46 TestHypothesesGrow::TestHypothesesGrow(std::string testName_, NCVTestSourceProvider<Ncv32u> &src_, in TestHypothesesGrow()
47 Ncv32u rectWidth_, Ncv32u rectHeight_, Ncv32f rectScale_, in TestHypothesesGrow()
48 … Ncv32u maxLenSrc_, Ncv32u lenSrc_, Ncv32u maxLenDst_, Ncv32u lenDst_) in TestHypothesesGrow()
87 NCVVectorAlloc<Ncv32u> h_vecSrc(*this->allocatorCPU.get(), this->maxLenSrc); in process()
89 NCVVectorAlloc<Ncv32u> d_vecSrc(*this->allocatorGPU.get(), this->maxLenSrc); in process()
104 …NCVVectorReuse<Ncv32u> h_vecDst_as32u(h_vecDst.getSegment(), lenDst * sizeof(NcvRect32u) / sizeof(… in process()
116 Ncv32u h_outElemNum_d = 0; in process()
117 Ncv32u h_outElemNum_h = 0; in process()
DTestCompact.h54 TestCompact(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
55 Ncv32u length, Ncv32u badElem, Ncv32u badElemPercentage);
67 NCVTestSourceProvider<Ncv32u> &src;
68 Ncv32u length;
69 Ncv32u badElem;
70 Ncv32u badElemPercentage;
DTestHypothesesFilter.h54 TestHypothesesFilter(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
55 Ncv32u numDstRects, Ncv32u minNeighbors, Ncv32f eps);
67 NCVTestSourceProvider<Ncv32u> &src;
68 Ncv32u numDstRects;
69 Ncv32u minNeighbors;
72 Ncv32u canvasWidth;
73 Ncv32u canvasHeight;
DTestDrawRects.h55 …TestDrawRects(std::string testName, NCVTestSourceProvider<T> &src, NCVTestSourceProvider<Ncv32u> &…
56 Ncv32u width, Ncv32u height, Ncv32u numRects, T color);
69 NCVTestSourceProvider<Ncv32u> &src32u;
70 Ncv32u width;
71 Ncv32u height;
72 Ncv32u numRects;
DTestHypothesesFilter.cpp46 TestHypothesesFilter::TestHypothesesFilter(std::string testName_, NCVTestSourceProvider<Ncv32u> &sr… in TestHypothesesFilter()
47 Ncv32u numDstRects_, Ncv32u minNeighbors_, Ncv32f eps_) in TestHypothesesFilter()
96 …NCVVectorAlloc<Ncv32u> h_random32u(*this->allocatorCPU.get(), this->numDstRects * sizeof(NcvRect32… in process()
99 Ncv32u srcSlotSize = 2 * this->minNeighbors + 1; in process()
110 Ncv32u randCnt = 0; in process()
113 for (Ncv32u i=0; i<this->numDstRects; i++) in process()
120 …Ncv32u numNeighbors = this->minNeighbors + 1 + (Ncv32u)(((1.0 * h_random32u.ptr()[i]) * (this->min… in process()
124 for (Ncv32u j=0; j<numNeighbors; j++) in process()
139 for (Ncv32u j=numNeighbors; j<srcSlotSize; j++) in process()
155 for (Ncv32u i=0; i<this->numDstRects*srcSlotSize-1; i++) in process()
[all …]
DTestRectStdDev.cpp47 Ncv32u width_, Ncv32u height_, NcvRect32u rect_, Ncv32f scaleFactor_, in TestRectStdDev()
89 Ncv32u normWidth = (Ncv32u)_normWidth; in process()
90 Ncv32u normHeight = (Ncv32u)_normHeight; in process()
93 Ncv32u widthII = this->width + 1; in process()
94 Ncv32u heightII = this->height + 1; in process()
95 Ncv32u widthSII = this->width + 1; in process()
96 Ncv32u heightSII = this->height + 1; in process()
103 NCVMatrixAlloc<Ncv32u> d_imgII(*this->allocatorGPU.get(), widthII, heightII); in process()
105 NCVMatrixAlloc<Ncv32u> h_imgII(*this->allocatorCPU.get(), widthII, heightII); in process()
120 Ncv32u bufSizeII, bufSizeSII; in process()
[all …]
DTestResize.cpp48 … Ncv32u width_, Ncv32u height_, Ncv32u scaleFactor_, NcvBool bTextureCache_) in TestResize()
113 if (sizeof(T) == sizeof(Ncv32u)) in process()
115 ncvStat = nppiStDecimate_32u_C1R((Ncv32u *)d_img.ptr(), d_img.pitch(), in process()
116 (Ncv32u *)d_small.ptr(), d_small.pitch(), in process()
137 if (sizeof(T) == sizeof(Ncv32u)) in process()
139 ncvStat = nppiStDecimate_32u_C1R_host((Ncv32u *)h_img.ptr(), h_img.pitch(), in process()
140 (Ncv32u *)h_small.ptr(), h_small.pitch(), in process()
161 for (Ncv32u i=0; bLoopVirgin && i < h_small.height(); i++) in process()
163 for (Ncv32u j=0; bLoopVirgin && j < h_small.width(); j++) in process()
189 template class TestResize<Ncv32u>;
DTestTranspose.cpp48 Ncv32u width_, Ncv32u height_) in TestTranspose()
102 if (sizeof(T) == sizeof(Ncv32u)) in process()
104 ncvStat = nppiStTranspose_32u_C1R((Ncv32u *)d_img.ptr(), d_img.pitch(), in process()
105 (Ncv32u *)d_dst.ptr(), d_dst.pitch(), in process()
124 if (sizeof(T) == sizeof(Ncv32u)) in process()
126 ncvStat = nppiStTranspose_32u_C1R_host((Ncv32u *)h_img.ptr(), h_img.pitch(), in process()
127 (Ncv32u *)h_dst.ptr(), h_dst.pitch(), in process()
148 for (Ncv32u i=0; bLoopVirgin && i < this->width; i++) in process()
150 for (Ncv32u j=0; bLoopVirgin && j < this->height; j++) in process()
176 template class TestTranspose<Ncv32u>;
DTestResize.h55 Ncv32u width, Ncv32u height, Ncv32u scaleFactor, NcvBool bTextureCache);
67 Ncv32u width;
68 Ncv32u height;
69 Ncv32u scaleFactor;
DNCVTest.hpp64 std::map<std::string, Ncv32u> statsNums;
163 … this->allocatorGPU.reset(new NCVMemStackAllocator(static_cast<Ncv32u>(devProp.textureAlignment))); in initMemory()
164 … this->allocatorCPU.reset(new NCVMemStackAllocator(static_cast<Ncv32u>(devProp.textureAlignment))); in initMemory()
179 Ncv32u maxGPUsize = (Ncv32u)this->allocatorGPU.get()->maxSize(); in initMemory()
180 Ncv32u maxCPUsize = (Ncv32u)this->allocatorCPU.get()->maxSize(); in initMemory()
185 …et(new NCVMemStackAllocator(NCVMemoryTypeDevice, maxGPUsize, static_cast<Ncv32u>(devProp.textureAl… in initMemory()
187 …ew NCVMemStackAllocator(NCVMemoryTypeHostPinned, maxCPUsize, static_cast<Ncv32u>(devProp.textureAl… in initMemory()
221 for (std::map<std::string,Ncv32u>::iterator it=report.statsNums.begin(); in dumpToFile()
/external/opencv3/modules/cudalegacy/src/
DNCV.cpp160 Ncv32u alignUp(Ncv32u what, Ncv32u alignment) in alignUp()
162 Ncv32u alignMask = alignment-1; in alignUp()
163 Ncv32u inverseAlignMask = ~alignMask; in alignUp()
164 Ncv32u res = (what + alignMask) & inverseAlignMask; in alignUp()
250 NCVStatus memSegCopyHelper2D(void *dst, Ncv32u dstPitch, NCVMemoryType dstType, in memSegCopyHelper2D()
251 const void *src, Ncv32u srcPitch, NCVMemoryType srcType, in memSegCopyHelper2D()
252 Ncv32u widthbytes, Ncv32u height, cudaStream_t cuStream) in memSegCopyHelper2D()
263 for (Ncv32u i=0; i<height; i++) in memSegCopyHelper2D()
329 NCVMemStackAllocator::NCVMemStackAllocator(Ncv32u alignment_) : in NCVMemStackAllocator()
344 NCVMemStackAllocator::NCVMemStackAllocator(NCVMemoryType memT, size_t capacity, Ncv32u alignment_, … in NCVMemStackAllocator()
[all …]
/external/opencv3/samples/gpu/
Dcascadeclassifier_nvidia_api.cpp87 Ncv32u width, Ncv32u height, in process()
111 for (Ncv32u i=0; i<(Ncv32u)srcdst->rows; i++) in process()
126 Ncv32u numDetections; in process()
145 for (Ncv32u i=0; i<(Ncv32u)srcdst->rows; i++) in process()
226 …NCVMemNativeAllocator gpuCascadeAllocator(NCVMemoryTypeDevice, static_cast<Ncv32u>(devProp.texture… in main()
228 …NCVMemNativeAllocator cpuCascadeAllocator(NCVMemoryTypeHostPinned, static_cast<Ncv32u>(devProp.tex… in main()
231 Ncv32u haarNumStages, haarNumNodes, haarNumFeatures; in main()
267 NCVMemStackAllocator gpuCounter(static_cast<Ncv32u>(devProp.textureAlignment)); in main()
269 NCVMemStackAllocator cpuCounter(static_cast<Ncv32u>(devProp.textureAlignment)); in main()
279 …NCVMemStackAllocator gpuAllocator(NCVMemoryTypeDevice, gpuCounter.maxSize(), static_cast<Ncv32u>(d… in main()
[all …]

12