MathFunctions.h (12539B)
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 namespace Eigen { 24 25 namespace internal { 26 27 // Make sure this is only available when targeting a GPU: we don't want to 28 // introduce conflicts between these packet_traits definitions and the ones 29 // we'll use on the host side (SSE, AVX, ...) 30 #if defined(SYCL_DEVICE_ONLY) 31 #define SYCL_PLOG(packet_type) \ 32 template <> \ 33 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type plog<packet_type>( \ 34 const packet_type& a) { \ 35 return cl::sycl::log(a); \ 36 } 37 38 SYCL_PLOG(cl::sycl::cl_float4) 39 SYCL_PLOG(cl::sycl::cl_double2) 40 #undef SYCL_PLOG 41 42 #define SYCL_PLOG1P(packet_type) \ 43 template <> \ 44 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type plog1p<packet_type>( \ 45 const packet_type& a) { \ 46 return cl::sycl::log1p(a); \ 47 } 48 49 SYCL_PLOG1P(cl::sycl::cl_float4) 50 SYCL_PLOG1P(cl::sycl::cl_double2) 51 #undef SYCL_PLOG1P 52 53 #define SYCL_PLOG10(packet_type) \ 54 template <> \ 55 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type plog10<packet_type>( \ 56 const packet_type& a) { \ 57 return cl::sycl::log10(a); \ 58 } 59 60 SYCL_PLOG10(cl::sycl::cl_float4) 61 SYCL_PLOG10(cl::sycl::cl_double2) 62 #undef SYCL_PLOG10 63 64 #define SYCL_PEXP(packet_type) \ 65 template <> \ 66 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pexp<packet_type>( \ 67 const packet_type& a) { \ 68 return cl::sycl::exp(a); \ 69 } 70 71 SYCL_PEXP(cl::sycl::cl_float4) 72 SYCL_PEXP(cl::sycl::cl_float) 73 SYCL_PEXP(cl::sycl::cl_double2) 74 #undef SYCL_PEXP 75 76 #define SYCL_PEXPM1(packet_type) \ 77 template <> \ 78 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pexpm1<packet_type>( \ 79 const packet_type& a) { \ 80 return cl::sycl::expm1(a); \ 81 } 82 83 SYCL_PEXPM1(cl::sycl::cl_float4) 84 SYCL_PEXPM1(cl::sycl::cl_double2) 85 #undef SYCL_PEXPM1 86 87 #define SYCL_PSQRT(packet_type) \ 88 template <> \ 89 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type psqrt<packet_type>( \ 90 const packet_type& a) { \ 91 return cl::sycl::sqrt(a); \ 92 } 93 94 SYCL_PSQRT(cl::sycl::cl_float4) 95 SYCL_PSQRT(cl::sycl::cl_double2) 96 #undef SYCL_PSQRT 97 98 #define SYCL_PRSQRT(packet_type) \ 99 template <> \ 100 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type prsqrt<packet_type>( \ 101 const packet_type& a) { \ 102 return cl::sycl::rsqrt(a); \ 103 } 104 105 SYCL_PRSQRT(cl::sycl::cl_float4) 106 SYCL_PRSQRT(cl::sycl::cl_double2) 107 #undef SYCL_PRSQRT 108 109 /** \internal \returns the hyperbolic sine of \a a (coeff-wise) */ 110 #define SYCL_PSIN(packet_type) \ 111 template <> \ 112 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type psin<packet_type>( \ 113 const packet_type& a) { \ 114 return cl::sycl::sin(a); \ 115 } 116 117 SYCL_PSIN(cl::sycl::cl_float4) 118 SYCL_PSIN(cl::sycl::cl_double2) 119 #undef SYCL_PSIN 120 121 /** \internal \returns the hyperbolic cosine of \a a (coeff-wise) */ 122 #define SYCL_PCOS(packet_type) \ 123 template <> \ 124 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pcos<packet_type>( \ 125 const packet_type& a) { \ 126 return cl::sycl::cos(a); \ 127 } 128 129 SYCL_PCOS(cl::sycl::cl_float4) 130 SYCL_PCOS(cl::sycl::cl_double2) 131 #undef SYCL_PCOS 132 133 /** \internal \returns the hyperbolic tan of \a a (coeff-wise) */ 134 #define SYCL_PTAN(packet_type) \ 135 template <> \ 136 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type ptan<packet_type>( \ 137 const packet_type& a) { \ 138 return cl::sycl::tan(a); \ 139 } 140 141 SYCL_PTAN(cl::sycl::cl_float4) 142 SYCL_PTAN(cl::sycl::cl_double2) 143 #undef SYCL_PTAN 144 145 /** \internal \returns the hyperbolic sine of \a a (coeff-wise) */ 146 #define SYCL_PASIN(packet_type) \ 147 template <> \ 148 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pasin<packet_type>( \ 149 const packet_type& a) { \ 150 return cl::sycl::asin(a); \ 151 } 152 153 SYCL_PASIN(cl::sycl::cl_float4) 154 SYCL_PASIN(cl::sycl::cl_double2) 155 #undef SYCL_PASIN 156 157 /** \internal \returns the hyperbolic cosine of \a a (coeff-wise) */ 158 #define SYCL_PACOS(packet_type) \ 159 template <> \ 160 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pacos<packet_type>( \ 161 const packet_type& a) { \ 162 return cl::sycl::acos(a); \ 163 } 164 165 SYCL_PACOS(cl::sycl::cl_float4) 166 SYCL_PACOS(cl::sycl::cl_double2) 167 #undef SYCL_PACOS 168 169 /** \internal \returns the hyperbolic tan of \a a (coeff-wise) */ 170 #define SYCL_PATAN(packet_type) \ 171 template <> \ 172 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type patan<packet_type>( \ 173 const packet_type& a) { \ 174 return cl::sycl::atan(a); \ 175 } 176 177 SYCL_PATAN(cl::sycl::cl_float4) 178 SYCL_PATAN(cl::sycl::cl_double2) 179 #undef SYCL_PATAN 180 181 /** \internal \returns the hyperbolic sine of \a a (coeff-wise) */ 182 #define SYCL_PSINH(packet_type) \ 183 template <> \ 184 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type psinh<packet_type>( \ 185 const packet_type& a) { \ 186 return cl::sycl::sinh(a); \ 187 } 188 189 SYCL_PSINH(cl::sycl::cl_float4) 190 SYCL_PSINH(cl::sycl::cl_double2) 191 #undef SYCL_PSINH 192 193 /** \internal \returns the hyperbolic cosine of \a a (coeff-wise) */ 194 #define SYCL_PCOSH(packet_type) \ 195 template <> \ 196 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pcosh<packet_type>( \ 197 const packet_type& a) { \ 198 return cl::sycl::cosh(a); \ 199 } 200 201 SYCL_PCOSH(cl::sycl::cl_float4) 202 SYCL_PCOSH(cl::sycl::cl_double2) 203 #undef SYCL_PCOSH 204 205 /** \internal \returns the hyperbolic tan of \a a (coeff-wise) */ 206 #define SYCL_PTANH(packet_type) \ 207 template <> \ 208 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type ptanh<packet_type>( \ 209 const packet_type& a) { \ 210 return cl::sycl::tanh(a); \ 211 } 212 213 SYCL_PTANH(cl::sycl::cl_float4) 214 SYCL_PTANH(cl::sycl::cl_double2) 215 #undef SYCL_PTANH 216 217 #define SYCL_PCEIL(packet_type) \ 218 template <> \ 219 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pceil<packet_type>( \ 220 const packet_type& a) { \ 221 return cl::sycl::ceil(a); \ 222 } 223 224 SYCL_PCEIL(cl::sycl::cl_float4) 225 SYCL_PCEIL(cl::sycl::cl_double2) 226 #undef SYCL_PCEIL 227 228 #define SYCL_PROUND(packet_type) \ 229 template <> \ 230 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pround<packet_type>( \ 231 const packet_type& a) { \ 232 return cl::sycl::round(a); \ 233 } 234 235 SYCL_PROUND(cl::sycl::cl_float4) 236 SYCL_PROUND(cl::sycl::cl_double2) 237 #undef SYCL_PROUND 238 239 #define SYCL_PRINT(packet_type) \ 240 template <> \ 241 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type print<packet_type>( \ 242 const packet_type& a) { \ 243 return cl::sycl::rint(a); \ 244 } 245 246 SYCL_PRINT(cl::sycl::cl_float4) 247 SYCL_PRINT(cl::sycl::cl_double2) 248 #undef SYCL_PRINT 249 250 #define SYCL_FLOOR(packet_type) \ 251 template <> \ 252 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pfloor<packet_type>( \ 253 const packet_type& a) { \ 254 return cl::sycl::floor(a); \ 255 } 256 257 SYCL_FLOOR(cl::sycl::cl_float4) 258 SYCL_FLOOR(cl::sycl::cl_double2) 259 #undef SYCL_FLOOR 260 261 #define SYCL_PMIN(packet_type, expr) \ 262 template <> \ 263 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pmin<packet_type>( \ 264 const packet_type& a, const packet_type& b) { \ 265 return expr; \ 266 } 267 268 SYCL_PMIN(cl::sycl::cl_float4, cl::sycl::fmin(a, b)) 269 SYCL_PMIN(cl::sycl::cl_double2, cl::sycl::fmin(a, b)) 270 #undef SYCL_PMIN 271 272 #define SYCL_PMAX(packet_type, expr) \ 273 template <> \ 274 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pmax<packet_type>( \ 275 const packet_type& a, const packet_type& b) { \ 276 return expr; \ 277 } 278 279 SYCL_PMAX(cl::sycl::cl_float4, cl::sycl::fmax(a, b)) 280 SYCL_PMAX(cl::sycl::cl_double2, cl::sycl::fmax(a, b)) 281 #undef SYCL_PMAX 282 283 #define SYCL_PLDEXP(packet_type) \ 284 template <> \ 285 EIGEN_DEVICE_FUNC EIGEN_STRONG_INLINE packet_type pldexp( \ 286 const packet_type& a, const packet_type& exponent) { \ 287 return cl::sycl::ldexp( \ 288 a, exponent.template convert<cl::sycl::cl_int, \ 289 cl::sycl::rounding_mode::automatic>()); \ 290 } 291 292 SYCL_PLDEXP(cl::sycl::cl_float4) 293 SYCL_PLDEXP(cl::sycl::cl_double2) 294 #undef SYCL_PLDEXP 295 296 #endif 297 } // end namespace internal 298 299 } // end namespace Eigen 300 301 #endif // EIGEN_MATH_FUNCTIONS_SYCL_H