arch/SYCL/MathFunctions.h
Go to the documentation of this file.
1 // This file is part of Eigen, a lightweight C++ template library
2 // for linear algebra.
3 //
4 // Mehdi Goli Codeplay Software Ltd.
5 // Ralph Potter Codeplay Software Ltd.
6 // Luke Iwanski Codeplay Software Ltd.
7 // Contact: <eigen@codeplay.com>
8 //
9 // This Source Code Form is subject to the terms of the Mozilla
10 // Public License v. 2.0. If a copy of the MPL was not distributed
11 // with this file, You can obtain one at http://mozilla.org/MPL/2.0/.
12 
13 /*****************************************************************
14  * MathFunctions.h
15  *
16  * \brief:
17  * MathFunctions
18  *
19  *****************************************************************/
20 
21 #ifndef EIGEN_MATH_FUNCTIONS_SYCL_H
22 #define EIGEN_MATH_FUNCTIONS_SYCL_H
23 // IWYU pragma: private
24 #include "../../InternalHeaderCheck.h"
25 
26 namespace Eigen {
27 
28 namespace internal {
29 
30 // Make sure this is only available when targeting a GPU: we don't want to
31 // introduce conflicts between these packet_traits definitions and the ones
32 // we'll use on the host side (SSE, AVX, ...)
33 #if defined(SYCL_DEVICE_ONLY)
34 #define SYCL_PLOG(packet_type) \
35  template <> \
36  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type plog<packet_type>(const packet_type& a) { \
37  return cl::sycl::log(a); \
38  }
39 
40 SYCL_PLOG(cl::sycl::cl_half8)
41 SYCL_PLOG(cl::sycl::cl_float4)
42 SYCL_PLOG(cl::sycl::cl_double2)
43 #undef SYCL_PLOG
44 
45 #define SYCL_PLOG1P(packet_type) \
46  template <> \
47  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type plog1p<packet_type>(const packet_type& a) { \
48  return cl::sycl::log1p(a); \
49  }
50 
51 SYCL_PLOG1P(cl::sycl::cl_half8)
52 SYCL_PLOG1P(cl::sycl::cl_float4)
53 SYCL_PLOG1P(cl::sycl::cl_double2)
54 #undef SYCL_PLOG1P
55 
56 #define SYCL_PLOG10(packet_type) \
57  template <> \
58  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type plog10<packet_type>(const packet_type& a) { \
59  return cl::sycl::log10(a); \
60  }
61 
62 SYCL_PLOG10(cl::sycl::cl_half8)
63 SYCL_PLOG10(cl::sycl::cl_float4)
64 SYCL_PLOG10(cl::sycl::cl_double2)
65 #undef SYCL_PLOG10
66 
67 #define SYCL_PEXP(packet_type) \
68  template <> \
69  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pexp<packet_type>(const packet_type& a) { \
70  return cl::sycl::exp(a); \
71  }
72 
73 SYCL_PEXP(cl::sycl::cl_half8)
74 SYCL_PEXP(cl::sycl::cl_half)
75 SYCL_PEXP(cl::sycl::cl_float4)
76 SYCL_PEXP(cl::sycl::cl_float)
77 SYCL_PEXP(cl::sycl::cl_double2)
78 #undef SYCL_PEXP
79 
80 #define SYCL_PEXPM1(packet_type) \
81  template <> \
82  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pexpm1<packet_type>(const packet_type& a) { \
83  return cl::sycl::expm1(a); \
84  }
85 
86 SYCL_PEXPM1(cl::sycl::cl_half8)
87 SYCL_PEXPM1(cl::sycl::cl_float4)
88 SYCL_PEXPM1(cl::sycl::cl_double2)
89 #undef SYCL_PEXPM1
90 
91 #define SYCL_PSQRT(packet_type) \
92  template <> \
93  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type psqrt<packet_type>(const packet_type& a) { \
94  return cl::sycl::sqrt(a); \
95  }
96 
97 SYCL_PSQRT(cl::sycl::cl_half8)
98 SYCL_PSQRT(cl::sycl::cl_float4)
99 SYCL_PSQRT(cl::sycl::cl_double2)
100 #undef SYCL_PSQRT
101 
102 #define SYCL_PRSQRT(packet_type) \
103  template <> \
104  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type prsqrt<packet_type>(const packet_type& a) { \
105  return cl::sycl::rsqrt(a); \
106  }
107 
108 SYCL_PRSQRT(cl::sycl::cl_half8)
109 SYCL_PRSQRT(cl::sycl::cl_float4)
110 SYCL_PRSQRT(cl::sycl::cl_double2)
111 #undef SYCL_PRSQRT
112 
114 #define SYCL_PSIN(packet_type) \
115  template <> \
116  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type psin<packet_type>(const packet_type& a) { \
117  return cl::sycl::sin(a); \
118  }
119 
120 SYCL_PSIN(cl::sycl::cl_half8)
121 SYCL_PSIN(cl::sycl::cl_float4)
122 SYCL_PSIN(cl::sycl::cl_double2)
123 #undef SYCL_PSIN
124 
126 #define SYCL_PCOS(packet_type) \
127  template <> \
128  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pcos<packet_type>(const packet_type& a) { \
129  return cl::sycl::cos(a); \
130  }
131 
132 SYCL_PCOS(cl::sycl::cl_half8)
133 SYCL_PCOS(cl::sycl::cl_float4)
134 SYCL_PCOS(cl::sycl::cl_double2)
135 #undef SYCL_PCOS
136 
138 #define SYCL_PTAN(packet_type) \
139  template <> \
140  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type ptan<packet_type>(const packet_type& a) { \
141  return cl::sycl::tan(a); \
142  }
143 
144 SYCL_PTAN(cl::sycl::cl_half8)
145 SYCL_PTAN(cl::sycl::cl_float4)
146 SYCL_PTAN(cl::sycl::cl_double2)
147 #undef SYCL_PTAN
148 
150 #define SYCL_PASIN(packet_type) \
151  template <> \
152  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pasin<packet_type>(const packet_type& a) { \
153  return cl::sycl::asin(a); \
154  }
155 
156 SYCL_PASIN(cl::sycl::cl_half8)
157 SYCL_PASIN(cl::sycl::cl_float4)
158 SYCL_PASIN(cl::sycl::cl_double2)
159 #undef SYCL_PASIN
160 
162 #define SYCL_PACOS(packet_type) \
163  template <> \
164  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pacos<packet_type>(const packet_type& a) { \
165  return cl::sycl::acos(a); \
166  }
167 
168 SYCL_PACOS(cl::sycl::cl_half8)
169 SYCL_PACOS(cl::sycl::cl_float4)
170 SYCL_PACOS(cl::sycl::cl_double2)
171 #undef SYCL_PACOS
172 
174 #define SYCL_PATAN(packet_type) \
175  template <> \
176  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type patan<packet_type>(const packet_type& a) { \
177  return cl::sycl::atan(a); \
178  }
179 
180 SYCL_PATAN(cl::sycl::cl_half8)
181 SYCL_PATAN(cl::sycl::cl_float4)
182 SYCL_PATAN(cl::sycl::cl_double2)
183 #undef SYCL_PATAN
184 
186 #define SYCL_PSINH(packet_type) \
187  template <> \
188  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type psinh<packet_type>(const packet_type& a) { \
189  return cl::sycl::sinh(a); \
190  }
191 
192 SYCL_PSINH(cl::sycl::cl_half8)
193 SYCL_PSINH(cl::sycl::cl_float4)
194 SYCL_PSINH(cl::sycl::cl_double2)
195 #undef SYCL_PSINH
196 
198 #define SYCL_PCOSH(packet_type) \
199  template <> \
200  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pcosh<packet_type>(const packet_type& a) { \
201  return cl::sycl::cosh(a); \
202  }
203 
204 SYCL_PCOSH(cl::sycl::cl_half8)
205 SYCL_PCOSH(cl::sycl::cl_float4)
206 SYCL_PCOSH(cl::sycl::cl_double2)
207 #undef SYCL_PCOSH
208 
210 #define SYCL_PTANH(packet_type) \
211  template <> \
212  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type ptanh<packet_type>(const packet_type& a) { \
213  return cl::sycl::tanh(a); \
214  }
215 
216 SYCL_PTANH(cl::sycl::cl_half8)
217 SYCL_PTANH(cl::sycl::cl_float4)
218 SYCL_PTANH(cl::sycl::cl_double2)
219 #undef SYCL_PTANH
220 
221 #define SYCL_PCEIL(packet_type) \
222  template <> \
223  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pceil<packet_type>(const packet_type& a) { \
224  return cl::sycl::ceil(a); \
225  }
226 
227 SYCL_PCEIL(cl::sycl::cl_half)
228 SYCL_PCEIL(cl::sycl::cl_float4)
229 SYCL_PCEIL(cl::sycl::cl_double2)
230 #undef SYCL_PCEIL
231 
232 #define SYCL_PROUND(packet_type) \
233  template <> \
234  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pround<packet_type>(const packet_type& a) { \
235  return cl::sycl::round(a); \
236  }
237 
238 SYCL_PROUND(cl::sycl::cl_half8)
239 SYCL_PROUND(cl::sycl::cl_float4)
240 SYCL_PROUND(cl::sycl::cl_double2)
241 #undef SYCL_PROUND
242 
243 #define SYCL_PRINT(packet_type) \
244  template <> \
245  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type print<packet_type>(const packet_type& a) { \
246  return cl::sycl::rint(a); \
247  }
248 
249 SYCL_PRINT(cl::sycl::cl_half8)
250 SYCL_PRINT(cl::sycl::cl_float4)
251 SYCL_PRINT(cl::sycl::cl_double2)
252 #undef SYCL_PRINT
253 
254 #define SYCL_FLOOR(packet_type) \
255  template <> \
256  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pfloor<packet_type>(const packet_type& a) { \
257  return cl::sycl::floor(a); \
258  }
259 
260 SYCL_FLOOR(cl::sycl::cl_half8)
261 SYCL_FLOOR(cl::sycl::cl_float4)
262 SYCL_FLOOR(cl::sycl::cl_double2)
263 #undef SYCL_FLOOR
264 
265 #define SYCL_PMIN(packet_type, expr) \
266  template <> \
267  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pmin<packet_type>(const packet_type& a, const packet_type& b) { \
268  return expr; \
269  }
270 
271 SYCL_PMIN(cl::sycl::cl_half8, cl::sycl::fmin(a, b))
272 SYCL_PMIN(cl::sycl::cl_float4, cl::sycl::fmin(a, b))
273 SYCL_PMIN(cl::sycl::cl_double2, cl::sycl::fmin(a, b))
274 #undef SYCL_PMIN
275 
276 #define SYCL_PMAX(packet_type, expr) \
277  template <> \
278  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pmax<packet_type>(const packet_type& a, const packet_type& b) { \
279  return expr; \
280  }
281 
282 SYCL_PMAX(cl::sycl::cl_half8, cl::sycl::fmax(a, b))
283 SYCL_PMAX(cl::sycl::cl_float4, cl::sycl::fmax(a, b))
284 SYCL_PMAX(cl::sycl::cl_double2, cl::sycl::fmax(a, b))
285 #undef SYCL_PMAX
286 
287 #define SYCL_PLDEXP(packet_type) \
288  template <> \
289  EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pldexp(const packet_type& a, const packet_type& exponent) { \
290  return cl::sycl::ldexp(a, exponent.template convert<cl::sycl::cl_int, cl::sycl::rounding_mode::automatic>()); \
291  }
292 
293 SYCL_PLDEXP(cl::sycl::cl_half8)
294 SYCL_PLDEXP(cl::sycl::cl_float4)
295 SYCL_PLDEXP(cl::sycl::cl_double2)
296 #undef SYCL_PLDEXP
297 
298 #endif
299 } // end namespace internal
300 
301 } // end namespace Eigen
302 
303 #endif // EIGEN_MATH_FUNCTIONS_SYCL_H
b
Scalar * b
Definition: benchVecAdd.cpp:17
a
const Scalar * a
Definition: level2_cplx_impl.h:32
Eigen::bfloat16_impl::fmin
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bfloat16 fmin(const bfloat16 &a, const bfloat16 &b)
Definition: BFloat16.h:664
Eigen::bfloat16_impl::fmax
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bfloat16 fmax(const bfloat16 &a, const bfloat16 &b)
Definition: BFloat16.h:670
Eigen
Namespace containing all symbols from the Eigen library.
Definition: bench_norm.cpp:70
internal
Definition: Eigen_Colamd.h:49