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"
13 #include <sycl/detail/stl_type_traits.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 
23 __SYCL_EXPORT cl::cl_float sycl_host_Dot(s::vec<float, 1>, s::vec<float, 1>);
24 __SYCL_EXPORT cl::cl_float sycl_host_Dot(s::vec<cl::cl_float, 2>,
25  s::vec<cl::cl_float, 2>);
26 __SYCL_EXPORT cl::cl_float sycl_host_Dot(s::vec<cl::cl_float, 3>,
27  s::vec<cl::cl_float, 3>);
28 __SYCL_EXPORT cl::cl_float sycl_host_Dot(s::vec<cl::cl_float, 4>,
29  s::vec<cl::cl_float, 4>);
30 __SYCL_EXPORT cl::cl_double sycl_host_Dot(s::vec<double, 1>, s::vec<double, 1>);
31 __SYCL_EXPORT cl::cl_double sycl_host_Dot(s::vec<cl::cl_double, 2>,
32  s::vec<cl::cl_double, 2>);
33 __SYCL_EXPORT cl::cl_double sycl_host_Dot(s::vec<cl::cl_double, 3>,
34  s::vec<cl::cl_double, 3>);
35 __SYCL_EXPORT cl::cl_double sycl_host_Dot(s::vec<cl::cl_double, 4>,
36  s::vec<cl::cl_double, 4>);
37 __SYCL_EXPORT cl::cl_half sycl_host_Dot(s::vec<s::half, 1>, s::vec<s::half, 1>);
38 __SYCL_EXPORT cl::cl_half sycl_host_Dot(s::vec<cl::cl_half, 2>,
39  s::vec<cl::cl_half, 2>);
40 __SYCL_EXPORT cl::cl_half sycl_host_Dot(s::vec<cl::cl_half, 3>,
41  s::vec<cl::cl_half, 3>);
42 __SYCL_EXPORT cl::cl_half sycl_host_Dot(s::vec<cl::cl_half, 4>,
43  s::vec<cl::cl_half, 4>);
44 
45 __SYCL_EXPORT cl::cl_int sycl_host_All(s::vec<int, 1>);
46 __SYCL_EXPORT cl::cl_int sycl_host_All(s::vec<cl::cl_int, 2>);
47 __SYCL_EXPORT cl::cl_int sycl_host_All(s::vec<cl::cl_int, 3>);
48 __SYCL_EXPORT cl::cl_int sycl_host_All(s::vec<cl::cl_int, 4>);
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 sycl::detail::enable_if_t<d::is_sgengeo<T>::value, T>
72 __length(T t) {
73  return std::sqrt(__FMul(t, t));
74 }
75 
76 template <typename T>
77 inline typename sycl::detail::enable_if_t<d::is_vgengeo<T>::value,
78  typename T::element_type>
79 __length(T t) {
80  return std::sqrt(sycl_host_Dot(t, t));
81 }
82 
83 template <typename T>
84 inline typename sycl::detail::enable_if_t<d::is_sgengeo<T>::value, T>
85 __normalize(T t) {
86  T r = __length(t);
87  return t / T(r);
88 }
89 
90 template <typename T>
91 inline typename sycl::detail::enable_if_t<d::is_vgengeo<T>::value, T>
92 __normalize(T t) {
93  typename T::element_type r = __length(t);
94  return t / T(r);
95 }
96 
97 template <typename T>
98 inline typename sycl::detail::enable_if_t<d::is_sgengeo<T>::value, T>
99 __fast_length(T t) {
100  return std::sqrt(__FMul(t, t));
101 }
102 
103 template <typename T>
104 inline typename sycl::detail::enable_if_t<d::is_vgengeo<T>::value,
105  typename T::element_type>
106 __fast_length(T t) {
107  return std::sqrt(sycl_host_Dot(t, t));
108 }
109 
110 template <typename T>
111 inline typename sycl::detail::enable_if_t<d::is_vgengeo<T>::value, T>
112 __fast_normalize(T t) {
113  if (sycl_host_All(t == T(0.0f)))
114  return t;
115  typename T::element_type r = std::sqrt(sycl_host_Dot(t, t));
116  return t / T(r);
117 }
118 
119 } // namespace
120 
121 // --------------- 4.13.6 Geometric functions. Host implementations ------------
122 // cross
123 __SYCL_EXPORT s::vec<cl::cl_float, 3>
124 sycl_host_cross(s::vec<cl::cl_float, 3> p0,
125  s::vec<cl::cl_float, 3> p1) __NOEXC {
126  return __cross(p0, p1);
127 }
128 __SYCL_EXPORT s::vec<cl::cl_float, 4>
129 sycl_host_cross(s::vec<cl::cl_float, 4> p0,
130  s::vec<cl::cl_float, 4> p1) __NOEXC {
131  return __cross(p0, p1);
132 }
133 __SYCL_EXPORT s::vec<cl::cl_double, 3>
134 sycl_host_cross(s::vec<cl::cl_double, 3> p0,
135  s::vec<cl::cl_double, 3> p1) __NOEXC {
136  return __cross(p0, p1);
137 }
138 __SYCL_EXPORT s::vec<cl::cl_double, 4>
139 sycl_host_cross(s::vec<cl::cl_double, 4> p0,
140  s::vec<cl::cl_double, 4> p1) __NOEXC {
141  return __cross(p0, p1);
142 }
143 __SYCL_EXPORT s::vec<cl::cl_half, 3>
144 sycl_host_cross(s::vec<cl::cl_half, 3> p0, s::vec<cl::cl_half, 3> p1) __NOEXC {
145  return __cross(p0, p1);
146 }
147 __SYCL_EXPORT s::vec<cl::cl_half, 4>
148 sycl_host_cross(s::vec<cl::cl_half, 4> p0, s::vec<cl::cl_half, 4> p1) __NOEXC {
149  return __cross(p0, p1);
150 }
151 
152 // FMul
153 __SYCL_EXPORT cl::cl_float sycl_host_FMul(cl::cl_float p0, cl::cl_float p1) {
154  return __FMul(p0, p1);
155 }
156 __SYCL_EXPORT cl::cl_double sycl_host_FMul(cl::cl_double p0, cl::cl_double p1) {
157  return __FMul(p0, p1);
158 }
159 __SYCL_EXPORT cl::cl_float sycl_host_FMul(cl::cl_half p0, cl::cl_half p1) {
160  return __FMul(p0, p1);
161 }
162 
163 // Dot
164 MAKE_GEO_1V_2V_RS(sycl_host_Dot, __FMul_impl, s::cl_float, s::cl_float,
165  s::cl_float)
166 MAKE_GEO_1V_2V_RS(sycl_host_Dot, __FMul_impl, s::cl_double, s::cl_double,
167  s::cl_double)
168 MAKE_GEO_1V_2V_RS(sycl_host_Dot, __FMul_impl, s::cl_half, s::cl_half,
169  s::cl_half)
170 
171 // length
172 __SYCL_EXPORT cl::cl_float sycl_host_length(cl::cl_float p) {
173  return __length(p);
174 }
175 __SYCL_EXPORT cl::cl_double sycl_host_length(cl::cl_double p) {
176  return __length(p);
177 }
178 __SYCL_EXPORT cl::cl_half sycl_host_length(cl::cl_half p) {
179  return __length(p);
180 }
181 __SYCL_EXPORT cl::cl_float sycl_host_length(s::vec<float, 1> p) {
182  return __length(p);
183 }
184 __SYCL_EXPORT cl::cl_float sycl_host_length(s::vec<cl::cl_float, 2> p) {
185  return __length(p);
186 }
187 __SYCL_EXPORT cl::cl_float sycl_host_length(s::vec<cl::cl_float, 3> p) {
188  return __length(p);
189 }
190 __SYCL_EXPORT cl::cl_float sycl_host_length(s::vec<cl::cl_float, 4> p) {
191  return __length(p);
192 }
193 __SYCL_EXPORT cl::cl_double sycl_host_length(s::vec<double, 1> p) {
194  return __length(p);
195 }
196 __SYCL_EXPORT cl::cl_double sycl_host_length(s::vec<cl::cl_double, 2> p) {
197  return __length(p);
198 }
199 __SYCL_EXPORT cl::cl_double sycl_host_length(s::vec<cl::cl_double, 3> p) {
200  return __length(p);
201 }
202 __SYCL_EXPORT cl::cl_double sycl_host_length(s::vec<cl::cl_double, 4> p) {
203  return __length(p);
204 }
205 __SYCL_EXPORT cl::cl_half sycl_host_length(s::vec<s::half, 1> p) {
206  return __length(p);
207 }
208 __SYCL_EXPORT cl::cl_half sycl_host_length(s::vec<cl::cl_half, 2> p) {
209  return __length(p);
210 }
211 __SYCL_EXPORT cl::cl_half sycl_host_length(s::vec<cl::cl_half, 3> p) {
212  return __length(p);
213 }
214 __SYCL_EXPORT cl::cl_half sycl_host_length(s::vec<cl::cl_half, 4> p) {
215  return __length(p);
216 }
217 
218 // distance
219 __SYCL_EXPORT cl::cl_float sycl_host_distance(cl::cl_float p0,
220  cl::cl_float p1) {
221  return sycl_host_length(p0 - p1);
222 }
223 __SYCL_EXPORT cl::cl_float sycl_host_distance(s::vec<float, 1> p0,
224  s::vec<float, 1> p1) {
225  return sycl_host_length(p0 - p1);
226 }
227 __SYCL_EXPORT cl::cl_float sycl_host_distance(s::vec<cl::cl_float, 2> p0,
228  s::vec<cl::cl_float, 2> p1) {
229  return sycl_host_length(p0 - p1);
230 }
231 __SYCL_EXPORT cl::cl_float sycl_host_distance(s::vec<cl::cl_float, 3> p0,
232  s::vec<cl::cl_float, 3> p1) {
233  return sycl_host_length(p0 - p1);
234 }
235 __SYCL_EXPORT cl::cl_float sycl_host_distance(s::vec<cl::cl_float, 4> p0,
236  s::vec<cl::cl_float, 4> p1) {
237  return sycl_host_length(p0 - p1);
238 }
239 __SYCL_EXPORT cl::cl_double sycl_host_distance(cl::cl_double p0,
240  cl::cl_double p1) {
241  return sycl_host_length(p0 - p1);
242 }
243 __SYCL_EXPORT cl::cl_float sycl_host_distance(s::vec<double, 1> p0,
244  s::vec<double, 1> p1) {
245  return sycl_host_length(p0 - p1);
246 }
247 __SYCL_EXPORT cl::cl_double sycl_host_distance(s::vec<cl::cl_double, 2> p0,
248  s::vec<cl::cl_double, 2> p1) {
249  return sycl_host_length(p0 - p1);
250 }
251 __SYCL_EXPORT cl::cl_double sycl_host_distance(s::vec<cl::cl_double, 3> p0,
252  s::vec<cl::cl_double, 3> p1) {
253  return sycl_host_length(p0 - p1);
254 }
255 __SYCL_EXPORT cl::cl_double sycl_host_distance(s::vec<cl::cl_double, 4> p0,
256  s::vec<cl::cl_double, 4> p1) {
257  return sycl_host_length(p0 - p1);
258 }
259 __SYCL_EXPORT cl::cl_half sycl_host_distance(cl::cl_half p0, cl::cl_half p1) {
260  return sycl_host_length(p0 - p1);
261 }
262 __SYCL_EXPORT cl::cl_float sycl_host_distance(s::vec<s::half, 1> p0,
263  s::vec<s::half, 1> p1) {
264  return sycl_host_length(p0 - p1);
265 }
266 __SYCL_EXPORT cl::cl_half sycl_host_distance(s::vec<cl::cl_half, 2> p0,
267  s::vec<cl::cl_half, 2> p1) {
268  return sycl_host_length(p0 - p1);
269 }
270 __SYCL_EXPORT cl::cl_half sycl_host_distance(s::vec<cl::cl_half, 3> p0,
271  s::vec<cl::cl_half, 3> p1) {
272  return sycl_host_length(p0 - p1);
273 }
274 __SYCL_EXPORT cl::cl_half sycl_host_distance(s::vec<cl::cl_half, 4> p0,
275  s::vec<cl::cl_half, 4> p1) {
276  return sycl_host_length(p0 - p1);
277 }
278 
279 // normalize
280 __SYCL_EXPORT cl::cl_float sycl_host_normalize(cl::cl_float p) {
281  return __normalize(p);
282 }
283 __SYCL_EXPORT cl::cl_float sycl_host_normalize(s::vec<float, 1> p) {
284  return __normalize(p);
285 }
286 __SYCL_EXPORT s::vec<cl::cl_float, 2>
287 sycl_host_normalize(s::vec<cl::cl_float, 2> p) {
288  return __normalize(p);
289 }
290 __SYCL_EXPORT s::vec<cl::cl_float, 3>
291 sycl_host_normalize(s::vec<cl::cl_float, 3> p) {
292  return __normalize(p);
293 }
294 __SYCL_EXPORT s::vec<cl::cl_float, 4>
295 sycl_host_normalize(s::vec<cl::cl_float, 4> p) {
296  return __normalize(p);
297 }
298 __SYCL_EXPORT cl::cl_double sycl_host_normalize(cl::cl_double p) {
299  return __normalize(p);
300 }
301 __SYCL_EXPORT cl::cl_double sycl_host_normalize(s::vec<double, 1> p) {
302  return __normalize(p);
303 }
304 __SYCL_EXPORT s::vec<cl::cl_double, 2>
305 sycl_host_normalize(s::vec<cl::cl_double, 2> p) {
306  return __normalize(p);
307 }
308 __SYCL_EXPORT s::vec<cl::cl_double, 3>
309 sycl_host_normalize(s::vec<cl::cl_double, 3> p) {
310  return __normalize(p);
311 }
312 __SYCL_EXPORT s::vec<cl::cl_double, 4>
313 sycl_host_normalize(s::vec<cl::cl_double, 4> p) {
314  return __normalize(p);
315 }
316 __SYCL_EXPORT cl::cl_half sycl_host_normalize(cl::cl_half p) {
317  return __normalize(p);
318 }
319 __SYCL_EXPORT s::vec<cl::cl_half, 2>
320 sycl_host_normalize(s::vec<cl::cl_half, 2> p) {
321  return __normalize(p);
322 }
323 __SYCL_EXPORT s::vec<cl::cl_half, 3>
324 sycl_host_normalize(s::vec<cl::cl_half, 3> p) {
325  return __normalize(p);
326 }
327 __SYCL_EXPORT s::vec<cl::cl_half, 4>
328 sycl_host_normalize(s::vec<cl::cl_half, 4> p) {
329  return __normalize(p);
330 }
331 
332 // fast_length
333 __SYCL_EXPORT cl::cl_float sycl_host_fast_length(cl::cl_float p) {
334  return __fast_length(p);
335 }
336 __SYCL_EXPORT cl::cl_float sycl_host_fast_length(s::vec<float, 1> p) {
337  return __fast_length(p);
338 }
339 __SYCL_EXPORT cl::cl_float sycl_host_fast_length(s::vec<cl::cl_float, 2> p) {
340  return __fast_length(p);
341 }
342 __SYCL_EXPORT cl::cl_float sycl_host_fast_length(s::vec<cl::cl_float, 3> p) {
343  return __fast_length(p);
344 }
345 __SYCL_EXPORT cl::cl_float sycl_host_fast_length(s::vec<cl::cl_float, 4> p) {
346  return __fast_length(p);
347 }
348 
349 // fast_normalize
350 __SYCL_EXPORT cl::cl_float sycl_host_fast_normalize(cl::cl_float p) {
351  if (p == 0.0f)
352  return p;
353  cl::cl_float r = std::sqrt(sycl_host_FMul(p, p));
354  return p / r;
355 }
356 __SYCL_EXPORT cl::cl_float sycl_host_fast_normalize(s::vec<float, 1> p) {
357  return __fast_normalize(p);
358 }
359 __SYCL_EXPORT s::vec<cl::cl_float, 2>
360 sycl_host_fast_normalize(s::vec<cl::cl_float, 2> p) {
361  return __fast_normalize(p);
362 }
363 __SYCL_EXPORT s::vec<cl::cl_float, 3>
364 sycl_host_fast_normalize(s::vec<cl::cl_float, 3> p) {
365  return __fast_normalize(p);
366 }
367 __SYCL_EXPORT s::vec<cl::cl_float, 4>
368 sycl_host_fast_normalize(s::vec<cl::cl_float, 4> p) {
369  return __fast_normalize(p);
370 }
371 
372 // fast_distance
373 __SYCL_EXPORT cl::cl_float sycl_host_fast_distance(cl::cl_float p0,
374  cl::cl_float p1) {
375  return sycl_host_fast_length(p0 - p1);
376 }
377 __SYCL_EXPORT cl::cl_float sycl_host_fast_distance(s::vec<float, 1> p0,
378  s::vec<float, 1> p1) {
379  return sycl_host_fast_length(p0 - p1);
380 }
381 __SYCL_EXPORT cl::cl_float sycl_host_fast_distance(s::vec<cl::cl_float, 2> p0,
382  s::vec<cl::cl_float, 2> p1) {
383  return sycl_host_fast_length(p0 - p1);
384 }
385 __SYCL_EXPORT cl::cl_float sycl_host_fast_distance(s::vec<cl::cl_float, 3> p0,
386  s::vec<cl::cl_float, 3> p1) {
387  return sycl_host_fast_length(p0 - p1);
388 }
389 __SYCL_EXPORT cl::cl_float sycl_host_fast_distance(s::vec<cl::cl_float, 4> p0,
390  s::vec<cl::cl_float, 4> p1) {
391  return sycl_host_fast_length(p0 - p1);
392 }
393 
394 } // namespace __host_std
__host_std::sycl_host_Dot
cl::cl_float sycl_host_Dot(s::vec< float, 1 >, s::vec< float, 1 >)
__NOEXC
#define __NOEXC
Definition: builtins.hpp:18
__host_std::sycl_host_fast_distance
cl::cl_float sycl_host_fast_distance(cl::cl_float p0, cl::cl_float p1)
Definition: builtins_geometric.cpp:373
__host_std
Definition: builtins.hpp:106
sycl
---— Error handling, matching OpenCL plugin semantics.
Definition: access.hpp:14
__host_std::__FMul_impl
__FMul_impl
Definition: builtins_geometric.cpp:166
sycl::_V1::cl_half
half cl_half
Definition: aliases.hpp:112
sycl::_V1::cl_int
std::int32_t cl_int
Definition: aliases.hpp:108
sycl::_V1::ext::intel::esimd::sqrt
__ESIMD_API simd< T, N > sqrt(simd< T, N > src, Sat sat={})
Square root.
Definition: math.hpp:389
__host_std::sycl_host_All
cl::cl_int sycl_host_All(s::vec< int, 1 >)
stl_type_traits.hpp
sycl::_V1::cl_double
double cl_double
Definition: aliases.hpp:114
sycl::_V1::cl_float
float cl_float
Definition: aliases.hpp:113
builtins_helper.hpp
__host_std::sycl_host_length
cl::cl_double sycl_host_length(cl::cl_double p)
Definition: builtins_geometric.cpp:175
__host_std::sycl_host_distance
cl::cl_float sycl_host_distance(cl::cl_float p0, cl::cl_float p1)
Definition: builtins_geometric.cpp:219
__host_std::sycl_host_normalize
cl::cl_float sycl_host_normalize(cl::cl_float p)
Definition: builtins_geometric.cpp:280
__host_std::MAKE_GEO_1V_2V_RS
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
__host_std::sycl_host_fast_length
cl::cl_float sycl_host_fast_length(cl::cl_float p)
Definition: builtins_geometric.cpp:333
sycl::_V1::image_channel_order::r
@ r
__host_std::sycl_host_fast_normalize
cl::cl_float sycl_host_fast_normalize(cl::cl_float p)
Definition: builtins_geometric.cpp:350
__host_std::sycl_host_cross
s::vec< cl::cl_float, 3 > sycl_host_cross(s::vec< cl::cl_float, 3 > p0, s::vec< cl::cl_float, 3 > p1) __NOEXC
Definition: builtins_geometric.cpp:124
__host_std::sycl_host_FMul
cl::cl_float sycl_host_FMul(cl::cl_float p0, cl::cl_float p1)
Definition: builtins_geometric.cpp:153