35
35
#define __SYCL_COMPILER_BESSEL_I0_SUPPORT 20241208L
36
36
#endif
37
37
38
- // The <sycl/ext/intel/math.hpp> header and related types like
39
- // _iml_half_internal are not compatible with NVIDIA GPUs and cause compilation
40
- // errors when building with -fsycl-targets=nvptx64-nvidia-cuda.
41
- #if !defined(__NVPTX__) && \
38
+ /* *
39
+ * The <sycl/ext/intel/math.hpp> header and related types like
40
+ * _iml_half_internal are not compatible with NVIDIA GPUs and cause compilation
41
+ * errors when building with -fsycl-targets=nvptx64-nvidia-cuda.
42
+ */
43
+ #if !defined(__NVPTX__)
44
+ #define __SYCL_EXT_INTEL_MATH_SUPPORT
45
+ #endif
46
+
47
+ #if defined(__SYCL_EXT_INTEL_MATH_SUPPORT) && \
42
48
(__SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT)
43
49
#include < sycl/ext/intel/math.hpp>
44
50
#endif
45
51
46
52
namespace dpnp ::kernels::i0
47
53
{
54
+ #if defined(__SYCL_EXT_INTEL_MATH_SUPPORT) && \
55
+ (__SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT)
56
+ using sycl::ext::intel::math::cyl_bessel_i0;
57
+
58
+ #else
59
+
48
60
/* *
49
61
* The below implementation of Bessel function of order 0
50
62
* is based on the source code from https://github.com/gcc-mirror/gcc
@@ -243,6 +255,10 @@ inline Tp cyl_bessel_i0(Tp x)
243
255
}
244
256
} // namespace impl
245
257
258
+ using impl::cyl_bessel_i0;
259
+
260
+ #endif
261
+
246
262
template <typename argT, typename resT>
247
263
struct I0Functor
248
264
{
@@ -257,13 +273,6 @@ struct I0Functor
257
273
258
274
resT operator ()(const argT &x) const
259
275
{
260
- #if !defined(__NVPTX__) && \
261
- (__SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT)
262
- using sycl::ext::intel::math::cyl_bessel_i0;
263
- #else
264
- using impl::cyl_bessel_i0;
265
- #endif
266
-
267
276
if constexpr (std::is_same_v<resT, sycl::half>) {
268
277
return static_cast <resT>(cyl_bessel_i0<float >(float (x)));
269
278
}
0 commit comments