DPC++ Runtime
Runtime libraries for oneAPI DPC++
builtins_geometric.cpp
Go to the documentation of this file.
1 //==--------- builtins_geometric.cpp - SYCL built-in geometric functions ---==//
2 //
3 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4 // See https://llvm.org/LICENSE.txt for license information.
5 // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6 //
7 //===----------------------------------------------------------------------===//
8 
9 // This file defines the host versions of functions defined
10 // in SYCL SPEC section - 4.13.6 Geometric functions.
11 
12 #include "builtins_helper.hpp"
14 
15 #include <cmath>
16 
17 namespace s = sycl;
18 namespace d = s::detail;
19 namespace cl = s::opencl;
20 
21 namespace __host_std {
22 
44 
49 
50 namespace {
51 
52 template <typename T> inline T __cross(T p0, T p1) {
53  T result(0);
54  result.x() = p0.y() * p1.z() - p0.z() * p1.y();
55  result.y() = p0.z() * p1.x() - p0.x() * p1.z();
56  result.z() = p0.x() * p1.y() - p0.y() * p1.x();
57  return result;
58 }
59 
60 template <typename T> inline void __FMul_impl(T &r, T p0, T p1) {
61  r += p0 * p1;
62 }
63 
64 template <typename T> inline T __FMul(T p0, T p1) {
65  T result = 0;
66  __FMul_impl(result, p0, p1);
67  return result;
68 }
69 
70 template <typename T>
71 inline typename std::enable_if_t<d::is_sgengeo_v<T>, T> __length(T t) {
72  return std::sqrt(__FMul(t, t));
73 }
74 
75 template <typename T>
76 inline typename std::enable_if_t<d::is_vgengeo_v<T>, typename T::element_type>
77 __length(T t) {
78  return std::sqrt(sycl_host_Dot(t, t));
79 }
80 
81 template <typename T>
82 inline typename std::enable_if_t<d::is_sgengeo_v<T>, T> __normalize(T t) {
83  T r = __length(t);
84  return t / T(r);
85 }
86 
87 template <typename T>
88 inline typename std::enable_if_t<d::is_vgengeo_v<T>, T> __normalize(T t) {
89  typename T::element_type r = __length(t);
90  return t / T(r);
91 }
92 
93 template <typename T>
94 inline typename std::enable_if_t<d::is_sgengeo_v<T>, T> __fast_length(T t) {
95  return std::sqrt(__FMul(t, t));
96 }
97 
98 template <typename T>
99 inline typename std::enable_if_t<d::is_vgengeo_v<T>, typename T::element_type>
100 __fast_length(T t) {
101  return std::sqrt(sycl_host_Dot(t, t));
102 }
103 
104 template <typename T>
105 inline typename std::enable_if_t<d::is_vgengeo_v<T>, T> __fast_normalize(T t) {
106  if (sycl_host_All(t == T(0.0f)))
107  return t;
108  typename T::element_type r = std::sqrt(sycl_host_Dot(t, t));
109  return t / T(r);
110 }
111 
112 } // namespace
113 
114 // --------------- 4.13.6 Geometric functions. Host implementations ------------
115 // cross
116 __SYCL_EXPORT s::vec<cl::cl_float, 3>
119  return __cross(p0, p1);
120 }
121 __SYCL_EXPORT s::vec<cl::cl_float, 4>
124  return __cross(p0, p1);
125 }
126 __SYCL_EXPORT s::vec<cl::cl_double, 3>
129  return __cross(p0, p1);
130 }
131 __SYCL_EXPORT s::vec<cl::cl_double, 4>
134  return __cross(p0, p1);
135 }
136 __SYCL_EXPORT s::vec<cl::cl_half, 3>
138  return __cross(p0, p1);
139 }
140 __SYCL_EXPORT s::vec<cl::cl_half, 4>
142  return __cross(p0, p1);
143 }
144 
145 // FMul
147  return __FMul(p0, p1);
148 }
150  return __FMul(p0, p1);
151 }
153  return __FMul(p0, p1);
154 }
155 
156 // Dot
158  s::cl_float)
160  s::cl_double)
162  s::cl_half)
163 
164 // length
165 __SYCL_EXPORT cl::cl_float sycl_host_length(cl::cl_float p) {
166  return __length(p);
167 }
169  return __length(p);
170 }
172  return __length(p);
173 }
175  return __length(p);
176 }
178  return __length(p);
179 }
181  return __length(p);
182 }
184  return __length(p);
185 }
187  return __length(p);
188 }
190  return __length(p);
191 }
193  return __length(p);
194 }
196  return __length(p);
197 }
199  return __length(p);
200 }
202  return __length(p);
203 }
205  return __length(p);
206 }
208  return __length(p);
209 }
210 
211 // distance
213  cl::cl_float p1) {
214  return sycl_host_length(p0 - p1);
215 }
218  return sycl_host_length(p0 - p1);
219 }
222  return sycl_host_length(p0 - p1);
223 }
226  return sycl_host_length(p0 - p1);
227 }
230  return sycl_host_length(p0 - p1);
231 }
233  cl::cl_double p1) {
234  return sycl_host_length(p0 - p1);
235 }
238  return sycl_host_length(p0 - p1);
239 }
242  return sycl_host_length(p0 - p1);
243 }
246  return sycl_host_length(p0 - p1);
247 }
250  return sycl_host_length(p0 - p1);
251 }
253  return sycl_host_length(p0 - p1);
254 }
257  return sycl_host_length(p0 - p1);
258 }
261  return sycl_host_length(p0 - p1);
262 }
265  return sycl_host_length(p0 - p1);
266 }
269  return sycl_host_length(p0 - p1);
270 }
271 
272 // normalize
274  return __normalize(p);
275 }
277  return __normalize(p);
278 }
279 __SYCL_EXPORT s::vec<cl::cl_float, 2>
281  return __normalize(p);
282 }
283 __SYCL_EXPORT s::vec<cl::cl_float, 3>
285  return __normalize(p);
286 }
287 __SYCL_EXPORT s::vec<cl::cl_float, 4>
289  return __normalize(p);
290 }
292  return __normalize(p);
293 }
295  return __normalize(p);
296 }
297 __SYCL_EXPORT s::vec<cl::cl_double, 2>
299  return __normalize(p);
300 }
301 __SYCL_EXPORT s::vec<cl::cl_double, 3>
303  return __normalize(p);
304 }
305 __SYCL_EXPORT s::vec<cl::cl_double, 4>
307  return __normalize(p);
308 }
310  return __normalize(p);
311 }
312 __SYCL_EXPORT s::vec<cl::cl_half, 2>
314  return __normalize(p);
315 }
316 __SYCL_EXPORT s::vec<cl::cl_half, 3>
318  return __normalize(p);
319 }
320 __SYCL_EXPORT s::vec<cl::cl_half, 4>
322  return __normalize(p);
323 }
324 
325 // fast_length
327  return __fast_length(p);
328 }
330  return __fast_length(p);
331 }
333  return __fast_length(p);
334 }
336  return __fast_length(p);
337 }
339  return __fast_length(p);
340 }
341 
342 // fast_normalize
344  if (p == 0.0f)
345  return p;
347  return p / r;
348 }
350  return __fast_normalize(p);
351 }
352 __SYCL_EXPORT s::vec<cl::cl_float, 2>
354  return __fast_normalize(p);
355 }
356 __SYCL_EXPORT s::vec<cl::cl_float, 3>
358  return __fast_normalize(p);
359 }
360 __SYCL_EXPORT s::vec<cl::cl_float, 4>
362  return __fast_normalize(p);
363 }
364 
365 // fast_distance
367  cl::cl_float p1) {
368  return sycl_host_fast_length(p0 - p1);
369 }
372  return sycl_host_fast_length(p0 - p1);
373 }
376  return sycl_host_fast_length(p0 - p1);
377 }
380  return sycl_host_fast_length(p0 - p1);
381 }
384  return sycl_host_fast_length(p0 - p1);
385 }
386 
387 } // namespace __host_std
#define __NOEXC
defined(__INTEL_PREVIEW_BREAKING_CHANGES)
Definition: types.hpp:346
cl::cl_float sycl_host_distance(cl::cl_float p0, cl::cl_float p1)
cl::cl_float sycl_host_fast_distance(cl::cl_float p0, cl::cl_float p1)
cl::cl_float sycl_host_FMul(cl::cl_float p0, cl::cl_float p1)
cl::cl_double sycl_host_length(cl::cl_double p)
cl::cl_float sycl_host_fast_length(cl::cl_float p)
cl::cl_int sycl_host_All(s::vec< int, 1 >)
cl::cl_float sycl_host_normalize(cl::cl_float p)
s::vec< cl::cl_float, 3 > sycl_host_cross(s::vec< cl::cl_float, 3 > p0, s::vec< cl::cl_float, 3 > p1) __NOEXC
MAKE_GEO_1V_2V_RS(sycl_host_Dot, __FMul_impl, s::cl_float, s::cl_float, s::cl_float) MAKE_GEO_1V_2V_RS(sycl_host_Dot
cl::cl_float sycl_host_fast_normalize(cl::cl_float p)
cl::cl_float sycl_host_Dot(s::vec< float, 1 >, s::vec< float, 1 >)
std::int32_t cl_int
Definition: aliases.hpp:134
std::enable_if_t< __FAST_MATH_GENFLOAT(T), T > sqrt(T x)
std::conditional_t< std::is_same_v< ElementType, half >, sycl::detail::half_impl::BIsRepresentationT, ElementType > element_type
Definition: multi_ptr.hpp:752
Definition: access.hpp:18