Point Cloud Library (PCL) 1.12.1
NCV.hpp
1/*
2 * Software License Agreement (BSD License)
3 *
4 * Point Cloud Library (PCL) - www.pointclouds.org
5 * Copyright (C) 2009-2010, NVIDIA Corporation, all rights reserved.
6 * Third party copyrights are property of their respective owners.
7 *
8 * All rights reserved.
9 *
10 * Redistribution and use in source and binary forms, with or without
11 * modification, are permitted provided that the following conditions
12 * are met:
13 *
14 * * Redistributions of source code must retain the above copyright
15 * notice, this list of conditions and the following disclaimer.
16 * * Redistributions in binary form must reproduce the above
17 * copyright notice, this list of conditions and the following
18 * disclaimer in the documentation and/or other materials provided
19 * with the distribution.
20 * * Neither the name of Willow Garage, Inc. nor the names of its
21 * contributors may be used to endorse or promote products derived
22 * from this software without specific prior written permission.
23 *
24 * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
25 * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT
26 * LIMITED TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS
27 * FOR A PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE
28 * COPYRIGHT OWNER OR CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT,
29 * INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING,
30 * BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES;
31 * LOSS OF USE, DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER
32 * CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT
33 * LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN
34 * ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF THE
35 * POSSIBILITY OF SUCH DAMAGE.
36 *
37 * $Id: $
38 * Ported to PCL by Koen Buys : Attention Work in progress!
39 */
40
41#ifndef PCL_GPU_PEOPLE__NCV_HPP_
42#define PCL_GPU_PEOPLE__NCV_HPP_
43
44#if (defined _WIN32 || defined WINCE) && defined CVAPI_EXPORTS
45 #define NCV_EXPORTS __declspec(dllexport)
46#else
47 #define NCV_EXPORTS
48#endif
49
50#ifdef _WIN32
51 #define WIN32_LEAN_AND_MEAN
52#endif
53
54#include <cuda_runtime.h>
55#include <sstream>
56#include <iostream>
57#include <pcl/console/print.h>
58
59//==============================================================================
60//
61// Compile-time assert functionality
62//
63//==============================================================================
64
65/**
66* Compile-time assert namespace
67*/
68namespace NcvCTprep
69{
70 template <bool x>
72
73 template <>
74 struct CT_ASSERT_FAILURE<true> {};
75
76 template <int x>
77 struct assertTest{};
78}
79
80#define NCV_CT_PREP_PASTE_AUX(a,b) a##b ///< Concatenation indirection macro
81#define NCV_CT_PREP_PASTE(a,b) NCV_CT_PREP_PASTE_AUX(a, b) ///< Concatenation macro
82
83/**
84* Performs compile-time assertion of a condition on the file scope
85*/
86#define NCV_CT_ASSERT(X) \
87 typedef NcvCTprep::assertTest<sizeof(NcvCTprep::CT_ASSERT_FAILURE< (bool)(X) >)> \
88 NCV_CT_PREP_PASTE(__ct_assert_typedef_, __LINE__)
89
90//==============================================================================
91//
92// Alignment macros
93//
94//==============================================================================
95
96#if !defined(__align__) && !defined(__CUDACC__)
97 #if defined(_WIN32) || defined(_WIN64)
98 #define __align__(n) __declspec(align(n))
99 #elif defined(__unix__)
100 #define __align__(n) __attribute__((__aligned__(n)))
101 #endif
102#endif
103
104//==============================================================================
105//
106// Integral and compound types of guaranteed size
107//
108//==============================================================================
109
110using NcvBool = bool;
111using Ncv64s = long long;
112
113#if defined(__APPLE__) && !defined(__CUDACC__)
114 using Ncv64u = std::uint64_t;
115#else
116 using Ncv64u = unsigned long long;
117#endif
118
119using Ncv32s = int;
120using Ncv32u = unsigned int;
121using Ncv16s = short;
122using Ncv16u = unsigned short;
123using Ncv8s = char;
124using Ncv8u = unsigned char;
125using Ncv32f = float;
126using Ncv64f = double;
127
129{
130 Ncv8u x;
131 Ncv8u y;
132 Ncv8u width;
133 Ncv8u height;
134 __host__ __device__ NcvRect8u() : x(0), y(0), width(0), height(0) {};
135 __host__ __device__ NcvRect8u(Ncv8u x, Ncv8u y, Ncv8u width, Ncv8u height) : x(x), y(y), width(width), height(height) {}
136};
137
139{
140 Ncv32s x; ///< x-coordinate of upper left corner.
141 Ncv32s y; ///< y-coordinate of upper left corner.
142 Ncv32s width; ///< Rectangle width.
143 Ncv32s height; ///< Rectangle height.
144 __host__ __device__ NcvRect32s() : x(0), y(0), width(0), height(0) {};
145 __host__ __device__ NcvRect32s(Ncv32s x, Ncv32s y, Ncv32s width, Ncv32s height) : x(x), y(y), width(width), height(height) {}
146};
147
149{
150 Ncv32u x; ///< x-coordinate of upper left corner.
151 Ncv32u y; ///< y-coordinate of upper left corner.
152 Ncv32u width; ///< Rectangle width.
153 Ncv32u height; ///< Rectangle height.
154 __host__ __device__ NcvRect32u() : x(0), y(0), width(0), height(0) {};
155 __host__ __device__ NcvRect32u(Ncv32u x, Ncv32u y, Ncv32u width, Ncv32u height) : x(x), y(y), width(width), height(height) {}
156};
157
159{
160 Ncv32s width; ///< Rectangle width.
161 Ncv32s height; ///< Rectangle height.
162 __host__ __device__ NcvSize32s() : width(0), height(0) {};
163 __host__ __device__ NcvSize32s(Ncv32s width, Ncv32s height) : width(width), height(height) {}
164};
165
167{
168 Ncv32u width; ///< Rectangle width.
169 Ncv32u height; ///< Rectangle height.
170 __host__ __device__ NcvSize32u() : width(0), height(0) {};
171 __host__ __device__ NcvSize32u(Ncv32u width, Ncv32u height) : width(width), height(height) {}
172 __host__ __device__ bool operator == (const NcvSize32u &another) const {return this->width == another.width && this->height == another.height;}
173};
174
176{
177 Ncv32s x; ///< Point X.
178 Ncv32s y; ///< Point Y.
179 __host__ __device__ NcvPoint2D32s() : x(0), y(0) {};
180 __host__ __device__ NcvPoint2D32s(Ncv32s x, Ncv32s y) : x(x), y(y) {}
181};
182
184{
185 Ncv32u x; ///< Point X.
186 Ncv32u y; ///< Point Y.
187 __host__ __device__ NcvPoint2D32u() : x(0), y(0) {};
188 __host__ __device__ NcvPoint2D32u(Ncv32u x, Ncv32u y) : x(x), y(y) {}
189};
190
191NCV_CT_ASSERT(sizeof(NcvBool) <= 4);
192NCV_CT_ASSERT(sizeof(Ncv64s) == 8);
193NCV_CT_ASSERT(sizeof(Ncv64u) == 8);
194NCV_CT_ASSERT(sizeof(Ncv32s) == 4);
195NCV_CT_ASSERT(sizeof(Ncv32u) == 4);
196NCV_CT_ASSERT(sizeof(Ncv16s) == 2);
197NCV_CT_ASSERT(sizeof(Ncv16u) == 2);
198NCV_CT_ASSERT(sizeof(Ncv8s) == 1);
199NCV_CT_ASSERT(sizeof(Ncv8u) == 1);
200NCV_CT_ASSERT(sizeof(Ncv32f) == 4);
201NCV_CT_ASSERT(sizeof(Ncv64f) == 8);
202NCV_CT_ASSERT(sizeof(NcvRect8u) == sizeof(Ncv32u));
203NCV_CT_ASSERT(sizeof(NcvRect32s) == 4 * sizeof(Ncv32s));
204NCV_CT_ASSERT(sizeof(NcvRect32u) == 4 * sizeof(Ncv32u));
205NCV_CT_ASSERT(sizeof(NcvSize32u) == 2 * sizeof(Ncv32u));
206NCV_CT_ASSERT(sizeof(NcvPoint2D32u) == 2 * sizeof(Ncv32u));
207
208
209//==============================================================================
210//
211// Persistent constants
212//
213//==============================================================================
214
215const Ncv32u K_WARP_SIZE = 32;
216const Ncv32u K_LOG2_WARP_SIZE = 5;
217
218//==============================================================================
219//
220// Error handling
221//
222//==============================================================================
223
224NCV_EXPORTS void ncvDebugOutput(const std::string &msg);
225
226using NCVDebugOutputHandler = void (const std::string &);
227
228NCV_EXPORTS void ncvSetDebugOutputHandler(NCVDebugOutputHandler* func);
229
230#define ncvAssertPrintCheck(pred, msg) \
231 do \
232 { \
233 if (!(pred)) \
234 { \
235 std::ostringstream oss; \
236 oss << "NCV Assertion Failed: " << msg << ", file=" << __FILE__ << ", line=" << __LINE__ << std::endl; \
237 ncvDebugOutput(oss.str()); \
238 } \
239 } while (0)
240
241#define ncvAssertPrintReturn(pred, msg, err) \
242 do \
243 { \
244 ncvAssertPrintCheck(pred, msg); \
245 if (!(pred)) return err; \
246 } while (0)
247
248#define ncvAssertReturn(pred, err) \
249 ncvAssertPrintReturn(pred, "retcode=" << (int)err, err)
250
251#define ncvAssertReturnNcvStat(ncvOp) \
252 do \
253 { \
254 NCVStatus _ncvStat = ncvOp; \
255 ncvAssertPrintReturn(NCV_SUCCESS==_ncvStat, "NcvStat=" << (int)_ncvStat, _ncvStat); \
256 } while (0)
257
258#define ncvAssertCUDAReturn(cudacall, errCode) \
259 do \
260 { \
261 cudaError_t res = cudacall; \
262 ncvAssertPrintReturn(cudaSuccess==res, "cudaError_t=" << res, errCode); \
263 } while (0)
264
265#define ncvAssertCUDALastErrorReturn(errCode) \
266 do \
267 { \
268 cudaError_t res = cudaGetLastError(); \
269 ncvAssertPrintReturn(cudaSuccess==res, "cudaError_t=" << res, errCode); \
270 } while (0)
271
272/**
273* \brief Return-codes for status notification, errors and warnings
274*/
275enum
276{
277 //NCV statuses
278 NCV_SUCCESS,
279 NCV_UNKNOWN_ERROR,
280
281 NCV_CUDA_ERROR,
282 NCV_NPP_ERROR,
283 NCV_FILE_ERROR,
284
285 NCV_NULL_PTR,
286 NCV_INCONSISTENT_INPUT,
287 NCV_TEXTURE_BIND_ERROR,
288 NCV_DIMENSIONS_INVALID,
289
290 NCV_INVALID_ROI,
291 NCV_INVALID_STEP,
292 NCV_INVALID_SCALE,
293
294 NCV_ALLOCATOR_NOT_INITIALIZED,
295 NCV_ALLOCATOR_BAD_ALLOC,
296 NCV_ALLOCATOR_BAD_DEALLOC,
297 NCV_ALLOCATOR_INSUFFICIENT_CAPACITY,
298 NCV_ALLOCATOR_DEALLOC_ORDER,
299 NCV_ALLOCATOR_BAD_REUSE,
300
301 NCV_MEM_COPY_ERROR,
302 NCV_MEM_RESIDENCE_ERROR,
303 NCV_MEM_INSUFFICIENT_CAPACITY,
304
305 NCV_HAAR_INVALID_PIXEL_STEP,
306 NCV_HAAR_TOO_MANY_FEATURES_IN_CLASSIFIER,
307 NCV_HAAR_TOO_MANY_FEATURES_IN_CASCADE,
308 NCV_HAAR_TOO_LARGE_FEATURES,
309 NCV_HAAR_XML_LOADING_EXCEPTION,
310
311 NCV_NOIMPL_HAAR_TILTED_FEATURES,
312 NCV_NOT_IMPLEMENTED,
313
314 NCV_WARNING_HAAR_DETECTIONS_VECTOR_OVERFLOW,
315
316 //NPP statuses
317 NPPST_SUCCESS = NCV_SUCCESS, ///< Successful operation (same as NPP_NO_ERROR)
318 NPPST_ERROR, ///< Unknown error
319 NPPST_CUDA_KERNEL_EXECUTION_ERROR, ///< CUDA kernel execution error
320 NPPST_NULL_POINTER_ERROR, ///< NULL pointer argument error
321 NPPST_TEXTURE_BIND_ERROR, ///< CUDA texture binding error or non-zero offset returned
322 NPPST_MEMCPY_ERROR, ///< CUDA memory copy error
323 NPPST_MEM_ALLOC_ERR, ///< CUDA memory allocation error
324 NPPST_MEMFREE_ERR, ///< CUDA memory deallocation error
325
326 //NPPST statuses
327 NPPST_INVALID_ROI, ///< Invalid region of interest argument
328 NPPST_INVALID_STEP, ///< Invalid image lines step argument (check sign, alignment, relation to image width)
329 NPPST_INVALID_SCALE, ///< Invalid scale parameter passed
330 NPPST_MEM_INSUFFICIENT_BUFFER, ///< Insufficient user-allocated buffer
331 NPPST_MEM_RESIDENCE_ERROR, ///< Memory residence error detected (check if pointers should be device or pinned)
332 NPPST_MEM_INTERNAL_ERROR, ///< Internal memory management error
333
334 NCV_LAST_STATUS ///< Marker to continue error numeration in other files
335};
336
337using NCVStatus = Ncv32u;
338
339#define NCV_SET_SKIP_COND(x) \
340 bool __ncv_skip_cond = x
341
342#define NCV_RESET_SKIP_COND(x) \
343 __ncv_skip_cond = x
344
345#define NCV_SKIP_COND_BEGIN \
346 if (!__ncv_skip_cond) {
347
348#define NCV_SKIP_COND_END \
349 }
350
351
352//==============================================================================
353//
354// Memory management classes template compound types
355//
356//==============================================================================
357
358
359/**
360* Calculates the aligned top bound value
361*/
362NCV_EXPORTS Ncv32u alignUp(Ncv32u what, Ncv32u alignment);
363
364
365/**
366* NCVMemoryType
367*/
368enum NCVMemoryType
369{
370 NCVMemoryTypeNone,
371 NCVMemoryTypeHostPageable,
372 NCVMemoryTypeHostPinned,
373 NCVMemoryTypeDevice
374};
375
376
377/**
378* NCVMemPtr
379*/
380struct NCV_EXPORTS NCVMemPtr
381{
382 void *ptr;
383 NCVMemoryType memtype;
384 void clear();
385};
386
387
388/**
389* NCVMemSegment
390*/
391struct NCV_EXPORTS NCVMemSegment
392{
394 std::size_t size;
395 void clear();
396};
397
398
399/**
400* INCVMemAllocator (Interface)
401*/
402class NCV_EXPORTS INCVMemAllocator
403{
404public:
405 virtual ~INCVMemAllocator() = 0;
406
407 virtual NCVStatus alloc(NCVMemSegment &seg, std::size_t size) = 0;
408 virtual NCVStatus dealloc(NCVMemSegment &seg) = 0;
409
410 virtual NcvBool isInitialized() const = 0;
411 virtual NcvBool isCounting() const = 0;
412
413 virtual NCVMemoryType memType() const = 0;
414 virtual Ncv32u alignment() const = 0;
415 virtual std::size_t maxSize() const = 0;
416};
417
419
420
421/**
422* NCVMemStackAllocator
423*/
424class NCV_EXPORTS NCVMemStackAllocator : public INCVMemAllocator
425{
428
429public:
430
431 explicit NCVMemStackAllocator(Ncv32u alignment);
432 NCVMemStackAllocator(NCVMemoryType memT, std::size_t capacity, Ncv32u alignment, void *reusePtr=nullptr);
434
435 virtual NCVStatus alloc(NCVMemSegment &seg, std::size_t size);
436 virtual NCVStatus dealloc(NCVMemSegment &seg);
437
438 virtual NcvBool isInitialized() const;
439 virtual NcvBool isCounting() const;
440
441 virtual NCVMemoryType memType() const;
442 virtual Ncv32u alignment() const;
443 virtual std::size_t maxSize() const;
444
445private:
446
447 NCVMemoryType _memType;
448 Ncv32u _alignment;
449 Ncv8u *allocBegin;
450 Ncv8u *begin;
451 Ncv8u *end;
452 std::size_t currentSize;
453 std::size_t _maxSize;
454 NcvBool bReusesMemory;
455};
456
457
458/**
459* NCVMemNativeAllocator
460*/
461class NCV_EXPORTS NCVMemNativeAllocator : public INCVMemAllocator
462{
463public:
464
465 NCVMemNativeAllocator(NCVMemoryType memT, Ncv32u alignment);
467
468 virtual NCVStatus alloc(NCVMemSegment &seg, std::size_t size);
469 virtual NCVStatus dealloc(NCVMemSegment &seg);
470
471 virtual NcvBool isInitialized() const;
472 virtual NcvBool isCounting() const;
473
474 virtual NCVMemoryType memType() const;
475 virtual Ncv32u alignment() const;
476 virtual std::size_t maxSize() const;
477
478private:
479
482
483 NCVMemoryType _memType;
484 Ncv32u _alignment;
485 std::size_t currentSize;
486 std::size_t _maxSize;
487};
488
489
490/**
491* Copy dispatchers
492*/
493NCV_EXPORTS NCVStatus memSegCopyHelper(void *dst, NCVMemoryType dstType,
494 const void *src, NCVMemoryType srcType,
495 std::size_t sz, cudaStream_t cuStream);
496
497
498NCV_EXPORTS NCVStatus memSegCopyHelper2D(void *dst, Ncv32u dstPitch, NCVMemoryType dstType,
499 const void *src, Ncv32u srcPitch, NCVMemoryType srcType,
500 Ncv32u widthbytes, Ncv32u height, cudaStream_t cuStream);
501
502
503/**
504* NCVVector (1D)
505*/
506template <class T>
508{
509 NCVVector(const NCVVector &) = delete;
510
511public:
512
514 {
515 clear();
516 }
517
518 virtual ~NCVVector() {}
519
520 void clear()
521 {
522 _ptr = nullptr;
523 _length = 0;
524 _memtype = NCVMemoryTypeNone;
525 }
526
527 NCVStatus copySolid(NCVVector<T> &dst, cudaStream_t cuStream, std::size_t howMuch=0) const
528 {
529 if (howMuch == 0)
530 {
531 ncvAssertReturn(dst._length == this->_length, NCV_MEM_COPY_ERROR);
532 howMuch = this->_length * sizeof(T);
533 }
534 else
535 {
536 ncvAssertReturn(dst._length * sizeof(T) >= howMuch &&
537 this->_length * sizeof(T) >= howMuch &&
538 howMuch > 0, NCV_MEM_COPY_ERROR);
539 }
540 ncvAssertReturn((this->_ptr != nullptr || this->_memtype == NCVMemoryTypeNone) &&
541 (dst._ptr != nullptr || dst._memtype == NCVMemoryTypeNone), NCV_NULL_PTR);
542
543 NCVStatus ncvStat = NCV_SUCCESS;
544 if (this->_memtype != NCVMemoryTypeNone)
545 {
546 ncvStat = memSegCopyHelper(dst._ptr, dst._memtype,
547 this->_ptr, this->_memtype,
548 howMuch, cuStream);
549 }
550
551 return ncvStat;
552 }
553
554 T *ptr() const {return this->_ptr;}
555 std::size_t length() const {return this->_length;}
556 NCVMemoryType memType() const {return this->_memtype;}
557
558protected:
559
561 std::size_t _length;
562 NCVMemoryType _memtype;
563};
564
565
566/**
567* NCVVectorAlloc
568*/
569template <class T>
570class NCVVectorAlloc : public NCVVector<T>
571{
572 NCVVectorAlloc() = delete;
573 NCVVectorAlloc(const NCVVectorAlloc &) = delete;
574 NCVVectorAlloc& operator=(const NCVVectorAlloc<T>&) = delete;
575
576public:
577
579 :
580 allocator(allocator)
581 {
582 NCVStatus ncvStat;
583
584 this->clear();
585 this->allocatedMem.clear();
586
587 ncvStat = allocator.alloc(this->allocatedMem, length * sizeof(T));
588 ncvAssertPrintReturn(ncvStat == NCV_SUCCESS, "NCVVectorAlloc ctor:: alloc failed", );
589
590 this->_ptr = (T *)this->allocatedMem.begin.ptr;
591 this->_length = length;
592 this->_memtype = this->allocatedMem.begin.memtype;
593 }
594
596 {
597 NCVStatus ncvStat;
598
599 ncvStat = allocator.dealloc(this->allocatedMem);
600 ncvAssertPrintCheck(ncvStat == NCV_SUCCESS, "NCVVectorAlloc dtor:: dealloc failed");
601
602 this->clear();
603 }
604
605 NcvBool isMemAllocated() const
606 {
607 return (this->allocatedMem.begin.ptr != nullptr) || (this->allocator.isCounting());
608 }
609
611 {
612 return allocator.alignment();
613 }
614
616 {
617 return allocatedMem;
618 }
619
620private:
621 INCVMemAllocator &allocator;
622 NCVMemSegment allocatedMem;
623};
624
625
626/**
627* NCVVectorReuse
628*/
629template <class T>
630class NCVVectorReuse : public NCVVector<T>
631{
632 NCVVectorReuse() = delete;
633 NCVVectorReuse(const NCVVectorReuse &) = delete;
634
635public:
636
637 explicit NCVVectorReuse(const NCVMemSegment &memSegment)
638 {
639 this->bReused = false;
640 this->clear();
641
642 this->_length = memSegment.size / sizeof(T);
643 this->_ptr = (T *)memSegment.begin.ptr;
644 this->_memtype = memSegment.begin.memtype;
645
646 this->bReused = true;
647 }
648
649 NCVVectorReuse(const NCVMemSegment &memSegment, Ncv32u length)
650 {
651 this->bReused = false;
652 this->clear();
653
654 ncvAssertPrintReturn(length * sizeof(T) <= memSegment.size, \
655 "NCVVectorReuse ctor:: memory binding failed due to size mismatch", );
656
657 this->_length = length;
658 this->_ptr = (T *)memSegment.begin.ptr;
659 this->_memtype = memSegment.begin.memtype;
660
661 this->bReused = true;
662 }
663
664 NcvBool isMemReused() const
665 {
666 return this->bReused;
667 }
668
669private:
670
671 NcvBool bReused;
672};
673
674
675/**
676* NCVMatrix (2D)
677*/
678template <class T>
680{
681 NCVMatrix(const NCVMatrix &) = delete;
682
683public:
684
686 {
687 clear();
688 }
689
690 virtual ~NCVMatrix() {}
691
692 void clear()
693 {
694 _ptr = nullptr;
695 _pitch = 0;
696 _width = 0;
697 _height = 0;
698 _memtype = NCVMemoryTypeNone;
699 }
700
701 Ncv32u stride() const
702 {
703 return _pitch / sizeof(T);
704 }
705
706 //a side effect of this function is that it copies everything in a single chunk, so the "padding" will be overwritten
707 NCVStatus copySolid(NCVMatrix<T> &dst, cudaStream_t cuStream, std::size_t howMuch=0) const
708 {
709 if (howMuch == 0)
710 {
711 ncvAssertReturn(dst._pitch == this->_pitch &&
712 dst._height == this->_height, NCV_MEM_COPY_ERROR);
713 howMuch = this->_pitch * this->_height;
714 }
715 else
716 {
717 ncvAssertReturn(dst._pitch * dst._height >= howMuch &&
718 this->_pitch * this->_height >= howMuch &&
719 howMuch > 0, NCV_MEM_COPY_ERROR);
720 }
721 ncvAssertReturn((this->_ptr != nullptr || this->_memtype == NCVMemoryTypeNone) &&
722 (dst._ptr != nullptr || dst._memtype == NCVMemoryTypeNone), NCV_NULL_PTR);
723
724 NCVStatus ncvStat = NCV_SUCCESS;
725 if (this->_memtype != NCVMemoryTypeNone)
726 {
727 ncvStat = memSegCopyHelper(dst._ptr, dst._memtype,
728 this->_ptr, this->_memtype,
729 howMuch, cuStream);
730 }
731
732 return ncvStat;
733 }
734
735 NCVStatus copy2D(NCVMatrix<T> &dst, NcvSize32u roi, cudaStream_t cuStream) const
736 {
737 ncvAssertReturn(this->width() >= roi.width && this->height() >= roi.height &&
738 dst.width() >= roi.width && dst.height() >= roi.height, NCV_MEM_COPY_ERROR);
739 ncvAssertReturn((this->_ptr != NULL || this->_memtype == NCVMemoryTypeNone) &&
740 (dst._ptr != NULL || dst._memtype == NCVMemoryTypeNone), NCV_NULL_PTR);
741
742 NCVStatus ncvStat = NCV_SUCCESS;
743 if (this->_memtype != NCVMemoryTypeNone)
744 {
745 ncvStat = memSegCopyHelper2D(dst._ptr, dst._pitch, dst._memtype,
746 this->_ptr, this->_pitch, this->_memtype,
747 roi.width * sizeof(T), roi.height, cuStream);
748 }
749
750 return ncvStat;
751 }
752
753 T &at(Ncv32u x, Ncv32u y) const
754 {
755 NcvBool bOutRange = (x >= this->_width || y >= this->_height);
756 ncvAssertPrintCheck(!bOutRange, "Error addressing matrix at [" << x << ", " << y << "]");
757 if (bOutRange)
758 {
759 return *this->_ptr;
760 }
761 return ((T *)((Ncv8u *)this->_ptr + y * this->_pitch))[x];
762 }
763
764 T *ptr() const {return this->_ptr;}
765 Ncv32u width() const {return this->_width;}
766 Ncv32u height() const {return this->_height;}
767 NcvSize32u size() const {return NcvSize32u(this->_width, this->_height);}
768 Ncv32u pitch() const {return this->_pitch;}
769 NCVMemoryType memType() const {return this->_memtype;}
770
771protected:
772
774 Ncv32u _width;
775 Ncv32u _height;
776 Ncv32u _pitch;
777 NCVMemoryType _memtype;
778};
779
780
781/**
782* NCVMatrixAlloc
783*/
784template <class T>
785class NCVMatrixAlloc : public NCVMatrix<T>
786{
787 NCVMatrixAlloc() = delete;
788 NCVMatrixAlloc(const NCVMatrixAlloc &) = delete;
789 NCVMatrixAlloc& operator=(const NCVMatrixAlloc &) = delete;
790public:
791
792 NCVMatrixAlloc(INCVMemAllocator &allocator, Ncv32u width, Ncv32u height, Ncv32u pitch=0)
793 :
794 allocator(allocator)
795 {
796 NCVStatus ncvStat;
797
798 this->clear();
799 this->allocatedMem.clear();
800
801 Ncv32u widthBytes = width * sizeof(T);
802 Ncv32u pitchBytes = alignUp(widthBytes, allocator.alignment());
803
804 if (pitch != 0)
805 {
806 ncvAssertPrintReturn(pitch >= pitchBytes &&
807 (pitch & (allocator.alignment() - 1)) == 0,
808 "NCVMatrixAlloc ctor:: incorrect pitch passed", );
809 pitchBytes = pitch;
810 }
811
812 Ncv32u requiredAllocSize = pitchBytes * height;
813
814 ncvStat = allocator.alloc(this->allocatedMem, requiredAllocSize);
815 ncvAssertPrintReturn(ncvStat == NCV_SUCCESS, "NCVMatrixAlloc ctor:: alloc failed", );
816
817 this->_ptr = (T *)this->allocatedMem.begin.ptr;
818 this->_width = width;
819 this->_height = height;
820 this->_pitch = pitchBytes;
821 this->_memtype = this->allocatedMem.begin.memtype;
822 }
823
825 {
826 NCVStatus ncvStat;
827
828 ncvStat = allocator.dealloc(this->allocatedMem);
829 ncvAssertPrintCheck(ncvStat == NCV_SUCCESS, "NCVMatrixAlloc dtor:: dealloc failed");
830
831 this->clear();
832 }
833
834 NcvBool isMemAllocated() const
835 {
836 return (this->allocatedMem.begin.ptr != nullptr) || (this->allocator.isCounting());
837 }
838
840 {
841 return allocator.alignment();
842 }
843
845 {
846 return allocatedMem;
847 }
848
849private:
850
851 INCVMemAllocator &allocator;
852 NCVMemSegment allocatedMem;
853};
854
855
856/**
857* NCVMatrixReuse
858*/
859template <class T>
860class NCVMatrixReuse : public NCVMatrix<T>
861{
862 NCVMatrixReuse() = delete;
863 NCVMatrixReuse(const NCVMatrixReuse &) = delete;
864
865public:
866
867 NCVMatrixReuse(const NCVMemSegment &memSegment, Ncv32u alignment, Ncv32u width, Ncv32u height, Ncv32u pitch=0, NcvBool bSkipPitchCheck=false)
868 {
869 this->bReused = false;
870 this->clear();
871
872 Ncv32u widthBytes = width * sizeof(T);
873 Ncv32u pitchBytes = alignUp(widthBytes, alignment);
874
875 if (pitch != 0)
876 {
877 if (!bSkipPitchCheck)
878 {
879 ncvAssertPrintReturn(pitch >= pitchBytes &&
880 (pitch & (alignment - 1)) == 0,
881 "NCVMatrixReuse ctor:: incorrect pitch passed", );
882 }
883 else
884 {
885 ncvAssertPrintReturn(pitch >= widthBytes, "NCVMatrixReuse ctor:: incorrect pitch passed", );
886 }
887 pitchBytes = pitch;
888 }
889
890 ncvAssertPrintReturn(pitchBytes * height <= memSegment.size, \
891 "NCVMatrixReuse ctor:: memory binding failed due to size mismatch", );
892
893 this->_width = width;
894 this->_height = height;
895 this->_pitch = pitchBytes;
896 this->_ptr = (T *)memSegment.begin.ptr;
897 this->_memtype = memSegment.begin.memtype;
898
899 this->bReused = true;
900 }
901
903 {
904 this->bReused = false;
905 this->clear();
906
907 ncvAssertPrintReturn(roi.x < mat.width() && roi.y < mat.height() && \
908 roi.x + roi.width <= mat.width() && roi.y + roi.height <= mat.height(),
909 "NCVMatrixReuse ctor:: memory binding failed due to mismatching ROI and source matrix dims", );
910
911 this->_width = roi.width;
912 this->_height = roi.height;
913 this->_pitch = mat.pitch();
914 this->_ptr = &mat.at(roi.x, roi.y);
915 this->_memtype = mat.memType();
916
917 this->bReused = true;
918 }
919
920 NcvBool isMemReused() const
921 {
922 return this->bReused;
923 }
924
925private:
926
927 NcvBool bReused;
928};
929
930/**
931* Operations with rectangles
932*/
933NCV_EXPORTS NCVStatus ncvGroupRectangles_host(NCVVector<NcvRect32u> &hypotheses, Ncv32u &numHypotheses,
934 Ncv32u minNeighbors, Ncv32f intersectEps, NCVVector<Ncv32u> *hypothesesWeights);
935
936NCV_EXPORTS NCVStatus ncvDrawRects_8u_host(Ncv8u *h_dst, Ncv32u dstStride, Ncv32u dstWidth, Ncv32u dstHeight,
937 NcvRect32u *h_rects, Ncv32u numRects, Ncv8u color);
938
939NCV_EXPORTS NCVStatus ncvDrawRects_32u_host(Ncv32u *h_dst, Ncv32u dstStride, Ncv32u dstWidth, Ncv32u dstHeight,
940 NcvRect32u *h_rects, Ncv32u numRects, Ncv32u color);
941
942NCV_EXPORTS NCVStatus ncvDrawRects_8u_device(Ncv8u *d_dst, Ncv32u dstStride, Ncv32u dstWidth, Ncv32u dstHeight,
943 NcvRect32u *d_rects, Ncv32u numRects, Ncv8u color, cudaStream_t cuStream);
944
945NCV_EXPORTS NCVStatus ncvDrawRects_32u_device(Ncv32u *d_dst, Ncv32u dstStride, Ncv32u dstWidth, Ncv32u dstHeight,
946 NcvRect32u *d_rects, Ncv32u numRects, Ncv32u color, cudaStream_t cuStream);
947
948#define CLAMP(x,a,b) ( (x) > (b) ? (b) : ( (x) < (a) ? (a) : (x) ) )
949#define CLAMP_TOP(x, a) (((x) > (a)) ? (a) : (x))
950#define CLAMP_BOTTOM(x, a) (((x) < (a)) ? (a) : (x))
951#define CLAMP_0_255(x) CLAMP(x,0,255)
952
953#define SUB_BEGIN(type, name) struct { __inline type name
954#define SUB_END(name) } name;
955#define SUB_CALL(name) name.name
956
957#define SQR(x) ((x)*(x))
958
959#define ncvSafeMatAlloc(name, type, alloc, width, height, err) \
960 NCVMatrixAlloc<type> name(alloc, width, height); \
961 ncvAssertReturn(name.isMemAllocated(), err);
962
963#endif // PCL_GPU_PEOPLE__NCV_HPP_
INCVMemAllocator (Interface)
Definition: NCV.hpp:403
virtual NcvBool isInitialized() const =0
virtual NCVStatus alloc(NCVMemSegment &seg, std::size_t size)=0
virtual NCVStatus dealloc(NCVMemSegment &seg)=0
virtual std::size_t maxSize() const =0
virtual ~INCVMemAllocator()=0
Definition: NCV.hpp:418
virtual Ncv32u alignment() const =0
virtual NcvBool isCounting() const =0
virtual NCVMemoryType memType() const =0
NCVMatrixAlloc.
Definition: NCV.hpp:786
NcvBool isMemAllocated() const
Definition: NCV.hpp:834
Ncv32u getAllocatorsAlignment() const
Definition: NCV.hpp:839
NCVMatrixAlloc(INCVMemAllocator &allocator, Ncv32u width, Ncv32u height, Ncv32u pitch=0)
Definition: NCV.hpp:792
~NCVMatrixAlloc()
Definition: NCV.hpp:824
NCVMemSegment getSegment() const
Definition: NCV.hpp:844
NCVMatrix (2D)
Definition: NCV.hpp:680
NCVMemoryType memType() const
Definition: NCV.hpp:769
NCVMemoryType _memtype
Definition: NCV.hpp:777
T * _ptr
Definition: NCV.hpp:773
Ncv32u _pitch
Definition: NCV.hpp:776
T & at(Ncv32u x, Ncv32u y) const
Definition: NCV.hpp:753
void clear()
Definition: NCV.hpp:692
NCVStatus copySolid(NCVMatrix< T > &dst, cudaStream_t cuStream, std::size_t howMuch=0) const
Definition: NCV.hpp:707
NcvSize32u size() const
Definition: NCV.hpp:767
T * ptr() const
Definition: NCV.hpp:764
Ncv32u width() const
Definition: NCV.hpp:765
Ncv32u _height
Definition: NCV.hpp:775
Ncv32u _width
Definition: NCV.hpp:774
NCVMatrix()
Definition: NCV.hpp:685
NCVStatus copy2D(NCVMatrix< T > &dst, NcvSize32u roi, cudaStream_t cuStream) const
Definition: NCV.hpp:735
virtual ~NCVMatrix()
Definition: NCV.hpp:690
Ncv32u height() const
Definition: NCV.hpp:766
Ncv32u stride() const
Definition: NCV.hpp:701
Ncv32u pitch() const
Definition: NCV.hpp:768
NCVMatrixReuse.
Definition: NCV.hpp:861
NCVMatrixReuse(const NCVMemSegment &memSegment, Ncv32u alignment, Ncv32u width, Ncv32u height, Ncv32u pitch=0, NcvBool bSkipPitchCheck=false)
Definition: NCV.hpp:867
NCVMatrixReuse(const NCVMatrix< T > &mat, NcvRect32u roi)
Definition: NCV.hpp:902
NcvBool isMemReused() const
Definition: NCV.hpp:920
NCVMemNativeAllocator.
Definition: NCV.hpp:462
virtual ~NCVMemNativeAllocator()
NCVMemNativeAllocator(NCVMemoryType memT, Ncv32u alignment)
virtual NCVStatus alloc(NCVMemSegment &seg, std::size_t size)
virtual NcvBool isInitialized() const
virtual Ncv32u alignment() const
virtual NCVStatus dealloc(NCVMemSegment &seg)
virtual std::size_t maxSize() const
virtual NcvBool isCounting() const
virtual NCVMemoryType memType() const
NCVMemStackAllocator.
Definition: NCV.hpp:425
NCVMemStackAllocator(NCVMemoryType memT, std::size_t capacity, Ncv32u alignment, void *reusePtr=nullptr)
virtual ~NCVMemStackAllocator()
virtual NcvBool isCounting() const
virtual NCVStatus alloc(NCVMemSegment &seg, std::size_t size)
virtual NcvBool isInitialized() const
virtual NCVStatus dealloc(NCVMemSegment &seg)
virtual std::size_t maxSize() const
NCVMemStackAllocator(Ncv32u alignment)
virtual NCVMemoryType memType() const
virtual Ncv32u alignment() const
NCVVectorAlloc.
Definition: NCV.hpp:571
Ncv32u getAllocatorsAlignment() const
Definition: NCV.hpp:610
~NCVVectorAlloc()
Definition: NCV.hpp:595
NcvBool isMemAllocated() const
Definition: NCV.hpp:605
NCVVectorAlloc(INCVMemAllocator &allocator, Ncv32u length)
Definition: NCV.hpp:578
NCVMemSegment getSegment() const
Definition: NCV.hpp:615
NCVVector (1D)
Definition: NCV.hpp:508
T * _ptr
Definition: NCV.hpp:560
T * ptr() const
Definition: NCV.hpp:554
NCVVector()
Definition: NCV.hpp:513
virtual ~NCVVector()
Definition: NCV.hpp:518
std::size_t _length
Definition: NCV.hpp:561
NCVMemoryType memType() const
Definition: NCV.hpp:556
void clear()
Definition: NCV.hpp:520
NCVStatus copySolid(NCVVector< T > &dst, cudaStream_t cuStream, std::size_t howMuch=0) const
Definition: NCV.hpp:527
NCVMemoryType _memtype
Definition: NCV.hpp:562
std::size_t length() const
Definition: NCV.hpp:555
NCVVectorReuse.
Definition: NCV.hpp:631
NcvBool isMemReused() const
Definition: NCV.hpp:664
NCVVectorReuse(const NCVMemSegment &memSegment)
Definition: NCV.hpp:637
NCVVectorReuse(const NCVMemSegment &memSegment, Ncv32u length)
Definition: NCV.hpp:649
Compile-time assert namespace.
Definition: NCV.hpp:69
NCVMemPtr.
Definition: NCV.hpp:381
NCVMemoryType memtype
Definition: NCV.hpp:383
void clear()
void * ptr
Definition: NCV.hpp:382
NCVMemSegment.
Definition: NCV.hpp:392
std::size_t size
Definition: NCV.hpp:394
NCVMemPtr begin
Definition: NCV.hpp:393
__host__ __device__ NcvPoint2D32s()
Definition: NCV.hpp:179
Ncv32s x
Point X.
Definition: NCV.hpp:177
__host__ __device__ NcvPoint2D32s(Ncv32s x, Ncv32s y)
Definition: NCV.hpp:180
Ncv32s y
Point Y.
Definition: NCV.hpp:178
Ncv32u x
Point X.
Definition: NCV.hpp:185
Ncv32u y
Point Y.
Definition: NCV.hpp:186
__host__ __device__ NcvPoint2D32u()
Definition: NCV.hpp:187
__host__ __device__ NcvPoint2D32u(Ncv32u x, Ncv32u y)
Definition: NCV.hpp:188
Ncv32s height
Rectangle height.
Definition: NCV.hpp:143
Ncv32s width
Rectangle width.
Definition: NCV.hpp:142
__host__ __device__ NcvRect32s()
Definition: NCV.hpp:144
Ncv32s x
x-coordinate of upper left corner.
Definition: NCV.hpp:140
Ncv32s y
y-coordinate of upper left corner.
Definition: NCV.hpp:141
__host__ __device__ NcvRect32s(Ncv32s x, Ncv32s y, Ncv32s width, Ncv32s height)
Definition: NCV.hpp:145
Ncv32u x
x-coordinate of upper left corner.
Definition: NCV.hpp:150
Ncv32u height
Rectangle height.
Definition: NCV.hpp:153
__host__ __device__ NcvRect32u()
Definition: NCV.hpp:154
Ncv32u width
Rectangle width.
Definition: NCV.hpp:152
Ncv32u y
y-coordinate of upper left corner.
Definition: NCV.hpp:151
__host__ __device__ NcvRect32u(Ncv32u x, Ncv32u y, Ncv32u width, Ncv32u height)
Definition: NCV.hpp:155
__host__ __device__ NcvRect8u(Ncv8u x, Ncv8u y, Ncv8u width, Ncv8u height)
Definition: NCV.hpp:135
Ncv8u y
Definition: NCV.hpp:131
__host__ __device__ NcvRect8u()
Definition: NCV.hpp:134
Ncv8u width
Definition: NCV.hpp:132
Ncv8u height
Definition: NCV.hpp:133
Ncv8u x
Definition: NCV.hpp:130
__host__ __device__ NcvSize32s()
Definition: NCV.hpp:162
Ncv32s height
Rectangle height.
Definition: NCV.hpp:161
__host__ __device__ NcvSize32s(Ncv32s width, Ncv32s height)
Definition: NCV.hpp:163
Ncv32s width
Rectangle width.
Definition: NCV.hpp:160
__host__ __device__ NcvSize32u(Ncv32u width, Ncv32u height)
Definition: NCV.hpp:171
__host__ __device__ bool operator==(const NcvSize32u &another) const
Definition: NCV.hpp:172
__host__ __device__ NcvSize32u()
Definition: NCV.hpp:170
Ncv32u height
Rectangle height.
Definition: NCV.hpp:169
Ncv32u width
Rectangle width.
Definition: NCV.hpp:168