Tests for sycl::get_kernel_id and sycl::get_kernel_ids

ci/computecpp.filter

@@ -12,6 +12,8 @@ hierarchical
 host_task
 image
 kernel
+kernel_args
+kernel_bundle
 math_builtin_api
 multi_ptr
 nd_item

ci/hipsycl.filter

@@ -17,6 +17,7 @@ image
 item
 kernel
 kernel_args
+kernel_bundle
 math_builtin_api
 multi_ptr
 nd_item

tests/CMakeLists.txt

@@ -141,10 +141,10 @@ endfunction()
 add_executable(test_all)
 
 # create test executable targets for each test project using the build_sycl function
-function(add_cts_test)
+function(add_cts_test_helper)
   get_filename_component(test_dir ${CMAKE_CURRENT_SOURCE_DIR} NAME)
-  set(test_exe_name test_${test_dir})
-  set(test_cases_list ${ARGN})
+  set(test_exe_name test_${ARGV0})
+  set(test_cases_list ${ARGV1})
 
   if(NOT ${test_dir} IN_LIST exclude_categories)
     message(STATUS "Adding test: " ${test_exe_name})
@@ -182,6 +182,36 @@ function(add_cts_test)
   target_sources(test_all PRIVATE $<TARGET_OBJECTS:${test_exe_name}_objects>)
 endfunction()
 
+# Create one *.exe-file from all of the provided *.cpp-files
+function(add_cts_test)
+  # To make check that any .cpp files are passed
+  # List created because, direct check on "${ARGN}" gives false-positive result
+  set(tests_list "${ARGN}")
+  if (tests_list)
+    get_filename_component(test_dir ${CMAKE_CURRENT_SOURCE_DIR} NAME)
+    set(test_exe_name ${test_dir})
+    set(test_cases_list "${ARGN}")
+
+    add_cts_test_helper(${test_exe_name} "${test_cases_list}")
+  endif()
+endfunction()
+
+# Create a separate *.exe-file from each of the provided *.cpp-files
+function(add_independent_cts_tests)
+  set(tests_list "${ARGN}")
+  foreach(ind_test IN LISTS tests_list)
+    if(EXISTS "${ind_test}")
+      get_filename_component(cpp_name "${ind_test}" NAME_WE)
+      set(test_exe_name "${cpp_name}")
+      set(test_cases_list "${ind_test}")
+
+      add_cts_test_helper(${test_exe_name} "${test_cases_list}")
+    else()
+      message(FATAL_ERROR "No file named ${ind_test}")
+    endif()
+  endforeach()
+endfunction()
+
 file(GLOB test_category_dirs RELATIVE ${CMAKE_CURRENT_SOURCE_DIR} *)
 list(REMOVE_ITEM test_category_dirs "common")
 foreach(dir ${test_category_dirs})

tests/common/common_by_reference.h

