diff --git a/tests/handler/handler_copy_common.h b/tests/handler/handler_copy_common.h index a57e777f7..ca735aa95 100644 --- a/tests/handler/handler_copy_common.h +++ b/tests/handler/handler_copy_common.h @@ -26,6 +26,7 @@ #include #include +#include "../../util/sycl_exceptions.h" #include "../common/common.h" namespace handler_copy_common { @@ -296,6 +297,27 @@ using range_helper = range_id_helper; template using id_helper = range_id_helper; +template +sycl::range<3> default_large_range() { + return range_helper<3>::cast(range_helper::make(5, 7, 9)); +} + +template +sycl::range<3> transform_large_range_into_small(sycl::range<3> largeBufRange) { + sycl::range<3> smallBufRange = sycl::range<3>(1, 1, 1); + + // Condense large range into small range so that both + // have the same size (= same number of items). + for (int d = 0; d < dim_large; ++d) { + if (transposed_copy) { + smallBufRange[std::min(d, dim_small - 1)] *= + largeBufRange[dim_large - d - 1]; + } else { + smallBufRange[std::min(d, dim_small - 1)] *= largeBufRange[d]; + } + } + return smallBufRange; +} /** * @brief The copy_test_context encapsulates all host and device data required * for testing, and provides utility functions for verifying the result @@ -622,23 +644,14 @@ class copy_test_context { * transposed_copy is set, in which case the destination will be transposed. */ void setup_ranges() { - constexpr auto dim_large = dim_src > dim_dst ? dim_src : dim_dst; - constexpr auto dim_small = dim_src <= dim_dst ? dim_src : dim_dst; - - auto largeBufRange = - range_helper<3>::cast(range_helper::make(5, 7, 9)); - auto smallBufRange = sycl::range<3>(1, 1, 1); - - // Condense large range into small range so that both - // have the same size (= same number of items). - for (int d = 0; d < dim_large; ++d) { - if (transposed_copy) { - smallBufRange[std::min(d, dim_small - 1)] *= - largeBufRange[dim_large - d - 1]; - } else { - smallBufRange[std::min(d, dim_small - 1)] *= largeBufRange[d]; - } - } + constexpr auto dim_large = std::max(dim_src, dim_dst); + constexpr auto dim_small = std::min(dim_src, dim_dst); + + auto largeBufRange = default_large_range(); + auto smallBufRange = + transform_large_range_into_small( + largeBufRange); + assert(smallBufRange.size() == largeBufRange.size()); auto largeCopyRange = largeBufRange; @@ -818,6 +831,28 @@ static void test_write_acc_copy_functions(log_helper lh, "copy(accessor<$dataT, $dim_src, $mode_src, $target>, " "accessor<$dataT, $dim_dst, $mode_dst, $target>)")); } +#if SYCL_CTS_ENABLE_DEPRECATED_FEATURES_TESTS + { + if constexpr (mode_src == mode_t::read) { + // Check copy(accessor, accessor) with constant_buffer target + copy_test_context ctx( + queue); + ctx.verify_d2d_copy( + [&](sycl::handler& cgh) { + auto r = + ctx.getSrcBuf() + .template get_access( + cgh, ctx.getSrcCopyRange(), ctx.getSrcCopyOffset()); + auto w = ctx.getDstBuf().template get_access( + cgh, ctx.getDstCopyRange(), ctx.getDstCopyOffset()); + cgh.copy(r, w); + }, + lh.set_line(__LINE__).set_op( + "copy(accessor<$dataT, $dim_src, $mode_src, constant_buffer>, " + "accessor<$dataT, $dim_dst, $mode_dst, $target>)")); + }; + } +#endif { // Check fill(accessor, dataT) const auto pattern = type_helper::make(117); @@ -934,5 +969,114 @@ static void test_all_variants(log_helper lh, sycl::queue& queue) { test_all_dimensions(lh, queue); } +// FIXME: re-enable when sycl::errc is implemented in computecpp +#ifndef SYCL_CTS_COMPILING_WITH_COMPUTECPP +/** + * @brief Class provides a test that checks if exception is thrown on explicit + * memory operation copy(acc, acc) in case of destination accessor range less + * than source accessor range + */ +template +class CheckCopyAccToAccException { + static constexpr int dim_src = DimSrcT::value; + static constexpr int dim_dst = DimDstT::value; + static constexpr mode_t mode_src = ModeSrcT::value; + static constexpr mode_t mode_dst = ModeDstT::value; + + sycl::range src_copy_range = range_helper::make(0, 0, 0); + sycl::range dst_copy_range = range_helper::make(0, 0, 0); + + void make_ranges() { + constexpr auto dim_large = std::max(dim_src, dim_dst); + constexpr auto dim_small = std::min(dim_src, dim_dst); + + auto large_range = default_large_range(); + auto small_range = + transform_large_range_into_small(large_range); + + if (dim_src > dim_dst) { + src_copy_range = range_helper::cast(large_range); + // Creating destination range less than source range to force exception on + // explicit memory operation copy(acc, acc) + dst_copy_range = + range_helper::cast(small_range - sycl::range<3>(1, 1, 1)); + } else { + src_copy_range = range_helper::cast(small_range); + // Creating destination range less than source range to force exception on + // explicit memory operation copy(acc, acc) + dst_copy_range = + range_helper::cast(large_range - sycl::range<3>(1, 1, 1)); + } + } + + std::string description(const std::string& type_name, + const std::string& mode_src_name, + const std::string& mode_dst_name, + std::string&& src_target_name) { + std::stringstream ss; + ss << "Check that exception with error code \"errc::invalid\" is thrown on " + "explicit memory operation copy(src_acc, dst_acc) in case of dst_acc " + "with incorrect range size (T: " + << type_name << " dim src: " << dim_src << "dim dist: " << dim_dst + << " acc mode src: " << mode_src_name + << "acc mode dst: " << mode_dst_name + << " source target: " << src_target_name << ")"; + return ss.str(); + } + + public: + CheckCopyAccToAccException() { make_ranges(); } + void operator()(sycl::queue& q, const std::string& type_name, + const std::string&, const std::string&, + const std::string& mode_src_name, + const std::string& mode_dst_name) { + std::shared_ptr src_buf_mem(new DataT[src_copy_range.size()], + std::default_delete()); + std::shared_ptr dst_buf_mem(new DataT[dst_copy_range.size()], + std::default_delete()); + + sycl::buffer src_buf(src_buf_mem, src_copy_range); + sycl::buffer dst_buf(dst_buf_mem, dst_copy_range); + + { + auto check_exception_with_invalid_dst_range = [&] { + q.submit([&](sycl::handler& cgh) { + auto src_acc = + src_buf.template get_access(cgh); + auto dst_acc = + dst_buf.template get_access(cgh); + cgh.copy(src_acc, dst_acc); + }); + }; + INFO(description(type_name, mode_src_name, mode_dst_name, "device")); + CHECK_THROWS_MATCHES( + check_exception_with_invalid_dst_range(), sycl::exception, + sycl_cts::util::equals_exception(sycl::errc::invalid)); + } + +#if SYCL_CTS_ENABLE_DEPRECATED_FEATURES_TESTS + if constexpr (mode_src == mode_t::read) { + auto check_exception_with_invalid_dst_range_constant_buffer = [&] { + q.submit([&](sycl::handler& cgh) { + auto src_acc = + src_buf.template get_access( + cgh); + auto dst_acc = + dst_buf.template get_access(cgh); + cgh.copy(src_acc, dst_acc); + }); + }; + INFO(description(type_name, mode_src_name, mode_dst_name, + "constant_buffer")); + CHECK_THROWS_MATCHES( + check_exception_with_invalid_dst_range_constant_buffer(), + sycl::exception, + sycl_cts::util::equals_exception(sycl::errc::invalid)); + } +#endif // SYCL_CTS_ENABLE_DEPRECATED_FEATURES_TESTS + } +}; +#endif // SYCL_CTS_COMPILING_WITH_COMPUTECPP } // namespace handler_copy_common #endif // __SYCLCTS_TESTS_HANDLER_COPY_COMMON_H diff --git a/tests/handler/handler_copy_core.cpp b/tests/handler/handler_copy_core.cpp index 18290b9fc..9b350d277 100644 --- a/tests/handler/handler_copy_core.cpp +++ b/tests/handler/handler_copy_core.cpp @@ -10,6 +10,12 @@ #include "catch2/catch_test_macros.hpp" +#include "../common/type_coverage.h" + +#include "../common/string_makers.h" + +#include "../common/disabled_for_test_case.h" + namespace handler_copy_core { using namespace handler_copy_common; @@ -34,4 +40,44 @@ TEST_CASE("Tests the API for sycl::handler::copy", "[handler]") { #endif } +// FIXME: re-enable when sycl::errc is implemented in computecpp +DISABLED_FOR_TEST_CASE(ComputeCpp) +("Check exception on copy(accessor, accessor) in case of invalid " + "destination accessor size", + "[handler]")({ + auto queue = util::get_cts_object::queue(); + + const auto types = + named_type_pack::generate("int" +#if SYCL_CTS_ENABLE_FULL_CONFORMANCE + , + "char", "short", "long", "float", + "sycl::char2", "sycl::short3", "sycl::int4", + "sycl::long8", "sycl::float8" +#endif + ); + + const auto dims = value_pack::generate_named( + "one dim range", "two dim range", "three dim range"); + + const auto src_modes = + value_pack::generate_named(); + + const auto dst_modes = + value_pack::generate_named(); + + for_all_combinations(types, dims, dims, src_modes, + dst_modes, queue); +}); + } // namespace handler_copy_core