3535#define __SYCL_COMPILER_BESSEL_I0_SUPPORT 20241208L
3636#endif
3737
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) && \
4248 (__SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT)
4349#include < sycl/ext/intel/math.hpp>
4450#endif
4551
4652namespace dpnp ::kernels::i0
4753{
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+
4860/* *
4961 * The below implementation of Bessel function of order 0
5062 * is based on the source code from https://github.com/gcc-mirror/gcc
@@ -243,6 +255,10 @@ inline Tp cyl_bessel_i0(Tp x)
243255}
244256} // namespace impl
245257
258+ using impl::cyl_bessel_i0;
259+
260+ #endif
261+
246262template <typename argT, typename resT>
247263struct I0Functor
248264{
@@ -257,13 +273,6 @@ struct I0Functor
257273
258274 resT operator ()(const argT &x) const
259275 {
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-
267276 if constexpr (std::is_same_v<resT, sycl::half>) {
268277 return static_cast <resT>(cyl_bessel_i0<float >(float (x)));
269278 }
0 commit comments