@@ -0,0 +1,135 @@
+/*******************************************************************************
+//
+//  SYCL 2020 Conformance Test Suite
+//
+// Provides verification for common by-reference semantics
+//
+*******************************************************************************/
+
+#ifndef __SYCLCTS_TESTS_COMMON_COMMON_BY_REFERENCE_H
+#define __SYCLCTS_TESTS_COMMON_COMMON_BY_REFERENCE_H
+
+#include "common.h"
+#include "common_semantics.h"
+#include "../../util/exceptions.h"
+
+#include <string>
+#include <utility>
+
+namespace common_by_reference {
+using namespace sycl_cts;
+
+/** @brief Can be used as condition in custom mutator
+ */
+enum class mutation : int {
+  const_correctness,
+  mutate_b,
+  mutate_a,
+  SIZE  // This should be last
+};
+
+/** @brief Predefined functor without mutation of passed object
+ */
+struct no_mutation {
+  template <typename T>
+  void operator()(T&, mutation) const {} // No mutation expected
+};
+
+/** @brief Check that object follows common reference semantics.
+ *  @attention Test for this check must be independent or last because passed
+ *             object will be moved during this check. So we can't use object
+ *             anymore.
+ */
+template <typename T, typename StorageT, typename MutatorT>
+void check_on_host(sycl_cts::util::logger &log, T& a, T& other,
+                   const std::string &testName, MutatorT mutator) {
+  common_semantics::check_on_host(log, a, other, testName);
+  // Create storage for following checks
+  StorageT storage;
+
+  auto mkInfo = [&](const std::string &check_name){
+    return testName + " Common reference semantics. " + check_name;
+  };
+
+  // Check that hashes are equal for a and copy, but different for a and other
+  std::hash<T> hasher;
+  T copy_to_hash = a;
+  if (hasher(a) != hasher(copy_to_hash)) {
+    FAIL(log, testName + "(Hashes of 'a' and 'b' are not equal)");
+  }
+  if (hasher(a) == hasher(other)) {
+    FAIL(log, "(Hashes should not be equal for 'a' and 'other')");
+  }
+
+  // Check that object is copy constructable
+  storage.store(a);
+  T copy_constructed_obj(a);
+  if (a != copy_constructed_obj) {
+    FAIL(log, testName + " is not copy constructible");
+  }
+  util::run_check(log, mkInfo("Copy construction."), [&]{
+    if (!storage.check(copy_constructed_obj)) {
+      FAIL(log, testName + " is not copy constructible (storage::check())");
+    }
+  });
+
+  // Check that obect is copy assignable
+  storage.store(a);
+  T copy_assigned_obj = a;
+  if (a != copy_assigned_obj) {
+    FAIL(log, testName + " is not copy assignable");
+  }
+  util::run_check(log, mkInfo("Copy assignment."), [&]{
+    if (!storage.check(copy_assigned_obj)) {
+      FAIL(log, testName + " is not copy assignable (storage::check())");
+    }
+  });
+
+  // Check const-correctness
+  const auto &const_obj = a;
+  storage.store(const_obj);
+  mutator(a, mutation::const_correctness);
+  util::run_check(log, mkInfo("Const-correcntess."), [&]{
+    if (!storage.check(const_obj)) {
+      FAIL(log, testName + " ignores const correctness (storage::check())");
+    }
+  });
+
+  // Check that mutation applied to both object and copy of object
+  T mutable_obj = a;
+  mutator(mutable_obj, mutation::mutate_b);
+  if (a != mutable_obj) {
+    FAIL(log, testName + " does not follow common reference semantics. (a != b"
+              "after mutation of 'b')");
+  }
+
+  mutator(a, mutation::mutate_a);
+  if (a != mutable_obj) {
+    FAIL(log, testName + " does not follow common reference semantics. (a != b"
+              "after mutation of 'a')");
+  }
+
+  // Move construction check
+  // Store object data before it is moved
+  storage.store(a);
+  T move_constructed_obj(std::move(a));
+  util::run_check(log, mkInfo("Move construction"), [&]{
+    if (!storage.check(move_constructed_obj)) {
+      FAIL(log, testName + " is not move constructible (storage::check())");
+    }
+  });
+
+  // Move assignment check
+  // Store object data before it is moved
+  storage.store(move_constructed_obj);
+  T move_assigned_obj = std::move(move_constructed_obj);
+  util::run_check(log, mkInfo("Move assignment."), [&]{
+    if (!storage.check(move_assigned_obj)) {
+      FAIL(log, testName + " is not move assignable (storage::check())");
+    }
+  });
+}
+
+}  // common_by_reference
+
+#endif  // __SYCLCTS_TESTS_COMMON_COMMON_BY_REFERENCE_H

tests/common/common_by_value.h → tests/common/common_semantics.h

@@ -2,26 +2,26 @@
 //
 //  SYCL 2020 Conformance Test Suite
 //
-// Provides verification for common by-value semantics
+// Provides generic semantics verification
 //
 *******************************************************************************/
 
-#ifndef __SYCLCTS_TESTS_COMMON_COMMON_BY_VALUE_H
-#define __SYCLCTS_TESTS_COMMON_COMMON_BY_VALUE_H
+#ifndef __SYCLCTS_TESTS_COMMON_SEMANTICS_H
+#define __SYCLCTS_TESTS_COMMON_SEMANTICS_H
 
 #include "common.h"
 
 #include <array>
 #include <string>
 #include <type_traits>
 
