SYCL: Use in-order queues in InterOp tests (kokkos#6246)

* SYCL: Use in-order queues in InterOp tests * Check that user-provided sycl::queues are in-order * Add comment * Improve comment. Co-authored-by: Damien L-G <dalg24+github@gmail.com> * Fix partition_by_args --------- Co-authored-by: Damien L-G <dalg24+github@gmail.com>
nliber · Jul 5, 2023 · 6a8488d · 6a8488d
1 parent 13948ec
commit 6a8488d
Show file tree

Hide file tree

Showing 6 changed files with 23 additions and 8 deletions.
diff --git a/core/src/SYCL/Kokkos_SYCL.cpp b/core/src/SYCL/Kokkos_SYCL.cpp
@@ -59,6 +59,13 @@ SYCL::SYCL(const sycl::queue& stream)
         ptr->finalize();
         delete ptr;
       }) {
+  // In principle could be guarded with
+  // #ifdef KOKKOS_IMPL_SYCL_USE_IN_ORDER_QUEUES
+  // but we chose to require user-provided queues to be in-order
+  // unconditionally so that code downstream does not break
+  // when the backend setting changes.
+  if (!stream.is_in_order())
+    Kokkos::abort("User provided sycl::queues must be in-order!");
   Impl::SYCLInternal::singleton().verify_is_initialized(
       "SYCL instance constructor");
   m_space_instance->initialize(stream);

diff --git a/core/src/SYCL/Kokkos_SYCL.hpp b/core/src/SYCL/Kokkos_SYCL.hpp
@@ -163,7 +163,8 @@ std::vector<SYCL> partition_space(const SYCL& sycl_space, Args...) {
   std::vector<SYCL> instances;
   instances.reserve(sizeof...(Args));
   for (unsigned int i = 0; i < sizeof...(Args); ++i)
-    instances.emplace_back(sycl::queue(context, device));
+    instances.emplace_back(
+        sycl::queue(context, device, sycl::property::queue::in_order()));
   return instances;
 }
 
@@ -183,7 +184,8 @@ std::vector<SYCL> partition_space(const SYCL& sycl_space,
   // otherwise.
   instances.reserve(weights.size());
   for (unsigned int i = 0; i < weights.size(); ++i)
-    instances.emplace_back(sycl::queue(context, device));
+    instances.emplace_back(
+        sycl::queue(context, device, sycl::property::queue::in_order()));
   return instances;
 }
 }  // namespace Experimental

diff --git a/core/unit_test/sycl/TestSYCL_InterOp_Init.cpp b/core/unit_test/sycl/TestSYCL_InterOp_Init.cpp
@@ -29,7 +29,8 @@ TEST(sycl, raw_sycl_interop) {
   Kokkos::Experimental::SYCL default_space;
   sycl::context default_context = default_space.sycl_queue().get_context();
 
-  sycl::queue queue(default_context, sycl::default_selector_v);
+  sycl::queue queue(default_context, sycl::default_selector_v,
+                    sycl::property::queue::in_order());
   constexpr int n = 100;
   int* p          = sycl::malloc_device<int>(n, queue);
   {

diff --git a/core/unit_test/sycl/TestSYCL_InterOp_Init_Context.cpp b/core/unit_test/sycl/TestSYCL_InterOp_Init_Context.cpp
@@ -27,7 +27,8 @@ TEST(sycl, raw_sycl_interop_context_1) {
   Kokkos::Experimental::SYCL default_space;
   sycl::context default_context = default_space.sycl_queue().get_context();
 
-  sycl::queue queue(default_context, sycl::default_selector_v);
+  sycl::queue queue(default_context, sycl::default_selector_v,
+                    sycl::property::queue::in_order());
   constexpr int n = 100;
   int* p          = sycl::malloc_device<int>(n, queue);
 
@@ -60,7 +61,8 @@ TEST(sycl, raw_sycl_interop_context_2) {
   Kokkos::Experimental::SYCL default_space;
   sycl::context default_context = default_space.sycl_queue().get_context();
 
-  sycl::queue queue(default_context, sycl::default_selector_v);
+  sycl::queue queue(default_context, sycl::default_selector_v,
+                    sycl::property::queue::in_order());
   constexpr int n = 100;
 
   Kokkos::Experimental::SYCL space(queue);

diff --git a/core/unit_test/sycl/TestSYCL_InterOp_Streams.cpp b/core/unit_test/sycl/TestSYCL_InterOp_Streams.cpp
@@ -25,7 +25,8 @@ TEST(sycl, raw_sycl_queues) {
   Kokkos::Experimental::SYCL default_space;
   sycl::context default_context = default_space.sycl_queue().get_context();
 
-  sycl::queue queue(default_context, sycl::default_selector_v);
+  sycl::queue queue(default_context, sycl::default_selector_v,
+                    sycl::property::queue::in_order());
   int* p            = sycl::malloc_device<int>(100, queue);
   using MemorySpace = typename TEST_EXECSPACE::memory_space;
 

diff --git a/core/unit_test/sycl/TestSYCL_TeamScratchStreams.cpp b/core/unit_test/sycl/TestSYCL_TeamScratchStreams.cpp
@@ -76,12 +76,14 @@ void sycl_queue_scratch_test(
   Kokkos::Experimental::SYCL default_space;
   sycl::context default_context = default_space.sycl_queue().get_context();
 
-  sycl::queue queue(default_context, sycl::default_selector_v);
+  sycl::queue queue(default_context, sycl::default_selector_v,
+                    sycl::property::queue::in_order());
 
   std::array<Kokkos::Experimental::SYCL, K> sycl;
   for (int i = 0; i < K; i++) {
     sycl[i] = Kokkos::Experimental::SYCL(
-        sycl::queue(default_context, sycl::default_selector_v));
+        sycl::queue(default_context, sycl::default_selector_v,
+                    sycl::property::queue::in_order()));
   }
 
   // Test that growing scratch size in subsequent calls doesn't crash things