Point Cloud Library (PCL) 1.13.0
Loading...
Searching...
No Matches
kernel_containers.h
1/*
2 * Software License Agreement (BSD License)
3 *
4 * Copyright (c) 2011, Willow Garage, Inc.
5 * All rights reserved.
6 *
7 * Redistribution and use in source and binary forms, with or without
8 * modification, are permitted provided that the following conditions
9 * are met:
10 *
11 * * Redistributions of source code must retain the above copyright
12 * notice, this list of conditions and the following disclaimer.
13 * * Redistributions in binary form must reproduce the above
14 * copyright notice, this list of conditions and the following
15 * disclaimer in the documentation and/or other materials provided
16 * with the distribution.
17 * * Neither the name of Willow Garage, Inc. nor the names of its
18 * contributors may be used to endorse or promote products derived
19 * from this software without specific prior written permission.
20 *
21 * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
22 * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT
23 * LIMITED TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS
24 * FOR A PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE
25 * COPYRIGHT OWNER OR CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT,
26 * INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING,
27 * BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES;
28 * LOSS OF USE, DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER
29 * CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT
30 * LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN
31 * ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF THE
32 * POSSIBILITY OF SUCH DAMAGE.
33 *
34 * Author: Anatoly Baskeheev, Itseez Ltd, (myname.mysurname@mycompany.com)
35 */
36
37#pragma once
38
39#if defined(__CUDACC__)
40#define __PCL_GPU_HOST_DEVICE__ __host__ __device__ __forceinline__
41#else
42#define __PCL_GPU_HOST_DEVICE__
43#endif
44
45#include <cstddef>
46
47namespace pcl {
48namespace gpu {
49template <typename T>
50struct DevPtr {
51 using elem_type = T;
52 const static std::size_t elem_size = sizeof(elem_type);
53
54 T* data;
55
56 __PCL_GPU_HOST_DEVICE__
57 DevPtr() : data(nullptr) {}
58
59 __PCL_GPU_HOST_DEVICE__
60 DevPtr(T* data_arg) : data(data_arg) {}
61
62 __PCL_GPU_HOST_DEVICE__ std::size_t
63 elemSize() const
64 {
65 return elem_size;
66 }
67
68 __PCL_GPU_HOST_DEVICE__
69 operator T*() { return data; }
70 __PCL_GPU_HOST_DEVICE__ operator const T*() const { return data; }
71};
72
73template <typename T>
74struct PtrSz : public DevPtr<T> {
75 __PCL_GPU_HOST_DEVICE__
76 PtrSz() : size(0) {}
77
78 __PCL_GPU_HOST_DEVICE__
79 PtrSz(T* data_arg, std::size_t size_arg) : DevPtr<T>(data_arg), size(size_arg) {}
80
81 std::size_t size;
82};
83
84template <typename T>
85struct PtrStep : public DevPtr<T> {
86 __PCL_GPU_HOST_DEVICE__
87 PtrStep() : step(0) {}
88
89 __PCL_GPU_HOST_DEVICE__
90 PtrStep(T* data_arg, std::size_t step_arg) : DevPtr<T>(data_arg), step(step_arg) {}
91
92 /** \brief stride between two consecutive rows in bytes. Step is stored always and
93 * everywhere in bytes!!! */
94 std::size_t step;
95
96 __PCL_GPU_HOST_DEVICE__ T*
97 ptr(int y = 0)
98 {
99 return (T*)((char*)DevPtr<T>::data + y * step);
100 }
101
102 __PCL_GPU_HOST_DEVICE__ const T*
103 ptr(int y = 0) const
104 {
105 return (const T*)((const char*)DevPtr<T>::data + y * step);
106 }
107
108 __PCL_GPU_HOST_DEVICE__ T&
109 operator()(int y, int x)
110 {
111 return ptr(y)[x];
112 }
113
114 __PCL_GPU_HOST_DEVICE__ const T&
115 operator()(int y, int x) const
116 {
117 return ptr(y)[x];
118 }
119};
120
121template <typename T>
122struct PtrStepSz : public PtrStep<T> {
123 __PCL_GPU_HOST_DEVICE__
124 PtrStepSz() : cols(0), rows(0) {}
125
126 __PCL_GPU_HOST_DEVICE__
127 PtrStepSz(int rows_arg, int cols_arg, T* data_arg, std::size_t step_arg)
128 : PtrStep<T>(data_arg, step_arg), cols(cols_arg), rows(rows_arg)
129 {}
130
131 int cols;
132 int rows;
133};
134} // namespace gpu
135
136namespace device {
139using pcl::gpu::PtrSz;
140} // namespace device
141} // namespace pcl
142
143#undef __PCL_GPU_HOST_DEVICE__
static const std::size_t elem_size
__PCL_GPU_HOST_DEVICE__ std::size_t elemSize() const
__PCL_GPU_HOST_DEVICE__ DevPtr(T *data_arg)
__PCL_GPU_HOST_DEVICE__ DevPtr()
__PCL_GPU_HOST_DEVICE__ PtrStep()
__PCL_GPU_HOST_DEVICE__ PtrStep(T *data_arg, std::size_t step_arg)
__PCL_GPU_HOST_DEVICE__ const T & operator()(int y, int x) const
std::size_t step
stride between two consecutive rows in bytes.
__PCL_GPU_HOST_DEVICE__ T * ptr(int y=0)
__PCL_GPU_HOST_DEVICE__ T & operator()(int y, int x)
__PCL_GPU_HOST_DEVICE__ const T * ptr(int y=0) const
__PCL_GPU_HOST_DEVICE__ PtrStepSz()
__PCL_GPU_HOST_DEVICE__ PtrStepSz(int rows_arg, int cols_arg, T *data_arg, std::size_t step_arg)
__PCL_GPU_HOST_DEVICE__ PtrSz(T *data_arg, std::size_t size_arg)
__PCL_GPU_HOST_DEVICE__ PtrSz()