Use Kokkos::printf for Kokkos >= 4.1.99

arborx · Oct 12, 2023 · 3e3655f · 3e3655f
1 parent 6a3a418
commit 3e3655f
Show file tree

Hide file tree

Showing 4 changed files with 34 additions and 20 deletions.
diff --git a/benchmarks/dbscan/ArborX_DBSCANVerification.hpp b/benchmarks/dbscan/ArborX_DBSCANVerification.hpp
@@ -46,7 +46,9 @@ bool verifyCorePointsNonnegativeIndex(ExecutionSpace const &exec_space,
         bool self_is_core_point = (offset(i + 1) - offset(i) >= core_min_size);
         if (self_is_core_point && labels(i) < 0)
         {
-#ifdef __SYCL_DEVICE_ONLY__
+#if KOKKOS_VERSION >= 40199
+          using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
           using sycl::ext::oneapi::experimental::printf;
 #endif
           printf("Core point is marked as noise: %d [%d]\n", i, labels(i));
@@ -82,7 +84,9 @@ bool verifyConnectedCorePointsShareIndex(ExecutionSpace const &exec_space,
 
             if (neigh_is_core_point && labels(i) != labels(j))
             {
-#ifdef __SYCL_DEVICE_ONLY__
+#if KOKKOS_VERSION >= 40199
+              using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
               using sycl::ext::oneapi::experimental::printf;
 #endif
               printf("Connected cores do not belong to the same cluster: "
@@ -134,22 +138,22 @@ bool verifyBorderAndNoisePoints(ExecutionSpace const &exec_space,
             }
           }
 
+#if KOKKOS_VERSION >= 40199
+          using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
+          using sycl::ext::oneapi::experimental::printf;
+#endif
+
           // Border point must be connected to a core point
           if (is_border && !have_shared_core)
           {
-#ifdef __SYCL_DEVICE_ONLY__
-            using sycl::ext::oneapi::experimental::printf;
-#endif
             printf("Border point does not belong to a cluster: %d [%d]\n", i,
                    labels(i));
             update++;
           }
           // Noise points must have index -1
           if (!is_border && labels(i) != -1)
           {
-#ifdef __SYCL_DEVICE_ONLY__
-            using sycl::ext::oneapi::experimental::printf;
-#endif
             printf("Noise point does not have index -1: %d [%d]\n", i,
                    labels(i));
             update++;

diff --git a/examples/callback/example_callback.cpp b/examples/callback/example_callback.cpp
@@ -56,7 +56,9 @@ struct PrintfCallback
   KOKKOS_FUNCTION void operator()(Predicate, int primitive,
                                   OutputFunctor const &out) const
   {
-#ifdef __SYCL_DEVICE_ONLY__
+#if KOKKOS_VERSION >= 40199
+    using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
     using sycl::ext::oneapi::experimental::printf;
 #endif
     printf("Found %d from functor\n", primitive);
@@ -95,7 +97,9 @@ int main(int argc, char *argv[])
         bvh, ExecutionSpace{}, FirstOctant{},
         KOKKOS_LAMBDA(auto /*predicate*/, int primitive,
                       auto /*output_functor*/) {
-#ifdef __SYCL_DEVICE_ONLY__
+#if KOKKOS_VERSION >= 40199
+          using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
           using sycl::ext::oneapi::experimental::printf;
 #endif
           printf("Found %d from generic lambda\n", primitive);
@@ -115,7 +119,9 @@ int main(int argc, char *argv[])
         bvh, ExecutionSpace{}, NearestToOrigin{k},
         KOKKOS_LAMBDA(auto /*predicate*/, int primitive,
                       auto /*output_functor*/) {
-#ifdef __SYCL_DEVICE_ONLY__
+#if KOKKOS_VERSION >= 40199
+          using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
           using sycl::ext::oneapi::experimental::printf;
 #endif
           printf("Found %d from generic lambda\n", primitive);
@@ -133,7 +139,9 @@ int main(int argc, char *argv[])
     bvh.query(
         ExecutionSpace{}, FirstOctant{},
         KOKKOS_LAMBDA(auto /*predicate*/, int j) {
-#ifdef __SYCL_DEVICE_ONLY__
+#if KOKKOS_VERSION >= 40199
+          using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
           using sycl::ext::oneapi::experimental::printf;
 #endif
           printf("%d %d %d\n", ++c(), -1, j);

diff --git a/examples/raytracing/example_raytracing.cpp b/examples/raytracing/example_raytracing.cpp
@@ -406,11 +406,14 @@ int main(int argc, char *argv[])
             fabs(energy_intersects(i));
         if (abs_error > rel_tol * fabs(energy_intersects(i)))
         {
-#ifndef KOKKOS_ENABLE_SYCL
+#if KOKKOS_VERSION >= 40199
+          using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
+          using sycl::ext::oneapi::experimental::printf;
+#endif
           printf("%d: %f != %f, relative error: %f\n", i,
                  energy_ordered_intersects(i), energy_intersects(i),
                  abs_error / fabs(energy_intersects(i)));
-#endif
           ++error;
         }
       },

diff --git a/examples/triangle_intersection/triangle_intersection.cpp b/examples/triangle_intersection/triangle_intersection.cpp
@@ -345,12 +345,14 @@ int main()
       KOKKOS_LAMBDA(int i, bool &update) {
         constexpr float eps = 1.e-3;
 
+#if KOKKOS_VERSION >= 40199
+        using Kokkos::printf;
+#elif defined(__SYCL_DEVICE_ONLY__)
+        using sycl::ext::oneapi::experimental::printf;
+#endif
         if (offsets(i) != i)
         {
-// FIXME_SYCL doesn't support printf
-#ifndef KOKKOS_ENABLE_SYCL
           printf("Offsets are wrong for query %d.\n", i);
-#endif
           update = false;
         }
         auto const &c = coefficients(i);
@@ -361,10 +363,7 @@ int main()
         if ((Kokkos::abs(p[0] - p_ref[0]) > eps) ||
             Kokkos::abs(p[1] - p_ref[1]) > eps)
         {
-// FIXME_SYCL doesn't support printf
-#ifndef KOKKOS_ENABLE_SYCL
           printf("Coefficients are wrong for query %d.\n", i);
-#endif
           update = false;
         }
       },