Move all logic with cyl_bessel_i0 support in i0.hpp

Author: vlad-perevezentsev

CHANGELOG.md

@@ -34,6 +34,7 @@ This release achieves 100% compliance with Python Array API specification (revis
 * Resolved an issue with an incorrect result returned due to missing dependency from the strided kernel on a copy event in `dpnp.erf` [#2378](https://github.com/IntelPython/dpnp/pull/2378)
 * Updated `conda create` commands build and install instructions of `Quick start guide` to avoid a compilation error [#2395](https://github.com/IntelPython/dpnp/pull/2395)
 * Added handling of empty string passed to a test env variable defining data type scope as a `False` value [#2415](https://github.com/IntelPython/dpnp/pull/2415)
+* Resolved build issues on non-Intel targets in `dpnp.i0` and `dpnp.kaiser` [#2439](https://github.com/IntelPython/dpnp/pull/2439)
 
 
 ## [0.17.0] - 02/26/2025

dpnp/backend/extensions/window/kaiser.cpp

@@ -32,18 +32,6 @@
 
 #include <sycl/sycl.hpp>
 
-/**
- * Version of SYCL DPC++ 2025.1 compiler where an issue with
- * sycl::ext::intel::math::cyl_bessel_i0(x) is fully resolved.
- */
-#ifndef __SYCL_COMPILER_BESSEL_I0_SUPPORT
-#define __SYCL_COMPILER_BESSEL_I0_SUPPORT 20241208L
-#endif
-
-#if __SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT
-#include <sycl/ext/intel/math.hpp>
-#endif
-
 #include "../kernels/elementwise_functions/i0.hpp"
 
 namespace dpnp::extensions::window
@@ -74,11 +62,7 @@ class KaiserFunctor
 
     void operator()(sycl::id<1> id) const
     {
-#if __SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT
-        using sycl::ext::intel::math::cyl_bessel_i0;
-#else
-        using dpnp::kernels::i0::impl::cyl_bessel_i0;
-#endif
+        using dpnp::kernels::i0::cyl_bessel_i0;
 
         const auto i = id.get(0);
         const T alpha = (N - 1) / T(2);

dpnp/backend/kernels/elementwise_functions/i0.hpp

@@ -35,12 +35,28 @@
 #define __SYCL_COMPILER_BESSEL_I0_SUPPORT 20241208L
 #endif
 
-#if __SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT
+/**
+ * Include <sycl/ext/intel/math.hpp> only when targeting to Intel devices.
+ * This header relies on intel-specific types like _iml_half_internal,
+ * which are not suppose to work with other targets (e.g., CUDA, AMD).
+ */
+#if defined(__SPIR__) && defined(__INTEL_LLVM_COMPILER) 
+#define __SYCL_EXT_INTEL_MATH_SUPPORT
+#endif
+
+#if defined(__SYCL_EXT_INTEL_MATH_SUPPORT) &&                                  \
+    (__SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT)
 #include <sycl/ext/intel/math.hpp>
 #endif
 
 namespace dpnp::kernels::i0
 {
+#if defined(__SYCL_EXT_INTEL_MATH_SUPPORT) &&                                  \
+    (__SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT)
+using sycl::ext::intel::math::cyl_bessel_i0;
+
+#else
+
 /**
  * The below implementation of Bessel function of order 0
  * is based on the source code from https://github.com/gcc-mirror/gcc
@@ -239,6 +255,10 @@ inline Tp cyl_bessel_i0(Tp x)
 }
 } // namespace impl
 
+using impl::cyl_bessel_i0;
+
+#endif
+
 template <typename argT, typename resT>
 struct I0Functor
 {
@@ -253,12 +273,6 @@ struct I0Functor
 
     resT operator()(const argT &x) const
     {
-#if __SYCL_COMPILER_VERSION >= __SYCL_COMPILER_BESSEL_I0_SUPPORT
-        using sycl::ext::intel::math::cyl_bessel_i0;
-#else
-        using impl::cyl_bessel_i0;
-#endif
-
         if constexpr (std::is_same_v<resT, sycl::half>) {
             return static_cast<resT>(cyl_bessel_i0<float>(float(x)));
         }