-namespace {
+namespace common_semantics {
 
 /**
  * @brief Check equality-comparable operations on the host side
  */
 template <typename T>
-void check_equality_comparable_generic(sycl_cts::util::logger& log, const T& a,
+void check_on_host(sycl_cts::util::logger& log, const T& a,
                                        const std::string& testName) {
   /** check for reflexivity
    */
@@ -64,11 +64,43 @@ void check_equality_comparable_generic(sycl_cts::util::logger& log, const T& a,
   }
 }
 
+/**
+ * @brief Check equality-comparable operations on the host side with extra
+ *        checks for symmetry
+ */
+template <typename T>
+void check_on_host(sycl_cts::util::logger& log, const T& a,
+                                       const T& other,
+                                       const std::string& testName) {
+    check_on_host(log, a, testName);
+
+    /** extra checks for symmetry (comparsion with other)
+     */
+    if (a == other) {
+      FAIL(log, (testName +
+                 " is not equality-comparable (operator==, different value)"));
+    }
+    if (other == a) {
+      FAIL(log, (testName +
+                 " is not equality-comparable (operator== symmetry failed," +
+                 " different value)"));
+    }
+    if (!(a != other)) {
+      FAIL(log, (testName +
+                 " is not equality-comparable (operator!=, different value)"));
+    }
+    if (!(other != a)) {
+      FAIL(log, (testName +
+                 " is not equality-comparable (operator!= symmetry failed," +
+                 " different value)"));
+    }
+}
+
 /**
  * @brief Check equality-comparable operations on the device side
  */
 template <typename T>
-class equality_comparable_on_device
+class on_device_checker
 {
   /**
    * @brief Provides a safe index for checking an operation
@@ -93,9 +125,8 @@ class equality_comparable_on_device
 
 public:
   template <typename kernelT>
-  static void check_on_kernel(sycl_cts::util::logger& log,
-                              const std::array<T, 2>& items,
-                              const std::string& testName) {
+  static void run(sycl_cts::util::logger& log, const std::array<T, 2>& items,
+                  const std::string& testName) {
     // Store comparison results from kernel into a success array
     success_array_t success;
     std::fill(std::begin(success), std::end(success), false);
@@ -222,4 +253,4 @@ class equality_comparable_on_device
 
 }  // namespace
 
-#endif  // __SYCLCTS_TESTS_COMMON_COMMON_BY_VALUE_H
+#endif  // __SYCLCTS_TESTS_COMMON_SEMANTICS_H

tests/group/group_constructors.cpp

@@ -7,7 +7,7 @@
 *******************************************************************************/
 
 #include "../common/common.h"
-#include "../common/common_by_value.h"
+#include "../common/common_semantics.h"
 #include "../common/invoke.h"
 #include "../../util/array.h"
 

tests/group/group_equality.cpp

@@ -7,7 +7,7 @@
 *******************************************************************************/
 
 #include "../common/common.h"
-#include "../common/common_by_value.h"
+#include "../common/common_semantics.h"
 #include "../common/invoke.h"
 
 #define TEST_NAME group_equality
@@ -43,14 +43,13 @@ class TEST_NAME : public util::test_base {
           store_instances<numItems, invoke_group<numDims, setup_kernel_t>>();
 
       // Check nd_item equality operator on the device side
-      equality_comparable_on_device<item_t>::
-          template check_on_kernel<group_equality_kernel<numDims>>(
+      common_semantics::on_device_checker<item_t>::template
+          run<group_equality_kernel<numDims>>(
               log, items, "group " + std::to_string(numDims) + " (device)");
 
       // Check group equality operator on the host side
-      check_equality_comparable_generic(log, items[0],
-                                        "group " + std::to_string(numDims) +
-                                        " (host)");
+      common_semantics::check_on_host(
+          log, items[0], "group " + std::to_string(numDims) + " (host)");
     }
   }
 
@@ -66,4 +65,4 @@ class TEST_NAME : public util::test_base {
 // construction of this proxy will register the above test
 util::test_proxy<TEST_NAME> proxy;
 
-}  // namespace TEST_NAME
+}  // namespace TEST_NAMESPACE

tests/h_item/h_item_equality.cpp

@@ -7,7 +7,7 @@
 *******************************************************************************/
 
 #include "../common/common.h"
-#include "../common/common_by_value.h"
+#include "../common/common_semantics.h"
 
 #include <array>
 #include <string>
@@ -75,7 +75,7 @@ class TEST_NAME : public util::test_base {
 
         // Retrieve two h_item objects and store them
         sycl::buffer<item_t> itemBuf(items.data(),
-                                         sycl::range<1>(items.size()));
+                                     sycl::range<1>(items.size()));
         testQueue.submit([&](sycl::handler& cgh) {
           auto itemAcc =
               itemBuf.template get_access<sycl::access_mode::write>(cgh);
@@ -90,7 +90,7 @@ class TEST_NAME : public util::test_base {
 
         // Perform comparisons on the stored h_item objects
         sycl::buffer<bool> successBuf(success.data(),
-                                          sycl::range<1>(success.size()));
+                                      sycl::range<1>(success.size()));
         testQueue.submit([&](sycl::handler& cgh) {
           auto itemAcc =
               itemBuf.template get_access<sycl::access_mode::read>(cgh);
@@ -126,8 +126,8 @@ class TEST_NAME : public util::test_base {
       }
 
       // Check h_item equality operator
-      check_equality_comparable_generic(log, items[0],
-                                        "h_item " + std::to_string(numDims));
+      common_semantics::check_on_host(log, items[0],
+                                      "h_item " + std::to_string(numDims));
       CHECK_VALUE(log, success[static_cast<size_t>(current_check::equal_self)],
                   true, numDims);
       CHECK_VALUE(log,