diff --git a/intel_test_drivers/case.json b/intel_test_drivers/case.json
index 21b7474de..c69989679 100644
--- a/intel_test_drivers/case.json
+++ b/intel_test_drivers/case.json
@@ -89,16 +89,21 @@
"binary" : "test_math_builtin_api",
"folder" : "math_builtin_api"
},
- "vector_OPERATORS_int64_t" : {
- "source" : "vector_operators_std__int64_t",
- "binary" : "test_vector_operators",
- "folder" : "vector_operators"
+ "exceptions_make_error_code" : {
+ "source" : "exceptions_make_error_code",
+ "binary" : "test_exceptions",
+ "folder" : "exceptions"
},
"usm_api_copy_queue_single_event" : {
"source" : "usm_api_copy_queue_single_event",
"binary" : "test_usm",
"folder" : "usm"
},
+ "vector_OPERATORS_int64_t" : {
+ "source" : "vector_operators_std__int64_t",
+ "binary" : "test_vector_operators",
+ "folder" : "vector_operators"
+ },
"atomic_api_64_base" : {
"source" : "atomic_api_64_base",
"binary" : "test_atomic",
@@ -284,6 +289,11 @@
"binary" : "test_vector_api",
"folder" : "vector_api"
},
+ "specialization_constants_same_command_group_fp16" : {
+ "source" : "specialization_constants_same_command_group_fp16",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"vector_swizzles_half" : {
"source" : "vector_swizzles_sycl__half",
"binary" : "test_vector_swizzles",
@@ -449,6 +459,11 @@
"binary" : "test_specialization_constants",
"folder" : "specialization_constants"
},
+ "specialization_constants_exceptions_throwing_core" : {
+ "source" : "specialization_constants_exceptions_throwing_core",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"group_constructors" : {
"source" : "group_constructors",
"binary" : "test_group",
@@ -489,6 +504,11 @@
"binary" : "test_queue",
"folder" : "queue"
},
+ "specialization_constants_exceptions_throwing_fp64" : {
+ "source" : "specialization_constants_exceptions_throwing_fp64",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"device_selector_constructors" : {
"source" : "device_selector_constructors",
"binary" : "test_device_selector",
@@ -624,26 +644,26 @@
"binary" : "test_usm",
"folder" : "usm"
},
- "queue_properties" : {
- "source" : "queue_properties",
- "binary" : "test_queue",
- "folder" : "queue"
- },
"accessor_api_local_fp16" : {
"source" : "accessor_api_local_fp16",
"binary" : "test_accessor",
"folder" : "accessor"
},
- "vector_LOAD_STORE_int" : {
- "source" : "vector_load_store_int",
- "binary" : "test_vector_load_store",
- "folder" : "vector_load_store"
+ "queue_properties" : {
+ "source" : "queue_properties",
+ "binary" : "test_queue",
+ "folder" : "queue"
},
"specialization_constants_class_with_member_fun" : {
"source" : "specialization_constants_class_with_member_fun",
"binary" : "test_specialization_constants",
"folder" : "specialization_constants"
},
+ "vector_LOAD_STORE_int" : {
+ "source" : "vector_load_store_int",
+ "binary" : "test_vector_load_store",
+ "folder" : "vector_load_store"
+ },
"vector_LOAD_STORE_unsigned_short" : {
"source" : "vector_load_store_unsigned_short",
"binary" : "test_vector_load_store",
@@ -704,6 +724,11 @@
"binary" : "test_vector_load_store",
"folder" : "vector_load_store"
},
+ "specialization_constants_via_handler_fp16" : {
+ "source" : "specialization_constants_via_handler_fp16",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"queue_api" : {
"source" : "queue_api",
"binary" : "test_queue",
@@ -914,6 +939,11 @@
"binary" : "test_nd_item",
"folder" : "nd_item"
},
+ "specialization_constants_via_handler_core" : {
+ "source" : "specialization_constants_via_handler_core",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"hierarchical_functor" : {
"source" : "hierarchical_functor",
"binary" : "test_hierarchical",
@@ -1149,6 +1179,11 @@
"binary" : "test_accessor",
"folder" : "accessor"
},
+ "specialization_constants_same_command_group_fp64" : {
+ "source" : "specialization_constants_same_command_group_fp64",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"specialization_constants_same_name_inter_link_1st_tu_fp16" : {
"source" : "specialization_constants_same_name_inter_link_1st_tu_fp16",
"binary" : "test_specialization_constants",
@@ -1164,6 +1199,11 @@
"binary" : "test_queue",
"folder" : "queue"
},
+ "specialization_constants_exceptions_throwing_fp16" : {
+ "source" : "specialization_constants_exceptions_throwing_fp16",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"usm_api_memcpy_handler_no_events" : {
"source" : "usm_api_memcpy_handler_no_events",
"binary" : "test_usm",
@@ -1184,16 +1224,16 @@
"binary" : "test_vector_swizzles",
"folder" : "vector_swizzles"
},
- "usm_malloc_host" : {
- "source" : "usm_malloc_host",
- "binary" : "test_usm",
- "folder" : "usm"
- },
"vector_ALIAS_double" : {
"source" : "vector_alias_double",
"binary" : "test_vector_alias",
"folder" : "vector_alias"
},
+ "usm_malloc_host" : {
+ "source" : "usm_malloc_host",
+ "binary" : "test_usm",
+ "folder" : "usm"
+ },
"specialization_constants_same_name_inter_link_2nd_tu_fp16" : {
"source" : "specialization_constants_same_name_inter_link_2nd_tu_fp16",
"binary" : "test_specialization_constants",
@@ -1254,16 +1294,16 @@
"binary" : "test_hierarchical",
"folder" : "hierarchical"
},
- "device_constructors" : {
- "source" : "device_constructors",
- "binary" : "test_device",
- "folder" : "device"
- },
"specialization_constants_same_name_stress_core" : {
"source" : "specialization_constants_same_name_stress_core",
"binary" : "test_specialization_constants",
"folder" : "specialization_constants"
},
+ "device_constructors" : {
+ "source" : "device_constructors",
+ "binary" : "test_device",
+ "folder" : "device"
+ },
"program_info" : {
"source" : "program_info",
"binary" : "test_program",
@@ -1489,6 +1529,11 @@
"binary" : "test_nd_item",
"folder" : "nd_item"
},
+ "specialization_constants_same_command_group_core" : {
+ "source" : "specialization_constants_same_command_group_core",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"vector_swizzles_int8_t" : {
"source" : "vector_swizzles_int8_t",
"binary" : "test_vector_swizzles",
@@ -1744,6 +1789,11 @@
"binary" : "test_vector_constructors",
"folder" : "vector_constructors"
},
+ "specialization_constants_via_handler_fp64" : {
+ "source" : "specialization_constants_via_handler_fp64",
+ "binary" : "test_specialization_constants",
+ "folder" : "specialization_constants"
+ },
"vector_SWIZZLE_ASSIGNMENT_int8_t" : {
"source" : "vector_swizzle_assignment_std__int8_t",
"binary" : "test_vector_swizzle_assignment",
diff --git a/intel_test_drivers/config/TEMPLATE_exceptions.xml b/intel_test_drivers/config/TEMPLATE_exceptions.xml
new file mode 100644
index 000000000..36d2e4c8f
--- /dev/null
+++ b/intel_test_drivers/config/TEMPLATE_exceptions.xml
@@ -0,0 +1,9 @@
+
+
+
+
+
+
+
diff --git a/sycl_cts.xml b/sycl_cts.xml
index eefb56812..7d4961023 100644
--- a/sycl_cts.xml
+++ b/sycl_cts.xml
@@ -215,12 +215,22 @@
+
+
+
+
+
+
+
+
+
+
diff --git a/sycl_cts_light.xml b/sycl_cts_light.xml
index bba38c532..d32a97b04 100644
--- a/sycl_cts_light.xml
+++ b/sycl_cts_light.xml
@@ -201,12 +201,22 @@
+
+
+
+
+
+
+
+
+
+
diff --git a/tests/accessor/accessor_api_image_common.h b/tests/accessor/accessor_api_image_common.h
index cb6605cd7..f09ece44c 100644
--- a/tests/accessor/accessor_api_image_common.h
+++ b/tests/accessor/accessor_api_image_common.h
@@ -17,6 +17,7 @@
#include
#include
#include
+#include
namespace {
@@ -407,35 +408,25 @@ T read_image_acc(const sycl::accessor
-T read_image_acc_sampled(const sycl::accessor &acc,
- const sycl::sampler& smpl,
- sycl::id idx,
- sycl::range,
- acc_coord_tag::use_int) {
- return acc.read(image_access::get_int(idx), smpl);
-}
-template
-T read_image_acc_sampled(const sycl::accessor &acc,
- const sycl::sampler& smpl,
- sycl::id idx,
- sycl::range,
- acc_coord_tag::use_float) {
- return acc.read(image_access::get_float(idx), smpl);
-}
-template
+ sycl::access_mode mode, typename coordT>
T read_image_acc_sampled(const sycl::accessor &acc,
const sycl::sampler& smpl,
sycl::id idx,
sycl::range range,
- coordT coordTag) {
- const auto pixelTag = acc_coord_tag::get_pixel_tag(coordTag);
- const auto& coords = image_access::get_normalized(pixelTag, idx, range);
- return acc.read(coords, smpl);
+ const coordT& coordTag) {
+ if constexpr (std::is_same_v) {
+ // Verify read using integer unnormalized coordinates
+ return acc.read(image_access::get_int(idx), smpl);
+ } else if constexpr (std::is_same_v) {
+ // Verify read using floating point unnormalized coordinates
+ return acc.read(image_access::get_float(idx), smpl);
+ } else {
+ // Verify read using normalized coordinates
+ const auto pixelTag = acc_coord_tag::get_pixel_tag(coordTag);
+ const auto& coords = image_access::get_normalized(pixelTag, idx, range);
+ return acc.read(coords, smpl);
+ }
}
-
template
T read_image_acc_sampled(const sycl::accessor &acc,
@@ -945,6 +936,7 @@ class image_accessor_api_sampled_r {
return (v000 + v001 + v010 + v011 + v100 + v101 + v110 + v111) / 8;
}
+ template
T get_expected_value(image_id_t idx) const {
const bool useLinear =
m_sampler.filtering_mode == sycl::filtering_mode::linear;
@@ -952,7 +944,48 @@ class image_accessor_api_sampled_r {
if (!useLinear) {
return get_expected_value_nearest(idx);
}
- return get_expected_value_linear(idx);
+ constexpr bool useUpper =
+ std::is_same_v;
+ if constexpr (!useUpper) {
+ // Use simplified equation for lower coordinate values
+ return get_expected_value_linear(idx);
+ } else {
+ /** Upper value is exactly 1 ULP lower than the lower value for the next
+ * coordinate. We can ignore this difference because:
+ * - there is no actual precision requirements defined for linear
+ * filtration mode in OpenCL spec.
+ * - we have data values pre-defined for floating type to be relatively
+ * not too big, so error propagation is relatively small
+ *
+ * Currently we have the coordinate values
+ * - "u" as the lower one and
+ * - "u + 1 - 1ULP" as the upper one
+ * for floating-point coordinates. Because it would be valuable to verify
+ * also values
+ * - "u - 0.5" and
+ * - "u + 0.5 - 1 ULP"
+ * as the border values for texel selection according to the OpenCL spec:
+ * i0 = address_mode((int)floor(u - 0.5))
+ * j0 = address_mode((int)floor(v - 0.5))
+ * k0 = address_mode((int)floor(w - 0.5))
+ * i1 = address_mode((int)floor(u - 0.5) + 1)
+ * j1 = address_mode((int)floor(v - 0.5) + 1)
+ * k1 = address_mode((int)floor(w - 0.5) + 1)
+ * we may need to provide an exact reference values someday.
+ *
+ * During future implementation of such reference functions we may gain
+ * accuracy from using Priest's compensated summation to avoid possible
+ * catastrophic cancellation.
+ * See
+ * Douglas M. Priest. "On Properties of Floating Point Arithmetics:
+ * Numerical Stability and the Cost of Accurate Computations."
+ * PhD thesis, Mathematics Department, University of California,
+ * Berkeley, CA, USA, November 1992. 126 pp.
+ * ftp://ftp.icsi.berkeley.edu/pub/theory/priest-thesis.ps.Z
+ * for details
+ */
+ return get_expected_value_linear(idx + 1);
+ }
}
/**
@@ -1033,8 +1066,9 @@ class image_accessor_api_sampled_r {
}
template
- bool check_read(sycl::id idx, const T& expected) const {
- T elem =
+ bool check_read(sycl::id idx) const {
+ const T expected = get_expected_value(idx);
+ const T elem =
read_image_acc_sampled(m_acc, m_sampler.instance, idx, m_range,
coordT{});
@@ -1065,8 +1099,6 @@ class image_accessor_api_sampled_r {
m_verificationRange(verificationRange) {}
void operator()(image_id_t idx) const {
- auto expected = get_expected_value(idx);
-
/** check coordinates with sampler read syntax
*/
const bool useNormalized =
@@ -1074,14 +1106,14 @@ class image_accessor_api_sampled_r {
sycl::coordinate_normalization_mode::normalized;
if (useNormalized) {
const bool worksForLower =
- check_read(idx, expected);
+ check_read(idx);
if (worksForLower)
- check_read(idx, expected);
+ check_read(idx);
} else {
const bool worksForInteger =
- check_read(idx, expected);
+ check_read(idx);
if (worksForInteger)
- check_read(idx, expected);
+ check_read(idx);
}
}
};
diff --git a/tests/exceptions/CMakeLists.txt b/tests/exceptions/CMakeLists.txt
new file mode 100644
index 000000000..82f462065
--- /dev/null
+++ b/tests/exceptions/CMakeLists.txt
@@ -0,0 +1,3 @@
+file(GLOB test_cases_list *.cpp)
+
+add_cts_test(${test_cases_list})
diff --git a/tests/exceptions/exceptions.h b/tests/exceptions/exceptions.h
new file mode 100644
index 000000000..9bfae60fd
--- /dev/null
+++ b/tests/exceptions/exceptions.h
@@ -0,0 +1,38 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Common code for exceptions tests
+//
+*******************************************************************************/
+
+#ifndef __SYCL_CTS_TEST_EXCEPTIONS_H
+#define __SYCL_CTS_TEST_EXCEPTIONS_H
+
+#include "../common/common.h"
+#include
+
+/** @brief Provide common code for all tests for exceptions
+ */
+namespace {
+
+inline const std::vector all_err_codes{
+ sycl::errc::success,
+ sycl::errc::runtime,
+ sycl::errc::kernel,
+ sycl::errc::accessor,
+ sycl::errc::nd_range,
+ sycl::errc::event,
+ sycl::errc::kernel_argument,
+ sycl::errc::build,
+ sycl::errc::invalid,
+ sycl::errc::memory_allocation,
+ sycl::errc::platform,
+ sycl::errc::profiling,
+ sycl::errc::feature_not_supported,
+ sycl::errc::kernel_not_supported,
+ sycl::errc::backend_mismatch};
+
+} // namespace
+
+#endif // __SYCL_CTS_TEST_EXCEPTIONS_H
diff --git a/tests/exceptions/exceptions_make_error_code.cpp b/tests/exceptions/exceptions_make_error_code.cpp
new file mode 100644
index 000000000..e990c731a
--- /dev/null
+++ b/tests/exceptions/exceptions_make_error_code.cpp
@@ -0,0 +1,117 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Provides tests for sycl::make_error_code function
+//
+*******************************************************************************/
+
+#include "exceptions.h"
+
+#define TEST_NAME exceptions_make_error_code
+
+namespace TEST_NAMESPACE {
+
+using namespace sycl_cts;
+
+/** @brief Provide verification for sycl::make_error_code function
+ * @param err_c Error code from sycl::errc enumeration
+ * @param log sycl_cts::util::logger class object
+ */
+void check_sycl_working(sycl::errc err_c, util::logger &log) {
+ auto make_err_c_result{sycl::make_error_code(err_c)};
+
+ CHECK_TYPE(log, make_err_c_result, std::error_code());
+ if (!noexcept(sycl::make_error_code(err_c))) {
+ FAIL(log, "sycl::make_error_code function are not marked as \"noexcept\"");
+ }
+ if (make_err_c_result != err_c) {
+ FAIL(
+ log,
+ "sycl::make_error_code function's error code are not equal to provided "
+ "error code from sycl::errc enumeration");
+ }
+}
+
+/** @brief Provide verification for same work std::error_code and
+ * sycl::make_error_code
+ * @param err_c Error code from sycl::errc enumeration
+ * @param log sycl_cts::util::logger class object
+ */
+void compare_sycl_and_std_working(sycl::errc err_c, util::logger &log) {
+ std::error_code err_c_result{static_cast(err_c), sycl::sycl_category()};
+ auto make_err_c_result{sycl::make_error_code(err_c)};
+
+ if (err_c_result.value() != make_err_c_result.value()) {
+ FAIL(log,
+ "error code value that received from std::error_code not equal to "
+ "value that received from sycl::make_error_code ");
+ }
+ if (err_c_result.message().empty()) {
+ FAIL(log, "error message from std::error_code are empty");
+ }
+ if (make_err_c_result.message().empty()) {
+ FAIL(log, "error message from sycl::make_error_code are empty");
+ }
+ if (err_c_result.message() != make_err_c_result.message()) {
+ FAIL(log,
+ "error message from std::error_code not equal to error message from "
+ "sycl::make_error_code");
+ }
+ if (err_c_result.default_error_condition() !=
+ std::error_condition(static_cast(err_c), sycl::sycl_category())) {
+ FAIL(log,
+ "default error condition that received from std::error_code not equal "
+ "to error condition that received from std::error_condition");
+ }
+ if (make_err_c_result.default_error_condition() !=
+ std::error_condition(static_cast(err_c), sycl::sycl_category())) {
+ FAIL(log,
+ "default error condition that received from sycl::make_error_code not "
+ "equal to error condition that received from std::error_condition");
+ }
+ if (err_c_result.category() != sycl::sycl_category()) {
+ FAIL(log,
+ "error category that received from std::error_code not equal to "
+ "sycl::sycl_category");
+ }
+ if (make_err_c_result.category() != sycl::sycl_category()) {
+ FAIL(log,
+ "error category that received from sycl::make_error_code not equal to "
+ "sycl::sycl_category");
+ }
+}
+
+/** Test instance
+ */
+class TEST_NAME : public util::test_base {
+ public:
+ /** return information about this test
+ */
+ void get_info(test_base::info &out) const override {
+ set_test_info(out, TOSTRING(TEST_NAME), TEST_FILE);
+ }
+
+ /** execute the test
+ */
+ void run(util::logger &log) override {
+ try {
+ for (auto &err_c : all_err_codes) {
+ compare_sycl_and_std_working(err_c, log);
+ check_sycl_working(err_c, log);
+ }
+ } catch (const sycl::exception &e) {
+ log_exception(log, e);
+ std::string errorMsg{"a SYCL exception was caught: " +
+ std::string(e.what())};
+ FAIL(log, errorMsg);
+ } catch (const std::exception &e) {
+ std::string errorMsg{"an exception was caught: " + std::string(e.what())};
+ FAIL(log, errorMsg);
+ }
+ }
+};
+
+util::test_proxy proxy;
+
+} // namespace TEST_NAMESPACE
diff --git a/tests/opencl_interop/opencl_interop_constructors.cpp b/tests/opencl_interop/opencl_interop_constructors.cpp
index 6a37e4234..ccd32d9a3 100644
--- a/tests/opencl_interop/opencl_interop_constructors.cpp
+++ b/tests/opencl_interop/opencl_interop_constructors.cpp
@@ -6,8 +6,10 @@
//
*******************************************************************************/
+#ifdef SYCL_BACKEND_OPENCL
#include "../../util/opencl_helper.h"
#include "../../util/test_base_opencl.h"
+#endif
#include "../common/common.h"
#define TEST_NAME opencl_interop_constructors
@@ -16,13 +18,16 @@ namespace opencl_interop_constructors__ {
using namespace sycl_cts;
class buffer_interop_constructor_kernel;
-class image_interop_constructor_kernel_default_event;
-class image_interop_constructor_kernel_provided_event;
-class sampler_interop_constructor_kernel;
/** tests the constructors for OpenCL inter-op
*/
-class TEST_NAME : public sycl_cts::util::test_base_opencl {
+class TEST_NAME :
+#ifdef SYCL_BACKEND_OPENCL
+ public sycl_cts::util::test_base_opencl
+#else
+ public util::test_base
+#endif
+{
public:
/** return information about this test
*/
@@ -33,15 +38,18 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
/** execute this test
*/
void run(util::logger &log) override {
+#ifdef SYCL_BACKEND_OPENCL
try {
+ auto queue = util::get_cts_object::queue();
+ if (queue.get_backend() != sycl::backend::opencl) {
+ log.note("Interop part is not supported on non-OpenCL backend types");
+ return;
+ }
+
cts_selector ctsSelector;
const auto ctsContext = util::get_cts_object::context(ctsSelector);
const auto ctsDevice = ctsContext.get_devices()[0];
- if (ctsContext.is_host()) {
- log.note("OpenCL interop doesn't work on host");
- return;
- }
std::string kernelSource = R"(
__kernel void opencl_interop_constructors_kernel(__global float *input)
@@ -51,23 +59,25 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
)";
std::string programBinaryFile =
"opencl_interop_constructors.bin";
- /** check platform (cl_platform_id) constructor
+ /** check make_platform (cl_platform_id)
*/
{
- sycl::platform platform(m_cl_platform_id);
+ sycl::platform platform = sycl::make_platform(m_cl_platform_id);
- cl_platform_id interopPlatformID = platform.get();
+ cl_platform_id interopPlatformID =
+ sycl::get_native(platform);
if (interopPlatformID != m_cl_platform_id) {
FAIL(log, "platform was not constructed correctly");
}
}
- /** check device (cl_device_id) constructor
+ /** check make_device (cl_device_id)
*/
{
- sycl::device device(m_cl_device_id);
+ sycl::device device = sycl::make_device(m_cl_device_id);
- cl_device_id interopDeviceID = device.get();
+ cl_device_id interopDeviceID =
+ sycl::get_native(device);
if (interopDeviceID != m_cl_device_id) {
FAIL(log, "device was not constructed correctly");
}
@@ -76,12 +86,13 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
}
}
- /** check context (cl_context) constructor
+ /** check make_context (cl_context)
*/
{
- sycl::context context(m_cl_context);
+ sycl::context context = sycl::make_context(m_cl_context);
- cl_context interopContext = context.get();
+ cl_context interopContext =
+ sycl::get_native(context);
if (interopContext != m_cl_context) {
FAIL(log, "context was not constructed correctly");
}
@@ -90,13 +101,14 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
}
}
- /** check context (cl_context, async_handler) constructor
+ /** check make_context (cl_context, async_handler)
*/
{
cts_async_handler asyncHandler;
- sycl::context context(m_cl_context, asyncHandler);
+ sycl::context context = sycl::make_context(m_cl_context, asyncHandler);
- cl_context interopContext = context.get();
+ cl_context interopContext =
+ sycl::get_native(context);
if (interopContext != m_cl_context) {
FAIL(log, "context was not constructed correctly");
}
@@ -105,12 +117,13 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
}
}
- /** check queue (cl_command_queue, const context&) constructor
+ /** check make_queue (cl_command_queue, const context&)
*/
{
- sycl::queue queue(m_cl_command_queue, ctsContext);
+ sycl::queue queue = sycl::make_queue(m_cl_command_queue, ctsContext);
- cl_command_queue interopQueue = queue.get();
+ cl_command_queue interopQueue =
+ sycl::get_native(queue);
if (interopQueue != m_cl_command_queue) {
FAIL(log, "queue was not constructed correctly");
}
@@ -118,7 +131,7 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
/** check that queue copy constructor preserve the same OpenCL queue
*/
sycl::queue queueCopy(queue);
- auto clQueueCopy = queueCopy.get();
+ auto clQueueCopy = sycl::get_native(queueCopy);
if (interopQueue != clQueueCopy) {
FAIL(log, "queue destination was not copy constructed correctly");
}
@@ -132,14 +145,15 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
}
}
- /** check queue (cl_command_queue, const context&, async_handler)
- * constructor
+ /** check make_queue (cl_command_queue, const context&, async_handler)
*/
{
cts_async_handler asyncHandler;
- sycl::queue queue(m_cl_command_queue, ctsContext, asyncHandler);
+ sycl::queue queue =
+ sycl::make_queue(m_cl_command_queue, ctsContext, asyncHandler);
- cl_command_queue interopQueue = queue.get();
+ cl_command_queue interopQueue =
+ sycl::get_native(queue);
if (interopQueue != m_cl_command_queue) {
FAIL(log, "queue was not constructed correctly");
}
@@ -148,18 +162,25 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
}
}
- /** check program (context, cl_program) constructor
+ /** check make_kernel_bundle (cl_program, context)
*/
{
cl_program clProgram{};
- if (online_compiler_supported(ctsDevice.get(), log)) {
- if (!create_built_program(kernelSource, ctsContext.get(),
- ctsDevice.get(), clProgram, log)) {
+ if (online_compiler_supported(
+ sycl::get_native(ctsDevice), log)) {
+ if (!create_built_program(
+ kernelSource,
+ sycl::get_native(ctsContext),
+ sycl::get_native(ctsDevice), clProgram,
+ log)) {
FAIL(log, "create_built_program failed");
}
} else {
- if (!create_program_with_binary(programBinaryFile, ctsContext.get(),
- ctsDevice.get(), clProgram, log)) {
+ if (!create_program_with_binary(
+ programBinaryFile,
+ sycl::get_native(ctsContext),
+ sycl::get_native(ctsDevice), clProgram,
+ log)) {
std::string errorMsg =
"create_program_with_binary failed.";
errorMsg +=
@@ -169,9 +190,12 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
}
}
- sycl::program program(ctsContext, clProgram);
+ auto kernel_bundle =
+ sycl::make_kernel_bundle(
+ clProgram, ctsContext);
- cl_program interopProgram = program.get();
+ cl_program interopProgram =
+ sycl::get_native(kernel_bundle);
if (interopProgram != clProgram) {
FAIL(log, "program was not constructed correctly");
}
@@ -180,18 +204,25 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
}
}
- /** check kernel (cl_kernel, const context&) constructor
+ /** check make_kernel (cl_kernel, const context&)
*/
{
cl_program clProgram{};
- if (online_compiler_supported(ctsDevice.get(), log)) {
- if (!create_built_program(kernelSource, ctsContext.get(),
- ctsDevice.get(), clProgram, log)) {
+ if (online_compiler_supported(
+ sycl::get_native(ctsDevice), log)) {
+ if (!create_built_program(
+ kernelSource,
+ sycl::get_native(ctsContext),
+ sycl::get_native(ctsDevice), clProgram,
+ log)) {
FAIL(log, "create_built_program failed");
}
} else {
- if (!create_program_with_binary(programBinaryFile, ctsContext.get(),
- ctsDevice.get(), clProgram, log)) {
+ if (!create_program_with_binary(
+ programBinaryFile,
+ sycl::get_native(ctsContext),
+ sycl::get_native(ctsDevice), clProgram,
+ log)) {
std::string errorMsg =
"create_program_with_binary failed.";
errorMsg +=
@@ -207,9 +238,10 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
FAIL(log, "create_kernel failed");
}
- sycl::kernel kernel(clKernel, ctsContext);
+ sycl::kernel kernel = sycl::make_kernel(clKernel, ctsContext);
- cl_kernel interopKernel = kernel.get();
+ cl_kernel interopKernel =
+ sycl::get_native(kernel);
if (interopKernel != clKernel) {
FAIL(log, "kernel was not constructed correctly");
}
@@ -218,7 +250,7 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
}
}
- /** check buffer (cl_mem, contex) constructor
+ /** check make_buffer (cl_mem, contex)
*/
{
const size_t size = 32;
@@ -228,13 +260,15 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
auto queue = util::get_cts_object::queue(ctsSelector);
cl_mem clBuffer = clCreateBuffer(
- queue.get_context().get(), CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR,
- size * sizeof(int), data, &error);
+ sycl::get_native(queue.get_context()),
+ CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, size * sizeof(int), data,
+ &error);
if (!CHECK_CL_SUCCESS(log, error)) {
FAIL(log, "create buffer failed");
}
- sycl::buffer buffer(clBuffer, queue.get_context());
+ sycl::buffer buffer =
+ sycl::make_buffer(clBuffer, queue.get_context());
// calculate element count, size and range for the interop buffer
sycl::range<1> interopRange{size};
@@ -247,19 +281,19 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
"opencl buffer was not interop constructed properly. "
"(is_sub_buffer) ");
}
- if (buffer.get_size() != interopSize) {
+ if (buffer.byte_size() != interopSize) {
FAIL(log,
- "opencl buffer was not interop constructed properly. (get_size) ");
+ "opencl buffer was not interop constructed properly. "
+ "(byte_size) ");
}
if (buffer.get_range() != interopRange) {
FAIL(
log,
"opencl buffer was not interop constructed properly. (get_range) ");
}
- if (buffer.get_count() != size) {
- FAIL(
- log,
- "opencl buffer was not interop constructed properly. (get_count) ");
+ if (buffer.size() != size) {
+ FAIL(log,
+ "opencl buffer was not interop constructed properly. (size) ");
}
queue.submit([&](sycl::handler &handler) {
@@ -278,7 +312,7 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
queue.wait_and_throw();
}
- /** check buffer (cl_mem, context, event) constructor
+ /** check make_buffer (cl_mem, context, event)
*/
{
const size_t size = 32;
@@ -294,13 +328,15 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
});
cl_mem clBuffer = clCreateBuffer(
- queue.get_context().get(), CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR,
- size * sizeof(int), data, &error);
+ sycl::get_native(queue.get_context()),
+ CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, size * sizeof(int), data,
+ &error);
if (!CHECK_CL_SUCCESS(log, error)) {
FAIL(log, "create buffer failed");
}
- sycl::buffer buffer(clBuffer, queue.get_context(), event);
+ sycl::buffer buffer =
+ sycl::make_buffer(clBuffer, queue.get_context(), event);
// calculate element count, size and range for the interop buffer
sycl::range<1> interopRange{size};
@@ -313,19 +349,19 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
"opencl buffer was not interop constructed properly. "
"(is_sub_buffer) ");
}
- if (buffer.get_size() != interopSize) {
+ if (buffer.byte_size() != interopSize) {
FAIL(log,
- "opencl buffer was not interop constructed properly. (get_size) ");
+ "opencl buffer was not interop constructed properly. "
+ "(byte_size) ");
}
if (buffer.get_range() != interopRange) {
FAIL(
log,
"opencl buffer was not interop constructed properly. (get_range) ");
}
- if (buffer.get_count() != size) {
- FAIL(
- log,
- "opencl buffer was not interop constructed properly. (get_count) ");
+ if (buffer.size() != size) {
+ FAIL(log,
+ "opencl buffer was not interop constructed properly. (size) ");
}
queue.submit([&](sycl::handler &handler) {
@@ -343,129 +379,16 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
queue.wait_and_throw();
}
+ // TODO: add checks for make_sampled_image and make_unsampled_image
- /** check image (cl_mem, const context&, event) constructor
- */
- {
- auto queue = util::get_cts_object::queue(ctsSelector);
- if (!queue.get_device()
- .get_info()) {
- log.note("Device does not support images");
- } else {
- constexpr size_t imageSideSize = 16;
- /* Size is *4 because image data is 4 channels (RGBA) */
- constexpr auto size = imageSideSize * imageSideSize * 4;
- float data[size] = {0.0f};
-
- const auto clContext = queue.get_context().get();
-
- cl_image_format clImageFormat;
- clImageFormat.image_channel_data_type = CL_FLOAT;
- clImageFormat.image_channel_order = CL_RGBA;
-
- cl_image_desc clImageDesc;
- clImageDesc.image_type = CL_MEM_OBJECT_IMAGE2D;
- clImageDesc.image_width = imageSideSize;
- clImageDesc.image_height = imageSideSize;
- clImageDesc.image_depth = 0;
- clImageDesc.image_array_size = 1;
- clImageDesc.image_row_pitch = 0;
- clImageDesc.image_slice_pitch = 0;
- clImageDesc.num_mip_levels = 0;
- clImageDesc.num_samples = 0;
- clImageDesc.buffer = nullptr;
-
- cl_int error = CL_SUCCESS;
- // Check constructing image with defaulted event
- {
- cl_mem clImage = clCreateImage(
- clContext, (CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR),
- &clImageFormat, &clImageDesc, data, &error);
- if (!CHECK_CL_SUCCESS(log, error)) {
- FAIL(log, "create image failed");
- }
-
- sycl::image<2> image(clImage, queue.get_context());
-
- queue.submit([&](sycl::handler &handler) {
- auto accessor =
- image.get_access(handler);
- handler.single_task<
- class image_interop_constructor_kernel_default_event>(
- []() {});
- });
-
- error = clReleaseMemObject(clImage);
- if (!CHECK_CL_SUCCESS(log, error)) {
- FAIL(log, "failed to release OpenCL image");
- }
- }
-
- // Check constructing image with specified event
- {
- sycl::event event;
- cl_mem clImage = clCreateImage(
- clContext, (CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR),
- &clImageFormat, &clImageDesc, data, &error);
- if (!CHECK_CL_SUCCESS(log, error)) {
- FAIL(log, "create image failed");
- }
-
- sycl::image<2> image(clImage, queue.get_context(), event);
-
- queue.submit([&](sycl::handler &handler) {
- auto accessor =
- image.get_access(handler);
- handler.single_task<
- class image_interop_constructor_kernel_provided_event>(
- []() {});
- });
-
- error = clReleaseMemObject(clImage);
- if (!CHECK_CL_SUCCESS(log, error)) {
- FAIL(log, "failed to release OpenCL image");
- }
- }
- queue.wait_and_throw();
- }
- }
-
- /** check sampler (cl_sampler, const context&) constructor
- */
- {
- auto queue = util::get_cts_object::queue(ctsSelector);
- if (!queue.get_device()
- .get_info()) {
- log.note("Device does not support images");
- } else {
- cl_sampler clSampler;
- create_sampler(clSampler, log);
-
- queue.submit([&](sycl::handler &handler) {
- sycl::sampler sampler(clSampler, queue.get_context());
-
- handler.single_task(
- []() {});
- });
-
- cl_int error = clReleaseSampler(clSampler);
- if (!CHECK_CL_SUCCESS(log, error)) {
- FAIL(log, "failed to release OpenCL sampler");
- }
-
- queue.wait_and_throw();
- }
- }
-
- /** check event (cl_event, const context&) constructor
+ /** check make_event (cl_event, const context&)
*/
{
- cl_event clEvent = clCreateUserEvent(ctsContext.get(), nullptr);
- sycl::event event(clEvent, ctsContext);
+ cl_event clEvent = clCreateUserEvent(
+ sycl::get_native(ctsContext), nullptr);
+ sycl::event event = sycl::make_event(clEvent, ctsContext);
- cl_event interopEvent = event.get();
+ cl_event interopEvent = sycl::get_native(event);
if (interopEvent != clEvent) {
FAIL(log, "event was not constructed correctly");
}
@@ -477,6 +400,9 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
"a SYCL exception was caught: " + std::string(e.what());
FAIL(log, errorMsg.c_str());
}
+#else
+ log.note("The test is skipped because OpenCL back-end is not supported");
+#endif // SYCL_BACKEND_OPENCL
}
};
diff --git a/tests/opencl_interop/opencl_interop_get.cpp b/tests/opencl_interop/opencl_interop_get.cpp
index 94f953edc..6cac9eff7 100644
--- a/tests/opencl_interop/opencl_interop_get.cpp
+++ b/tests/opencl_interop/opencl_interop_get.cpp
@@ -6,24 +6,28 @@
//
*******************************************************************************/
+#ifdef SYCL_BACKEND_OPENCL
#include "../../util/opencl_helper.h"
#include "../../util/test_base_opencl.h"
+#endif
#include "../common/common.h"
#define TEST_NAME opencl_interop_get
-struct program_get_kernel {
- void operator()() const {}
-};
-
namespace opencl_interop_get__ {
using namespace sycl_cts;
class event_kernel;
-/** tests the get() methods for OpenCL inter-op
+/** tests the get_native() methods for OpenCL inter-op
*/
-class TEST_NAME : public sycl_cts::util::test_base_opencl {
+class TEST_NAME :
+#ifdef SYCL_BACKEND_OPENCL
+ public sycl_cts::util::test_base_opencl
+#else
+ public util::test_base
+#endif
+{
public:
/** return information about this test
*/
@@ -34,83 +38,75 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
/** execute this test
*/
void run(util::logger &log) override {
+#ifdef SYCL_BACKEND_OPENCL
try {
+ auto queue = util::get_cts_object::queue();
+ if (queue.get_backend() != sycl::backend::opencl) {
+ log.note("Interop part is not supported on non-OpenCL backend types");
+ return;
+ }
cts_selector ctsSelector;
const auto ctsContext = util::get_cts_object::context(ctsSelector);
const auto ctsDevice = ctsContext.get_devices()[0];
- if (ctsContext.is_host()) {
- log.note("OpenCL interop doesn't work on host");
- return;
- }
-
- /** check platform get() method
+ /** check get_native() for platform
*/
{
auto platform = util::get_cts_object::platform(ctsSelector);
- if (!platform.is_host()) {
- auto interopPlatformID = platform.get();
- check_return_type(log, interopPlatformID,
- "sycl::platform::get()");
-
- if (interopPlatformID == 0) {
- FAIL(log,
- "sycl::platform::get() did not return a valid "
- "cl_platform_id");
- }
+ auto interopPlatformID =
+ sycl::get_native(platform);
+ check_return_type(log, interopPlatformID,
+ "get_native(platform)");
+
+ if (interopPlatformID == 0) {
+ FAIL(log,
+ "get_native(platform) did not return a valid "
+ "cl_platform_id");
}
}
- /** check device get() method
+ /** check get_native() for device
*/
{
auto device = util::get_cts_object::device(ctsSelector);
- if (!device.is_host()) {
- auto interopDeviceID = device.get();
- check_return_type(log, interopDeviceID,
- "sycl::device::get()");
-
- if (interopDeviceID == 0) {
- FAIL(log,
- "sycl::device::get() did not return a valid cl_device_id");
- }
+ auto interopDeviceID = sycl::get_native(device);
+ check_return_type(log, interopDeviceID,
+ "get_native(device)");
+
+ if (interopDeviceID == 0) {
+ FAIL(log, "get_native(device) did not return a valid cl_device_id");
}
}
- /** check context get() method
+ /** check get_native() for context
*/
{
auto context = util::get_cts_object::context(ctsSelector);
- if (!context.is_host()) {
- auto interopContext = context.get();
- check_return_type(log, interopContext,
- "sycl::context::get()");
-
- if (interopContext == nullptr) {
- FAIL(log,
- "sycl::context::get() did not return a valid cl_context");
- }
+ auto interopContext = sycl::get_native(context);
+ check_return_type(log, interopContext,
+ "get_native(context)");
+
+ if (interopContext == nullptr) {
+ FAIL(log, "get_native(context) did not return a valid cl_context");
}
}
- /** check queue get() method
+ /** check get_native() for queue
*/
{
auto queue = util::get_cts_object::queue(ctsSelector);
- if (!queue.is_host()) {
- auto interopQueue = queue.get();
- check_return_type(log, interopQueue,
- "sycl::queue::get()");
-
- if (interopQueue == nullptr) {
- FAIL(log,
- "sycl::queue::get() did not return a valid "
- "cl_command_queue");
- }
+ auto interopQueue = sycl::get_native(queue);
+ check_return_type(log, interopQueue,
+ "get_native(queue)");
+
+ if (interopQueue == nullptr) {
+ FAIL(log,
+ "get_native(queue) did not return a valid "
+ "cl_command_queue");
}
}
- /** check program get() method
+ /** check get_native() for kernel_bundle
*/
{
if (!util::get_cts_object::queue(ctsSelector)
@@ -118,42 +114,49 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
.get_info()) {
log.note("online compiler not available -- skipping check");
} else {
- auto program =
- util::get_cts_object::program::compiled(
- ctsContext);
+ auto bundle = sycl::get_kernel_bundle(
+ ctsContext);
if (!program.is_host()) {
- auto interopProgram = program.get();
+ auto interopProgram =
+ sycl::get_native(bundle);
check_return_type(log, interopProgram,
- "sycl::program::get()");
+ "get_native(kernel_bundle)");
if (interopProgram == nullptr) {
- FAIL(
- log,
- "sycl::program::get() did not return a valid cl_program");
+ FAIL(log,
+ "get_native(kernel_bundle) did not return a valid "
+ "cl_program");
}
}
}
}
- /** check kernel get() method
+ /** check get_native() for kernel
*/
{
if (!ctsContext.is_host()) {
cl_program clProgram{};
- if (online_compiler_supported(ctsDevice.get(), log)) {
+ if (online_compiler_supported(
+ sycl::get_native(ctsDevice), log)) {
std::string kernelSource = R"(
__kernel void opencl_interop_get_kernel() {}
)";
- if (!create_built_program(kernelSource, ctsContext.get(),
- ctsDevice.get(), clProgram, log)) {
+ if (!create_built_program(
+ kernelSource,
+ sycl::get_native(ctsContext),
+ sycl::get_native(ctsDevice),
+ clProgram, log)) {
FAIL(log, "create_built_program failed");
}
} else {
std::string programBinaryFile = "opencl_interop_get.bin";
- if (!create_program_with_binary(programBinaryFile, ctsContext.get(),
- ctsDevice.get(), clProgram, log)) {
+ if (!create_program_with_binary(
+ programBinaryFile,
+ sycl::get_native(ctsContext),
+ sycl::get_native(ctsDevice),
+ clProgram, log)) {
std::string errorMsg =
"create_program_with_binary failed.";
errorMsg +=
@@ -169,20 +172,19 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
FAIL(log, "create_kernel failed");
}
- sycl::kernel kernel(clKernel, ctsContext);
+ sycl::kernel kernel = sycl::make_kernel(clKernel, ctsContext);
- auto interopKernel = kernel.get();
+ auto interopKernel = sycl::get_native(kernel);
check_return_type(log, interopKernel,
- "sycl::kernel::get()");
+ "get_native(kernel)");
if (interopKernel == nullptr) {
- FAIL(log,
- "sycl::kernel::get() did not return a valid cl_kernel");
+ FAIL(log, "get_native(kernel) did not return a valid cl_kernel");
}
}
}
- /** check event get() method
+ /** check get_native() for event
*/
{
auto ctsQueue = util::get_cts_object::queue(ctsSelector);
@@ -192,12 +194,11 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
});
if (!event.is_host()) {
- auto interopEvent = event.get();
- check_return_type(log, interopEvent,
- "sycl::event::get()");
+ auto interopEvent = sycl::get_native(event);
+ check_return_type(log, interopEvent, "get_native(event)");
if (interopEvent == nullptr) {
- FAIL(log, "sycl::event::get() did not return a valid cl_event");
+ FAIL(log, "get_native(event) did not return a valid cl_event");
}
}
@@ -210,6 +211,9 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
"a SYCL exception was caught: " + std::string(e.what());
FAIL(log, errorMsg.c_str());
}
+#else
+ log.note("The test is skipped because OpenCL back-end is not supported");
+#endif // SYCL_BACKEND_OPENCL
}
};
diff --git a/tests/opencl_interop/opencl_interop_kernel.cpp b/tests/opencl_interop/opencl_interop_kernel.cpp
index dd5839be3..9e39c997e 100644
--- a/tests/opencl_interop/opencl_interop_kernel.cpp
+++ b/tests/opencl_interop/opencl_interop_kernel.cpp
@@ -6,8 +6,10 @@
//
*******************************************************************************/
+#ifdef SYCL_BACKEND_OPENCL
#include "../../util/opencl_helper.h"
#include "../../util/test_base_opencl.h"
+#endif
#include "../common/common.h"
#define TEST_NAME opencl_interop_kernel
@@ -15,31 +17,6 @@
namespace opencl_interop_kernel__ {
using namespace sycl_cts;
-/** check inter-op types
- */
-template
-using globalPtrType = typename sycl::global_ptr::pointer;
-template
-using constantPtrType = typename sycl::constant_ptr::pointer;
-template
-using localPtrType = typename sycl::local_ptr::pointer;
-template
-using privatePtrType = typename sycl::private_ptr::pointer;
-template
-using globalMultiPtrType = typename sycl::multi_ptr<
- T, sycl::access::address_space::global_space>::pointer;
-template
-using constantMultiPtrType = typename sycl::multi_ptr<
- T, sycl::access::address_space::constant_space>::pointer;
-template
-using localMultiPtrType = typename sycl::multi_ptr<
- T, sycl::access::address_space::local_space>::pointer;
-template
-using privateMultiPtrType = typename sycl::multi_ptr<
- T, sycl::access::address_space::private_space>::pointer;
-template
-using vectorType = typename sycl::vec::vector_t;
-
/**
* @brief Trivially-copyable standard layout custom type
*/
@@ -48,25 +25,15 @@ struct simple_struct {
float b;
};
-// Forward declaration of the kernel
-template
-struct program_kernel_interop {
- void operator()() const {}
-};
-
-/** simple OpenCL test kernel
- */
-const std::string kernelName = "sample";
-std::string kernel_source = R"(
-__kernel void sample(__global float * input)
-{
- input[get_global_id(0)] = get_global_id(0);
-}
-)";
-
/** tests the kernel execution for OpenCL inter-op
*/
-class TEST_NAME : public sycl_cts::util::test_base_opencl {
+class TEST_NAME :
+#ifdef SYCL_BACKEND_OPENCL
+ public sycl_cts::util::test_base_opencl
+#else
+ public util::test_base
+#endif
+{
public:
/** return information about this test
*/
@@ -77,14 +44,15 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
/** execute this test
*/
void run(util::logger &log) override {
+#ifdef SYCL_BACKEND_OPENCL
try {
- cts_selector ctsSelector;
- const auto ctsContext = util::get_cts_object::context(ctsSelector);
-
- if (ctsContext.is_host()) {
- log.note("OpenCL interop doesn't work on host");
+ auto queue = util::get_cts_object::queue();
+ if (queue.get_backend() != sycl::backend::opencl) {
+ log.note("Interop part is not supported on non-OpenCL backend types");
return;
}
+ cts_selector ctsSelector;
+ const auto ctsContext = util::get_cts_object::context(ctsSelector);
{
const size_t bufferSize = 32;
@@ -98,7 +66,8 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
sycl::range<1>(bufferSize));
cl_program clProgram{};
- if (online_compiler_supported(device.get(), log)) {
+ if (online_compiler_supported(
+ sycl::get_native(device), log)) {
std::string kernelSource = R"(
struct simple_struct {
int a;
@@ -112,16 +81,22 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
{}
)";
- if (!create_built_program(kernelSource, context.get(), device.get(),
- clProgram, log)) {
+ if (!create_built_program(
+ kernelSource,
+ sycl::get_native(context),
+ sycl::get_native(device), clProgram,
+ log)) {
FAIL(log, "create_built_program failed");
}
} else {
std::string programBinaryFile =
"opencl_interop_kernel.bin";
- if (!create_program_with_binary(programBinaryFile, context.get(),
- device.get(), clProgram, log)) {
+ if (!create_program_with_binary(
+ programBinaryFile,
+ sycl::get_native(context),
+ sycl::get_native(device), clProgram,
+ log)) {
std::string errorMsg =
"create_program_with_binary failed.";
errorMsg +=
@@ -137,7 +112,7 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
FAIL(log, "create_kernel failed");
}
- sycl::kernel kernel(clKernel, context);
+ sycl::kernel kernel = sycl::make_kernel(clKernel, context);
/** test single_task(kernel)
*/
@@ -185,336 +160,7 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
queue.wait_and_throw();
}
- {
- if (!util::get_cts_object::queue(ctsSelector)
- .get_device()
- .get_info()) {
- log.note("Device does not support images");
- } else {
- static constexpr size_t imageSideSize = 32;
- static constexpr size_t imgAccElemSize = 4; // rgba
- static constexpr auto imageSize =
- (imgAccElemSize * imageSideSize * imageSideSize);
- float imageData[imageSize] = {0.0f};
-
- auto queue = util::get_cts_object::queue(ctsSelector);
- auto context = queue.get_context();
- auto device = queue.get_device();
-
- sycl::image<2> image(
- imageData, sycl::image_channel_order::rgba,
- sycl::image_channel_type::fp32,
- sycl::range<2>(imageSideSize, imageSideSize));
-
- cl_program clProgram{};
- if (online_compiler_supported(device.get(), log)) {
- std::string kernelSource = R"(
- struct simple_struct {
- int a;
- float b;
- };
-
- __kernel void opencl_interop_image_kernel_kernel(read_only image2d_t arg0,
- sampler_t arg1)
- {}
- )";
-
- if (!create_built_program(kernelSource, context.get(), device.get(),
- clProgram, log)) {
- FAIL(log, "create_built_program failed");
- }
- } else {
- std::string programBinaryFile =
- "opencl_interop_image_kernel.bin";
-
- if (!create_program_with_binary(programBinaryFile, context.get(),
- device.get(), clProgram, log)) {
- std::string errorMsg =
- "create_program_with_binary failed.";
- errorMsg +=
- " Since online compile is not supported, expecting to find " +
- programBinaryFile + " in same path as the executable binary";
- FAIL(log, errorMsg.c_str());
- }
- }
-
- cl_kernel clKernel{};
- if (!create_kernel(clProgram, "opencl_interop_image_kernel_kernel",
- clKernel, log)) {
- FAIL(log, "create_kernel failed");
- }
-
- sycl::kernel kernel(clKernel, context);
-
- /** test single_task(kernel)
- */
- queue.submit([&](sycl::handler &handler) {
- auto imageAccessor =
- image
- .get_access(
- handler);
-
- sycl::sampler sampler(
- sycl::coordinate_normalization_mode::unnormalized,
- sycl::addressing_mode::none,
- sycl::filtering_mode::nearest);
-
- /** check the set_arg() methods
- */
-
- // set_args(int, image)
- handler.set_arg(0, imageAccessor);
- // set_args(int, sampler)
- handler.set_arg(1, sampler);
-
- handler.single_task(kernel);
- });
-
- /** test parallel_for(const nd range&, kernel)
- */
- queue.submit([&](sycl::handler &handler) {
- auto imageAccessor =
- image
- .get_access(
- handler);
-
- sycl::sampler sampler(
- sycl::coordinate_normalization_mode::unnormalized,
- sycl::addressing_mode::none,
- sycl::filtering_mode::nearest);
-
- /** check the set_args() method
- */
- handler.set_args(imageAccessor, sampler);
-
- sycl::range<1> myRange(1024);
- handler.parallel_for(myRange, kernel);
- });
-
- queue.wait_and_throw();
- }
- }
-
- auto ctsQueue = util::get_cts_object::queue(ctsSelector);
- auto context = ctsQueue.get_context();
- auto deviceList = context.get_devices();
-
- // Do ALL devices support online compiler / linker?
- bool compiler_available = is_compiler_available(deviceList);
- bool linker_available = is_linker_available(deviceList);
-
- const std::string compileOptions = "-cl-opt-disable";
- const std::string linkOptions = "-cl-fast-relaxed-math";
-
- {
- log.note(
- "link an OpenCL and a SYCL program without compile and link "
- "options");
-
- if (!compiler_available) {
- log.note("online compiler not available -- skipping check");
- }
-
- else {
- // obtain an existing OpenCL C program object
- cl_program myClProgram = nullptr;
- if (!create_compiled_program(
- kernel_source, context.get(),
- ctsQueue.get_device().get(), myClProgram, log)) {
- FAIL(log, "Didn't create the cl_program");
- }
-
- // Create a SYCL program object from a cl_program object
- sycl::program myExternProgram(context,
- myClProgram);
-
- if (myExternProgram.get_state() !=
- sycl::program_state::compiled) {
- FAIL(log, "Compiled interop program should be in compiled state");
- }
-
- // Add in the SYCL program object for our kernel
- sycl::program mySyclProgram(context);
- mySyclProgram.compile_with_kernel_type>();
-
- if (mySyclProgram.get_state() != sycl::program_state::compiled) {
- FAIL(log, "Compiled SYCL program should be in compiled state");
- }
-
- // Link myClProgram with the SYCL program object
- try {
- sycl::program myLinkedProgram({myExternProgram, mySyclProgram});
-
- if (myLinkedProgram.get_state() !=
- sycl::program_state::linked) {
- FAIL(log, "Program was not linked");
- }
-
- ctsQueue.submit([&](sycl::handler &cgh) {
- cgh.single_task(program_kernel_interop<0>());
- });
- ctsQueue.wait_and_throw();
-
- } catch (const sycl::feature_not_supported &fnse_link) {
- if (!linker_available) {
- log.note("online linker not available -- skipping check");
- } else {
- throw;
- }
- }
- }
- }
-
- {
- log.note(
- "link an OpenCL and a SYCL program with compile and link options");
-
- if (!compiler_available) {
- log.note("online compiler not available -- skipping check");
- }
-
- else {
- // obtain an existing OpenCL C program object
- cl_program myClProgram = nullptr;
- if (!create_compiled_program(
- kernel_source, context.get(),
- ctsQueue.get_device().get(), myClProgram, log)) {
- FAIL(log, "Didn't create the cl_program");
- }
-
- // Create a SYCL program object from a cl_program object
- sycl::program myExternProgram(context,
- myClProgram);
-
- if (myExternProgram.get_state() !=
- sycl::program_state::compiled) {
- FAIL(log, "Compiled interop program should be in compiled state");
- }
-
- // Add in the SYCL program object for our kernel
- sycl::program mySyclProgram(context);
- mySyclProgram.compile_with_kernel_type>(
- compileOptions);
-
- if (mySyclProgram.get_state() != sycl::program_state::compiled) {
- FAIL(log, "Compiled SYCL program should be in compiled state");
- }
-
- if (mySyclProgram.get_compile_options() != compileOptions) {
- FAIL(log,
- "Compiled SYCL program did not store the compile options");
- }
-
- // Link myClProgram with the SYCL program object
- try {
- sycl::program myLinkedProgram({myExternProgram, mySyclProgram},
- linkOptions);
-
- if (myLinkedProgram.get_state() !=
- sycl::program_state::linked) {
- FAIL(log, "Program was not linked");
- }
-
- if (myLinkedProgram.get_link_options() != linkOptions) {
- FAIL(log, "Linked program did not store the link options");
- }
-
- ctsQueue.submit([&](sycl::handler &cgh) {
- cgh.single_task(program_kernel_interop<1>());
- });
- ctsQueue.wait_and_throw();
-
- } catch (const sycl::feature_not_supported &fnse_link) {
- if (!linker_available) {
- log.note("online linker not available -- skipping check");
- } else {
- throw;
- }
- }
- }
- }
-
- if (!context.is_host()) {
- log.note("check compiling and building from source");
-
- { // Check compile_with_source(source)
- sycl::program prog(context);
- try {
- prog.compile_with_source(kernel_source);
- } catch (const sycl::feature_not_supported &fnse_compile) {
- if (!compiler_available) {
- log.note("online compiler not available -- skipping check");
- } else {
- throw;
- }
- }
- }
- { // Check compile_with_source(source, options)
- sycl::program prog(context);
- try {
- prog.compile_with_source(kernel_source, compileOptions);
- } catch (const sycl::feature_not_supported &fnse_compile) {
- if (!compiler_available) {
- log.note("online compiler not available -- skipping check");
- } else {
- throw;
- }
- }
- }
- { // Check build_with_source(source)
- sycl::program prog(context);
- try {
- prog.build_with_source(kernel_source);
- } catch (const sycl::feature_not_supported &fnse_build) {
- if (!compiler_available || !linker_available) {
- log.note(
- "online compiler or linker not available -- skipping check");
- } else {
- throw;
- }
- }
- }
- { // Check build_with_source(source, options)
- sycl::program prog(context);
-
- try {
- prog.build_with_source(kernel_source, linkOptions);
- } catch (const sycl::feature_not_supported &fnse_build) {
- if (!compiler_available || !linker_available) {
- log.note(
- "online compiler or linker not available -- skipping check");
- } else {
- throw;
- }
- }
- }
-
- { // Check retrieving kernel
- sycl::program prog(context);
-
- try {
- prog.build_with_source(kernel_source);
-
- // Check has_kernel(string_class)
- bool hasKernel = prog.has_kernel(kernelName);
- if (!hasKernel) {
- FAIL(log,
- "Program was not built properly (has_kernel(string_class))");
- }
-
- // Check get_kernel(string_class)
- sycl::kernel k = prog.get_kernel(kernelName);
-
- } catch (const sycl::feature_not_supported &fnse_build) {
- if (!compiler_available || !linker_available) {
- log.note(
- "online compiler or linker not available -- skipping check");
- } else {
- throw;
- }
- }
- }
- }
+ // TODO: add checks to sampled_image_accessor, unsampled_image_accessor
} catch (const sycl::exception &e) {
log_exception(log, e);
@@ -522,6 +168,9 @@ class TEST_NAME : public sycl_cts::util::test_base_opencl {
"a SYCL exception was caught: " + std::string(e.what());
FAIL(log, errorMsg.c_str());
}
+#else
+ log.note("The test is skipped because OpenCL back-end is not supported");
+#endif // SYCL_BACKEND_OPENCL
}
};
diff --git a/tests/specialization_constants/specialization_constants_common.h b/tests/specialization_constants/specialization_constants_common.h
index da839c599..45ca00e07 100644
--- a/tests/specialization_constants/specialization_constants_common.h
+++ b/tests/specialization_constants/specialization_constants_common.h
@@ -117,6 +117,11 @@ inline constexpr auto get_init_value_helper(int x) {
return x;
}
+template <>
+inline constexpr auto get_init_value_helper(int x) {
+ return (x%2 != 0);
+}
+
template <>
inline constexpr auto get_init_value_helper(int x) {
testing_types::no_cnstr instance{};
diff --git a/tests/specialization_constants/specialization_constants_exceptions_throwing_common.h b/tests/specialization_constants/specialization_constants_exceptions_throwing_common.h
new file mode 100644
index 000000000..4001b821b
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_exceptions_throwing_common.h
@@ -0,0 +1,99 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Common code for expected exceptions throwing by specialization constants.
+// In this tests we check that exception with code sycl::errc::invalid is thrown
+// any other exception causes test to fail and gets logged.
+//
+*******************************************************************************/
+
+#ifndef __SYCLCTS_TESTS_SPEC_CONST_THROW_EXCEPT_COMMON_H
+#define __SYCLCTS_TESTS_SPEC_CONST_THROW_EXCEPT_COMMON_H
+
+#include "specialization_constants_common.h"
+
+template
+class dummy_specialization_constants_exceptions {};
+template
+using spec_const_exception_dummy_functor =
+ ::dummy_functor>;
+
+template
+class check_spec_constant_exception_throw_for_type {
+ public:
+ void operator()(sycl_cts::util::logger &log, const std::string &type_name) {
+ using namespace get_spec_const;
+ const std::string err_message_prefix{
+ "unexpected SYCL exception was thrown in case "};
+
+ // case 1: Try to get specialization constant via handler that is bound to a
+ // kernel_bundle
+ {
+ bool exception_was_thrown = false;
+ T res = T(get_init_value_helper(0));
+ const int case_num = 1;
+ auto queue = sycl_cts::util::get_cts_object::queue();
+
+ queue.submit([&](sycl::handler &cgh) {
+ auto context = queue.get_context();
+ auto k_bundle =
+ sycl::get_kernel_bundle(context);
+ cgh.use_kernel_bundle(k_bundle);
+ // We expect that exception with sycl::errc::invalid will be thrown
+ try {
+ res = cgh.get_specialization_constant>();
+ } catch (const sycl::exception &e) {
+ if (e.code() != sycl::errc::invalid) {
+ const auto errorMsg =
+ "unexpected SYCL exception was thrown in case " +
+ std::to_string(case_num) + " for " +
+ type_name_string::get(type_name);
+ FAIL(log, errorMsg);
+ throw;
+ } else {
+ exception_was_thrown = true;
+ }
+ }
+ cgh.single_task(spec_const_exception_dummy_functor{});
+ });
+ CHECK_VALUE_SCALAR(log, exception_was_thrown, true);
+ }
+
+ // case 2: Try to set specialization constant via handler that is bound to a
+ // kernel_bundle
+ {
+ bool exception_was_thrown = false;
+ T sc_val = T(get_init_value_helper(0));
+ const int case_num = 2;
+ auto queue = sycl_cts::util::get_cts_object::queue();
+
+ queue.submit([&](sycl::handler &cgh) {
+ auto context = queue.get_context();
+ auto k_bundle =
+ sycl::get_kernel_bundle(context);
+ cgh.use_kernel_bundle(k_bundle);
+ // We expect that exception with sycl::errc::invalid will be thrown
+ try {
+ cgh.set_specialization_constant>(sc_val);
+ } catch (const sycl::exception &e) {
+ if (static_cast(e.code().value()) !=
+ sycl::errc::invalid) {
+ const auto errorMsg =
+ "unexpected SYCL exception was thrown in case " +
+ std::to_string(case_num) + " for " +
+ type_name_string::get(type_name);
+ FAIL(log, errorMsg);
+ throw;
+ } else {
+ exception_was_thrown = true;
+ }
+ }
+ cgh.single_task(spec_const_exception_dummy_functor{});
+ });
+ CHECK_VALUE_SCALAR(log, exception_was_thrown, true);
+ }
+ }
+};
+
+#endif // __SYCLCTS_TESTS_SPEC_CONST_THROW_EXCEPT_COMMON_H
diff --git a/tests/specialization_constants/specialization_constants_exceptions_throwing_core.cpp b/tests/specialization_constants/specialization_constants_exceptions_throwing_core.cpp
new file mode 100644
index 000000000..42b6c665d
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_exceptions_throwing_core.cpp
@@ -0,0 +1,57 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Provides tests for specialization constants throwing exceptions when expected
+//
+*******************************************************************************/
+
+#include "../common/common.h"
+
+#include "specialization_constants_exceptions_throwing_common.h"
+
+#define TEST_NAME specialization_constants_exceptions_throwing_core
+
+namespace TEST_NAMESPACE {
+using namespace sycl_cts;
+
+/** test that specialization constants throws exceptions when expected
+ */
+class TEST_NAME : public util::test_base {
+ public:
+ /** return information about this test
+ */
+ void get_info(test_base::info &out) const override {
+ set_test_info(out, TOSTRING(TEST_NAME), TEST_FILE);
+ }
+
+ /** execute the test
+ */
+ void run(util::logger &log) override {
+ try {
+#ifndef SYCL_CTS_FULL_CONFORMANCE
+ for_all_types(
+ get_spec_const::testing_types::types, log);
+#else
+ for_all_types_vectors_marray(
+ get_spec_const::testing_types::types, log);
+#endif
+ for_all_types(
+ get_spec_const::testing_types::composite_types, log);
+ } catch (const sycl::exception &e) {
+ log_exception(log, e);
+ std::string errorMsg =
+ "a SYCL exception was thrown: " + std::string(e.what());
+ FAIL(log, errorMsg);
+ } catch (const std::exception &e) {
+ std::string errorMsg =
+ "an exception was thrown: " + std::string(e.what());
+ FAIL(log, errorMsg);
+ }
+ }
+};
+
+// construction of this proxy will register the test above
+util::test_proxy proxy;
+
+} // namespace TEST_NAMESPACE
diff --git a/tests/specialization_constants/specialization_constants_exceptions_throwing_fp16.cpp b/tests/specialization_constants/specialization_constants_exceptions_throwing_fp16.cpp
new file mode 100644
index 000000000..e176415aa
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_exceptions_throwing_fp16.cpp
@@ -0,0 +1,65 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Provides tests for specialization constants throwing exceptions for sycl::half
+// when expected
+//
+*******************************************************************************/
+
+#include "../common/common.h"
+
+#include "specialization_constants_exceptions_throwing_common.h"
+
+#define TEST_NAME specialization_constants_exceptions_throwing_fp16
+
+namespace TEST_NAMESPACE {
+using namespace sycl_cts;
+
+/** test that specialization constants throws exceptions with sycl::half
+ when expected
+ */
+class TEST_NAME : public util::test_base {
+ public:
+ /** return information about this test
+ */
+ void get_info(test_base::info &out) const override {
+ set_test_info(out, TOSTRING(TEST_NAME), TEST_FILE);
+ }
+
+ /** execute the test
+ */
+ void run(util::logger &log) override {
+ try {
+ auto queue = util::get_cts_object::queue();
+ if (!queue.get_device().has(sycl::aspect::fp16)) {
+ log.note(
+ "Device does not support half precision floating point "
+ "operations");
+ return;
+ }
+#ifndef SYCL_CTS_FULL_CONFORMANCE
+ check_spec_constant_exception_throw_for_type fp16_test{};
+ fp16_test(log, "sycl::half");
+#else
+ for_type_vectors_marray(log, "sycl::half");
+#endif
+
+ } catch (const sycl::exception &e) {
+ log_exception(log, e);
+ std::string errorMsg =
+ "a SYCL exception was thrown: " + std::string(e.what());
+ FAIL(log, errorMsg);
+ } catch (const std::exception &e) {
+ std::string errorMsg =
+ "an exception was thrown: " + std::string(e.what());
+ FAIL(log, errorMsg);
+ }
+ }
+};
+
+// construction of this proxy will register the test above
+util::test_proxy proxy;
+
+} // namespace TEST_NAMESPACE
diff --git a/tests/specialization_constants/specialization_constants_exceptions_throwing_fp64.cpp b/tests/specialization_constants/specialization_constants_exceptions_throwing_fp64.cpp
new file mode 100644
index 000000000..2d1ad0216
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_exceptions_throwing_fp64.cpp
@@ -0,0 +1,65 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Provides tests for specialization constants throwing exceptions for double
+// when expected
+//
+*******************************************************************************/
+
+#include "../common/common.h"
+
+#include "specialization_constants_exceptions_throwing_common.h"
+
+#define TEST_NAME specialization_constants_exceptions_throwing_fp64
+
+namespace TEST_NAMESPACE {
+using namespace sycl_cts;
+
+/** test that specialization constants throws exceptions with double
+ when expected
+ */
+class TEST_NAME : public util::test_base {
+ public:
+ /** return information about this test
+ */
+ void get_info(test_base::info &out) const override {
+ set_test_info(out, TOSTRING(TEST_NAME), TEST_FILE);
+ }
+
+ /** execute the test
+ */
+ void run(util::logger &log) override {
+ try {
+ auto queue = util::get_cts_object::queue();
+ if (!queue.get_device().has(sycl::aspect::fp64)) {
+ log.note(
+ "Device does not support double precision floating point "
+ "operations");
+ return;
+ }
+#ifndef SYCL_CTS_FULL_CONFORMANCE
+ check_spec_constant_exception_throw_for_type fp64_test{};
+ fp64_test(log, "double");
+#else
+ for_type_vectors_marray(log, "double");
+#endif
+
+ } catch (const sycl::exception &e) {
+ log_exception(log, e);
+ std::string errorMsg =
+ "a SYCL exception was thrown: " + std::string(e.what());
+ FAIL(log, errorMsg);
+ } catch (const std::exception &e) {
+ std::string errorMsg =
+ "an exception was thrown: " + std::string(e.what());
+ FAIL(log, errorMsg);
+ }
+ }
+};
+
+// construction of this proxy will register the test above
+util::test_proxy proxy;
+
+} // namespace TEST_NAMESPACE
diff --git a/tests/specialization_constants/specialization_constants_same_command_group_common.h b/tests/specialization_constants/specialization_constants_same_command_group_common.h
new file mode 100644
index 000000000..dd34d5973
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_same_command_group_common.h
@@ -0,0 +1,96 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Common checks for specialization constants usage via handler
+//
+*******************************************************************************/
+
+#ifndef __SYCLCTS_TESTS_SPEC_CONST_SAME_COMMAND_GROUP_COMMON_H
+#define __SYCLCTS_TESTS_SPEC_CONST_SAME_COMMAND_GROUP_COMMON_H
+
+#include "../common/common.h"
+#include "specialization_constants_common.h"
+
+namespace specialization_constants_same_command_group_common {
+using namespace sycl_cts;
+using namespace get_spec_const;
+
+template class kernel;
+
+template class command_group_object {
+ T *value; // to not initialize for struct with no default constructor
+ sycl::buffer *result_buffer;
+
+public:
+ bool set_const;
+ void set_value(T *value_) {
+ value = value_;
+ set_const = true;
+ }
+ void set_buffer(sycl::buffer *buffer) { result_buffer = buffer; }
+ void operator()(sycl::handler &cgh) {
+ if (set_const)
+ cgh.set_specialization_constant>(*value);
+ auto res_acc =
+ result_buffer->template get_access(cgh);
+ cgh.single_task>([=](sycl::kernel_handler h) {
+ res_acc[0] = h.get_specialization_constant>();
+ });
+ }
+};
+
+template class check_specialization_constants_same_command_group {
+public:
+ void operator()(util::logger &log, const std::string &type_name) {
+ T ref_A { get_init_value_helper(5) };
+ T ref_B { get_init_value_helper(10) };
+ auto queue = util::get_cts_object::queue();
+ sycl::range<1> range(1);
+ {
+ T result1 { get_init_value_helper(0) };
+ T result2 = { get_init_value_helper(0) };
+ {
+ command_group_object cmo;
+ sycl::buffer result_buffer1(&result1, range);
+ sycl::buffer result_buffer2(&result2, range);
+
+ cmo.set_value(&ref_A);
+ cmo.set_buffer(&result_buffer1);
+ queue.submit(cmo);
+
+ cmo.set_value(&ref_B);
+ cmo.set_buffer(&result_buffer2);
+ queue.submit(cmo);
+ }
+ if (!check_equal_values(ref_A, result1))
+ FAIL(log, "case 1 failed for value A for " + type_name);
+ if (!check_equal_values(ref_B, result2))
+ FAIL(log, "case 1 failed for value B for " + type_name);
+ }
+
+ {
+ T result1 = { get_init_value_helper(0) };
+ T result2 = { get_init_value_helper(0) };
+ {
+ command_group_object cmo;
+ sycl::buffer result_buffer1(&result1, range);
+ sycl::buffer result_buffer2(&result2, range);
+
+ cmo.set_value(&ref_A);
+ cmo.set_buffer(&result_buffer1);
+ queue.submit(cmo);
+
+ cmo.set_const = false;
+ cmo.set_buffer(&result_buffer2);
+ queue.submit(cmo);
+ }
+ if (!check_equal_values(ref_A, result1))
+ FAIL(log, "case 2 failed for value A for " + type_name);
+ if (!check_equal_values(T(get_init_value_helper(default_val)), result2))
+ FAIL(log, "case 2 failed for default value for " + type_name);
+ }
+ }
+};
+} /* namespace specialization_constants_same_command_group_common */
+#endif // __SYCLCTS_TESTS_SPEC_CONST_SAME_COMMAND_GROUP_COMMON_H
diff --git a/tests/specialization_constants/specialization_constants_same_command_group_core.cpp b/tests/specialization_constants/specialization_constants_same_command_group_core.cpp
new file mode 100644
index 000000000..12c799db7
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_same_command_group_core.cpp
@@ -0,0 +1,63 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Provides tests for specialization constants usage with same command group
+// function
+//
+*******************************************************************************/
+
+#include "../common/common.h"
+#include "../common/type_coverage.h"
+#include "specialization_constants_same_command_group_common.h"
+
+#define TEST_NAME specialization_constants_same_command_group_core
+
+namespace TEST_NAMESPACE {
+using namespace sycl_cts;
+
+/** test specialization constants
+ */
+class TEST_NAME : public sycl_cts::util::test_base {
+public:
+ /** return information about this test
+ */
+ void get_info(test_base::info &out) const override {
+ set_test_info(out, TOSTRING(TEST_NAME), TEST_FILE);
+ }
+
+ /** execute the test
+ */
+ void run(util::logger &log) override {
+ using namespace specialization_constants_same_command_group_common;
+ try {
+
+#ifndef SYCL_CTS_FULL_CONFORMANCE
+ for_all_types<
+ check_specialization_constants_same_command_group>(
+ get_spec_const::testing_types::types, log);
+#else
+ for_all_types_vectors_marray<
+ check_specialization_constants_same_command_group>(
+ get_spec_const::testing_types::types, log);
+#endif
+ for_all_types(
+ get_spec_const::testing_types::composite_types, log);
+
+ } catch (const sycl::exception &e) {
+ log_exception(log, e);
+ std::string errorMsg =
+ "a SYCL exception was caught: " + std::string(e.what());
+ FAIL(log, errorMsg.c_str());
+ } catch (const std::exception &e) {
+ std::string errorMsg =
+ "an exception was caught: " + std::string(e.what());
+ FAIL(log, errorMsg.c_str());
+ }
+ }
+};
+
+// construction of this proxy will register the above test
+util::test_proxy proxy;
+
+} /* namespace spec_const__ */
diff --git a/tests/specialization_constants/specialization_constants_same_command_group_fp16.cpp b/tests/specialization_constants/specialization_constants_same_command_group_fp16.cpp
new file mode 100644
index 000000000..bcb20dfaa
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_same_command_group_fp16.cpp
@@ -0,0 +1,64 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Provides tests for specialization constants usage with same command group
+// function for sycl::half
+//
+*******************************************************************************/
+
+#include "../common/common.h"
+#include "../common/type_coverage.h"
+#include "specialization_constants_same_command_group_common.h"
+
+#define TEST_NAME specialization_constants_same_command_group_fp16
+
+namespace TEST_NAMESPACE {
+using namespace sycl_cts;
+
+/** test specialization constants for sycl::half
+ */
+class TEST_NAME : public sycl_cts::util::test_base {
+public:
+ /** return information about this test
+ */
+ void get_info(test_base::info &out) const override {
+ set_test_info(out, TOSTRING(TEST_NAME), TEST_FILE);
+ }
+
+ /** execute the test
+ */
+ void run(util::logger &log) override {
+ using namespace specialization_constants_same_command_group_common;
+ try {
+ auto queue = util::get_cts_object::queue();
+ if (!queue.get_device().has(sycl::aspect::fp16)) {
+ log.note("Device does not support half precision floating point "
+ "operations");
+ return;
+ }
+#ifndef SYCL_CTS_FULL_CONFORMANCE
+ check_specialization_constants_same_command_group fp16_test{};
+ fp16_test(log, "sycl::half");
+#else
+ for_type_vectors_marray(log, "sycl::half");
+#endif
+
+ } catch (const sycl::exception &e) {
+ log_exception(log, e);
+ std::string errorMsg =
+ "a SYCL exception was caught: " + std::string(e.what());
+ FAIL(log, errorMsg.c_str());
+ } catch (const std::exception &e) {
+ std::string errorMsg =
+ "an exception was caught: " + std::string(e.what());
+ FAIL(log, errorMsg.c_str());
+ }
+ }
+};
+
+// construction of this proxy will register the above test
+util::test_proxy proxy;
+
+} /* namespace spec_const__ */
diff --git a/tests/specialization_constants/specialization_constants_same_command_group_fp64.cpp b/tests/specialization_constants/specialization_constants_same_command_group_fp64.cpp
new file mode 100644
index 000000000..c672aaeab
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_same_command_group_fp64.cpp
@@ -0,0 +1,64 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Provides tests for specialization constants usage with same command group
+// function for double
+//
+*******************************************************************************/
+
+#include "../common/common.h"
+#include "../common/type_coverage.h"
+#include "specialization_constants_same_command_group_common.h"
+
+#define TEST_NAME specialization_constants_same_command_group_fp64
+
+namespace TEST_NAMESPACE {
+using namespace sycl_cts;
+
+/** test specialization constants for double
+ */
+class TEST_NAME : public sycl_cts::util::test_base {
+public:
+ /** return information about this test
+ */
+ void get_info(test_base::info &out) const override {
+ set_test_info(out, TOSTRING(TEST_NAME), TEST_FILE);
+ }
+
+ /** execute the test
+ */
+ void run(util::logger &log) override {
+ using namespace specialization_constants_same_command_group_common;
+ try {
+ auto queue = util::get_cts_object::queue();
+ if (!queue.get_device().has(sycl::aspect::fp64)) {
+ log.note("Device does not support double precision floating point "
+ "operations");
+ return;
+ }
+#ifndef SYCL_CTS_FULL_CONFORMANCE
+ check_specialization_constants_same_command_group fp64_test{};
+ fp64_test(log, "double");
+#else
+ for_type_vectors_marray(log, "double");
+#endif
+
+ } catch (const sycl::exception &e) {
+ log_exception(log, e);
+ std::string errorMsg =
+ "a SYCL exception was caught: " + std::string(e.what());
+ FAIL(log, errorMsg.c_str());
+ } catch (const std::exception &e) {
+ std::string errorMsg =
+ "an exception was caught: " + std::string(e.what());
+ FAIL(log, errorMsg.c_str());
+ }
+ }
+};
+
+// construction of this proxy will register the above test
+util::test_proxy proxy;
+
+} /* namespace spec_const__ */
diff --git a/tests/specialization_constants/specialization_constants_via_handler_common.h b/tests/specialization_constants/specialization_constants_via_handler_common.h
new file mode 100644
index 000000000..79b15c976
--- /dev/null
+++ b/tests/specialization_constants/specialization_constants_via_handler_common.h
@@ -0,0 +1,237 @@
+/*******************************************************************************
+//
+// SYCL 2020 Conformance Test Suite
+//
+// Common checks for specialization constants usage via handler
+//
+*******************************************************************************/
+
+#ifndef __SYCLCTS_TESTS_SPEC_CONST_HANDLER_COMMON_H
+#define __SYCLCTS_TESTS_SPEC_CONST_HANDLER_COMMON_H
+
+#include "../../util/allocation.h"
+#include "../../util/math_helper.h"
+#include "../common/common.h"
+#include "specialization_constants_common.h"
+
+namespace specialization_constants_via_handler_common {
+using namespace sycl_cts;
+using namespace get_spec_const;
+
+inline constexpr int val_A = 5;
+
+template
+constexpr sycl::specialization_id sc_multiple(get_init_value_helper(case_num));
+
+template class kernel;
+
+template
+bool check_kernel_handler_by_reference(sycl::kernel_handler &h) {
+ T ref { get_init_value_helper(0) };
+ fill_init_values(ref, val_A);
+ return check_equal_values(
+ ref, h.get_specialization_constant>());
+}
+
+template
+bool check_kernel_handler_by_value(sycl::kernel_handler h) {
+ T ref { get_init_value_helper(0) };
+ fill_init_values(ref, val_A);
+ return check_equal_values(
+ ref, h.get_specialization_constant>());
+}
+
+template class check_spec_constant_with_handler_for_type {
+public:
+ void operator()(util::logger &log, const std::string &type_name) {
+ auto queue = util::get_cts_object::queue();
+ sycl::range<1> range(1);
+ T result { get_init_value_helper(0) };
+ T ref { get_init_value_helper(0) };
+ T ref_other { get_init_value_helper(0) };
+ int val_B = 10;
+ fill_init_values(ref, val_A);
+ fill_init_values(ref_other, val_B);
+ // case 1: Set the value in the handler and read it from the same handler.
+ {
+ const int case_num = 1;
+ {
+ result = get_init_value_helper(0);
+ queue.submit([&](sycl::handler &cgh) {
+ cgh.set_specialization_constant>(ref);
+ result = cgh.get_specialization_constant>();
+ });
+ }
+ if (!check_equal_values(ref, result))
+ FAIL(log, "case " + std::to_string(case_num) + " for " +
+ type_name_string::get(type_name));
+ }
+ // case 2: Set the value in the handler twice and read it from the same
+ // handler.
+ {
+ const int case_num = 2;
+ {
+ result = get_init_value_helper(0);
+ queue.submit([&](sycl::handler &cgh) {
+ cgh.set_specialization_constant>(ref);
+ cgh.set_specialization_constant>(ref_other);
+ result = cgh.get_specialization_constant>();
+ });
+ }
+ if (!check_equal_values(ref_other, result))
+ FAIL(log, "case " + std::to_string(case_num) + " for " +
+ type_name_string::get(type_name));
+ }
+
+ // case 3: Set the value in the handler, launch a kernel, and read the value
+ // from the kernel.
+ {
+ const int case_num = 3;
+ {
+ result = get_init_value_helper(0);
+ sycl::buffer result_buffer(&result, range);
+ queue.submit([&](sycl::handler &cgh) {
+ auto res_acc =
+ result_buffer.template get_access(cgh);
+ cgh.set_specialization_constant>(ref);
+ cgh.single_task>([=](sycl::kernel_handler h) {
+ res_acc[0] =
+ h.get_specialization_constant>();
+ });
+ });
+ }
+ if (!check_equal_values(ref, result))
+ FAIL(log, "case " + std::to_string(case_num) + " for " +
+ type_name_string::get(type_name));
+ }
+
+ // case 4: Set the value in the handler twice, launch a kernel, and read the
+ // value from the kernel.
+ {
+ const int case_num = 4;
+ {
+ result = get_init_value_helper(0);
+ sycl::buffer result_buffer(&result, range);
+ queue.submit([&](sycl::handler &cgh) {
+ auto res_acc =
+ result_buffer.template get_access(cgh);
+ cgh.set_specialization_constant>(ref);
+ cgh.set_specialization_constant>(ref_other);
+ cgh.single_task>([=](sycl::kernel_handler h) {
+ res_acc[0] =
+ h.get_specialization_constant>();
+ });
+ });
+ }
+ if (!check_equal_values(ref_other, result))
+ FAIL(log, "case " + std::to_string(case_num) + " for " +
+ type_name_string::get(type_name));
+ }
+
+ // case 5: Set the value in the handler, launch a kernel,
+ // and read the value from the kernel twice.
+ {
+ const int case_num = 5;
+ constexpr size_t size = 2;
+ // to not initialize for struct with no default constructor
+ util::remove_initialization result_vec_same[size]{};
+ {
+ sycl::buffer result_buffer(result_vec_same->data(),
+ sycl::range<1>(size));
+ queue.submit([&](sycl::handler &cgh) {
+ auto res_acc =
+ result_buffer.template get_access(cgh);
+ cgh.set_specialization_constant>(ref);
+ cgh.single_task>([=](sycl::kernel_handler h) {
+ res_acc[0] =
+ h.get_specialization_constant>();
+ res_acc[1] =
+ h.get_specialization_constant>();
+ });
+ });
+ }
+ if (!check_equal_values(ref, result_vec_same[0].value) ||
+ !check_equal_values(ref, result_vec_same[1].value))
+ FAIL(log, "case " + std::to_string(case_num) + " for " +
+ type_name_string::get(type_name));
+ }
+
+ // case 6: Do not set the value of the spec constant, and read it from the
+ // handler. Expecting default value.
+ {
+ const int case_num = 6;
+ {
+ result = get_init_value_helper(0);
+ queue.submit([&](sycl::handler &cgh) {
+ result = cgh.get_specialization_constant>();
+ });
+ }
+ if (!check_equal_values(T(get_init_value_helper(default_val)), result))
+ FAIL(log, "case " + std::to_string(case_num) + " for " +
+ type_name_string::get(type_name));
+ }
+
+ // case 7: Do not set the value of the spec constant, launch a kernel, and
+ // read the value from the kernel. Expecting default value.
+ {
+ const int case_num = 7;
+ {
+ result = get_init_value_helper(0);
+ sycl::buffer result_buffer(&result, range);
+ queue.submit([&](sycl::handler &cgh) {
+ auto res_acc =
+ result_buffer.template get_access(cgh);
+ cgh.single_task>([=](sycl::kernel_handler h) {
+ res_acc[0] =
+ h.get_specialization_constant>();
+ });
+ });
+ }
+ if (!check_equal_values(T(get_init_value_helper(default_val)), result))
+ FAIL(log, "case " + std::to_string(case_num) + " for " +
+ type_name_string::get(type_name));
+ }
+
+ // case 8: Pass kernel handler object by reference to another function
+ {
+ const int case_num = 8;
+ bool func_result = false;
+ {
+ sycl::buffer result_buffer(&func_result, range);
+ queue.submit([&](sycl::handler &cgh) {
+ auto res_acc =
+ result_buffer.template get_access(cgh);
+ cgh.set_specialization_constant>(ref);
+ cgh.single_task>([=](sycl::kernel_handler h) {
+ res_acc[0] = check_kernel_handler_by_reference