Merge pull request #869 from bader/nd-item-equality

Update nd_item constructors and equality tests.
KhronosGroup · Feb 22, 2024 · 864b0fa · 864b0fa
2 parents 6114e29 + dff0fb6
commit 864b0fa
Show file tree

Hide file tree

Showing 2 changed files with 31 additions and 60 deletions.
diff --git a/tests/nd_item/nd_item_constructors.cpp b/tests/nd_item/nd_item_constructors.cpp
@@ -203,50 +203,6 @@ void test_constructors(util::logger& log) {
   CHECK(success_deprecated[to_integral(current_check::move_constructor)]);
   CHECK(success_deprecated[to_integral(current_check::move_assignment)]);
 #endif
-
-  // nd_item is not default constructible, store two objects into the array
-  static constexpr size_t numItems = 2;
-  using setup_kernel_t = nd_item_setup_kernel<numDims>;
-  auto items =
-      store_instances<numItems, invoke_nd_item<numDims, setup_kernel_t>>();
-  {
-    const auto& item = items[0];
-    const auto& itemReadOnly = item;
-    state_storage<numDims> expected(itemReadOnly);
-
-    {  // Check copy constructor
-      sycl::nd_item<numDims> copied(itemReadOnly);
-      CHECK(expected.check_equality(copied));
-#if SYCL_CTS_ENABLE_DEPRECATED_FEATURES_TESTS
-      CHECK(expected.check_equality_deprecated(copied));
-#endif
-    }
-    {  // Check copy assignment
-      auto copied = itemReadOnly;
-      CHECK(expected.check_equality(copied));
-#if SYCL_CTS_ENABLE_DEPRECATED_FEATURES_TESTS
-      CHECK(expected.check_equality_deprecated(copied));
-#endif
-    }
-    {  // Check move constructor; invalidates item
-      sycl::nd_item<numDims> moved(item);
-      CHECK(expected.check_equality(moved));
-#if SYCL_CTS_ENABLE_DEPRECATED_FEATURES_TESTS
-      CHECK(expected.check_equality_deprecated(moved));
-#endif
-    }
-  }
-  {
-    const auto& item = items[1];
-    state_storage<numDims> expected(item);
-
-    // Check move assignment; invalidates item
-    auto moved = std::move(item);
-    CHECK(expected.check_equality(moved));
-#if SYCL_CTS_ENABLE_DEPRECATED_FEATURES_TESTS
-    CHECK(expected.check_equality_deprecated(moved));
-#endif
-  }
 }
 
 class TEST_NAME : public util::test_base {

diff --git a/tests/nd_item/nd_item_equality.cpp b/tests/nd_item/nd_item_equality.cpp
@@ -28,9 +28,6 @@
 namespace TEST_NAMESPACE {
 using namespace sycl_cts;
 
-template <int numDims>
-struct nd_item_setup_kernel;
-
 template <int numDims>
 struct nd_item_equality_kernel;
 
@@ -46,23 +43,41 @@ class TEST_NAME : public util::test_base {
 
   template <int numDims>
   void test_equality(util::logger& log) {
+    using item_t = sycl::nd_item<numDims>;
+    using kernel_t = nd_item_equality_kernel<numDims>;
+
+    // Store comparison results from kernel into a success array
+    std::array<bool,
+               to_integral(common_by_value_semantics::current_check::size)>
+        success;
+    std::fill(std::begin(success), std::end(success), true);
+
     {
-      using item_t = sycl::nd_item<numDims>;
+      sycl::buffer<bool> successBuf(success.data(),
+                                    sycl::range<1>(success.size()));
+
+      const auto oneElemRange =
+          util::get_cts_object::range<numDims>::get(1, 1, 1);
 
-      // nd_item is not default constructible, store two objects into the array
-      static constexpr size_t numItems = 2;
-      using setup_kernel_t = nd_item_setup_kernel<numDims>;
-      auto items =
-          store_instances<numItems, invoke_nd_item<numDims, setup_kernel_t>>();
+      auto queue = util::get_cts_object::queue();
+      queue
+          .submit([&](sycl::handler& cgh) {
+            auto successAcc =
+                successBuf.get_access<sycl::access_mode::write>(cgh);
 
-      // Check nd_item equality operator on the device side
-      common_by_value_semantics::on_device_checker<item_t>::template run<
-          nd_item_equality_kernel<numDims>>(
-          log, items, "nd_item " + std::to_string(numDims) + " (device)");
+            cgh.parallel_for<kernel_t>(
+                sycl::nd_range<numDims>(oneElemRange, oneElemRange),
+                [=](item_t item) {
+                  common_by_value_semantics::check_equality(item, successAcc);
+                });
+          })
+          .wait_and_throw();
+    }
 
-      // Check nd_item equality operator on the host side
-      common_by_value_semantics::check_on_host(
-          log, items[0], "nd_item " + std::to_string(numDims) + " (host)");
+    for (int i = 0; i < success.size(); ++i) {
+      INFO(std::string(TOSTRING(TEST_NAME)) + " is " +
+           common_by_value_semantics::get_error_string(i));
+      CHECK(success[i]);
     }
   }