37 #ifndef PCL_GPU_UTILS_DEVICE_VECTOR_MATH_HPP_
38 #define PCL_GPU_UTILS_DEVICE_VECTOR_MATH_HPP_
55 #define PCL_GPU_IMPLEMENT_COMPOUND_VEC3_OP(type, scalar, op) \
56 __device__ __host__ __forceinline__ type & operator op (type & v1, const type & v2) { v1.x op v2.x; v1.y op v2.y; v1.z op v2.z; return v1; } \
57 __device__ __host__ __forceinline__ type & operator op (type & v, scalar val) { v.x op val; v.y op val; v.z op val; return v; }
59 PCL_GPU_IMPLEMENT_COMPOUND_VEC3_OP(float3,
float, -=)
60 PCL_GPU_IMPLEMENT_COMPOUND_VEC3_OP(float3,
float, +=)
61 PCL_GPU_IMPLEMENT_COMPOUND_VEC3_OP(float3,
float, *=)
63 PCL_GPU_IMPLEMENT_COMPOUND_VEC3_OP(short3,
short, -=)
65 PCL_GPU_IMPLEMENT_COMPOUND_VEC3_OP(int3,
int, +=)
67 #undef PCL_GPU_IMPLEMENT_COMPOUND_VEC3_OP
69 __device__ __host__ __forceinline__
float dot(
const float3& v1,
const float3& v2)
71 return v1.x * v2.x + v1.y * v2.y + v1.z * v2.z;
74 __device__ __host__ __forceinline__ float3
cross(
const float3& v1,
const float3& v2)
76 return make_float3(v1.y * v2.z - v1.z * v2.y, v1.z * v2.x - v1.x * v2.z, v1.x * v2.y - v1.y * v2.x);
82 __device__ __host__ __forceinline__
float dot(
const float4& v1,
const float4& v2)
84 return v1.x * v2.x + v1.y * v2.y + v1.z * v2.z + v1.w * v2.w;
90 #define PCL_GPU_IMPLEMENT_VEC_BINOP(type, scalar, op, cop) \
91 __device__ __host__ __forceinline__ type operator op (const type & v1, const type & v2) { type r = v1; r cop v2; return r; } \
92 __device__ __host__ __forceinline__ type operator op (const type & v1, scalar c) { type r = v1; r cop c; return r; }
94 PCL_GPU_IMPLEMENT_VEC_BINOP(float3,
float, -, -=)
95 PCL_GPU_IMPLEMENT_VEC_BINOP(float3,
float, +, +=)
96 PCL_GPU_IMPLEMENT_VEC_BINOP(float3,
float, *, *=)
98 PCL_GPU_IMPLEMENT_VEC_BINOP(short3,
short, -, -=)
100 PCL_GPU_IMPLEMENT_VEC_BINOP(int3,
int, +, +=)
102 #undef PCL_GPU_IMPLEMENT_VEC_BINOP
108 template<
typename T> __device__ __host__ __forceinline__
float norm(
const T& val)
110 return sqrtf(
dot(val, val));
113 template<
typename T> __host__ __device__ __forceinline__
float inverse_norm(
const T& v)
115 return rsqrtf(
dot(v, v));
118 template<
typename T> __host__ __device__ __forceinline__ T
normalized(
const T& v)
123 template<
typename T> __host__ __device__ __forceinline__ T
normalized_safe(
const T& v)
125 return (
dot(v, v) > 0) ? (v * rsqrtf(
dot(v, v))) : v;
__host__ __device__ __forceinline__ float inverse_norm(const T &v)
__host__ __device__ __forceinline__ T normalized_safe(const T &v)
__device__ __forceinline__ float3 normalized(const float3 &v)
__device__ __forceinline__ float dot(const float3 &v1, const float3 &v2)
__device__ __host__ __forceinline__ float norm(const float3 &v1, const float3 &v2)
__device__ __host__ __forceinline__ float3 cross(const float3 &v1, const float3 &v2)