From cbf6abe9f5df612eb99c3731c637256390344985 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Wed, 1 Feb 2023 03:39:21 -0800 Subject: [PATCH 01/15] Extend support for implicitly device copyable types Signed-off-by: Maronas, Marcos --- sycl/include/sycl/types.hpp | 8 ++++++++ 1 file changed, 8 insertions(+) diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp index 4fc68a0c50073..e9788a1266fce 100644 --- a/sycl/include/sycl/types.hpp +++ b/sycl/include/sycl/types.hpp @@ -2411,6 +2411,10 @@ struct is_device_copyable> : detail::bool_constant::value && is_device_copyable>::value> {}; +template +struct is_device_copyable[N]> + : is_device_copyable> {}; + // marray is device copyable if element type is device copyable and it is also // not trivially copyable (if the element type is trivially copyable, the marray // is device copyable by default). @@ -2433,6 +2437,10 @@ struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") std::is_trivially_destructible::value && !is_device_copyable::value>> : std::true_type {}; +template +struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") + IsDeprecatedDeviceCopyable : IsDeprecatedDeviceCopyable {}; + #ifdef __SYCL_DEVICE_ONLY__ // Checks that the fields of the type T with indices 0 to (NumFieldsToCheck - 1) // are device copyable. From cd4b93d29626ed32ede976a8e7c2a2c3c650d99b Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Wed, 1 Feb 2023 09:09:05 -0800 Subject: [PATCH 02/15] Adds tests for implicit device copyable types. Signed-off-by: Maronas, Marcos --- .../implicit_device_copyable_types.cpp | 161 ++++++++++++++++++ 1 file changed, 161 insertions(+) create mode 100644 sycl/test/basic_tests/implicit_device_copyable_types.cpp diff --git a/sycl/test/basic_tests/implicit_device_copyable_types.cpp b/sycl/test/basic_tests/implicit_device_copyable_types.cpp new file mode 100644 index 0000000000000..3861d8bcb4107 --- /dev/null +++ b/sycl/test/basic_tests/implicit_device_copyable_types.cpp @@ -0,0 +1,161 @@ +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out + +#include +#include + +struct ACopyable { + int i; + ACopyable() = default; + ACopyable(int _i) : i(_i) {} + ACopyable(const ACopyable &x) : i(x.i) {} +}; + +template <> struct sycl::is_device_copyable : std::true_type {}; + +int main() { + sycl::queue q; + { + std::pair pair_arr[5]; + std::pair pair; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::pair p0 = pair_arr[0]; + std::pair p = pair; + }); + }).wait_and_throw(); + } + + { + std::pair pair_arr[5]; + std::pair pair; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::pair p0 = pair_arr[0]; + std::pair p = pair; + }); + }).wait_and_throw(); + } + + { + std::tuple tuple_arr[5]; + std::tuple tuple; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::tuple t0 = tuple_arr[0]; + std::tuple t = tuple; + }); + }).wait_and_throw(); + } + + { + std::tuple tuple_arr[5]; + std::tuple tuple; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::tuple t0 = tuple_arr[0]; + std::tuple t = tuple; + }); + }).wait_and_throw(); + } + + { + std::variant variant_arr[5]; + std::variant variant; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::variant v0 = variant_arr[0]; + std::variant v = variant; + }); + }).wait_and_throw(); + } + + { + std::variant variant_arr[5]; + std::variant variant; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + // std::variant with complex types relies on virtual functions, so + // they cannot be used within sycl kernels + auto size = sizeof(variant_arr[0]); + size = sizeof(variant); + }); + }).wait_and_throw(); + } + + { + std::array arr_arr[5]; + std::array arr; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::array arr0 = arr_arr[0]; + std::array a = arr; + }); + }).wait_and_throw(); + } + + { + std::array arr_arr[5]; + std::array arr; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::array arr0 = arr_arr[0]; + std::array a = arr; + }); + }).wait_and_throw(); + } + + { + sycl::queue q{}; + std::optional opt_arr[5]; + std::optional opt; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::optional o0 = opt_arr[0]; + std::optional o = opt; + }); + }).wait_and_throw(); + } + + { + sycl::queue q{}; + std::optional opt_arr[5]; + std::optional opt; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::optional o0 = opt_arr[0]; + std::optional o = opt; + }); + }).wait_and_throw(); + } + + { + std::string_view strv_arr[5]; + std::string_view strv; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::string_view str0 = strv_arr[0]; + std::string_view str = strv; + }); + }).wait_and_throw(); + } + +#if __cpp_lib_span >= 202002 + { + std::vector v(5); + std::span s{v.begin(), 4}; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { int x = s[0]; }); + }).wait_and_throw(); + } +#endif + + { + std::vector v(5); + sycl::span s{v.data(), 4}; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { int x = s[0]; }); + }).wait_and_throw(); + } + + return 0; +} From 6b0264073863d4d6c72d3371cdb534dc7eecea83 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Fri, 3 Feb 2023 02:31:59 -0800 Subject: [PATCH 03/15] Extend support for SYCL2020 implicitly device copyable types. Signed-off-by: Maronas, Marcos --- sycl/include/sycl/types.hpp | 53 ++++++++++++++++++++++++++++++++++--- 1 file changed, 50 insertions(+), 3 deletions(-) diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp index e9788a1266fce..f53d07ce7cbe1 100644 --- a/sycl/include/sycl/types.hpp +++ b/sycl/include/sycl/types.hpp @@ -58,6 +58,7 @@ #include #include #include +#include #ifndef __SYCL_DEVICE_ONLY__ #include #endif @@ -2401,6 +2402,35 @@ struct is_device_copyable< template inline constexpr bool is_device_copyable_v = is_device_copyable::value; +// std::array is implicitly device copyable type. +template +struct is_device_copyable> : std::true_type {}; + +// std::array is implicitly device copyable type if T is device copyable +// and it is not trivially copyable (if it is trivially copyable it is device +// copyable by default) +template +struct is_device_copyable< + std::array, + std::enable_if_t>::value>> + : is_device_copyable {}; + +// std::optional is implicitly device copyable type if T is device copyable +// and it is not trivially copyable (if it is trivially copyable it is device +// copyable by default) +template +struct is_device_copyable< + std::optional, + std::enable_if_t>::value>> + : is_device_copyable {}; + +// std::pair is implicitly device copyable type if T1 and T2 are device +// copyable +template +struct is_device_copyable> + : detail::bool_constant::value && + is_device_copyable::value> {}; + // std::tuple<> is implicitly device copyable type. template <> struct is_device_copyable> : std::true_type {}; @@ -2411,9 +2441,18 @@ struct is_device_copyable> : detail::bool_constant::value && is_device_copyable>::value> {}; -template -struct is_device_copyable[N]> - : is_device_copyable> {}; +// std::variant<> is implicitly device copyable type +template <> struct is_device_copyable> : std::true_type {}; + +// std::variant is implicitly device copyable type if each type T of +// Ts... is device copyable, and it is not trivially copyable (if it is +// trivially copyable it is device copyable by default) +template +struct is_device_copyable, + std::enable_if_t>::value>> + : detail::bool_constant::value && + is_device_copyable>::value> {}; // marray is device copyable if element type is device copyable and it is also // not trivially copyable (if the element type is trivially copyable, the marray @@ -2424,6 +2463,14 @@ struct is_device_copyable< !std::is_trivially_copyable::value>> : std::true_type {}; +// array is device copyable if element type is device copyable and it is also +// not trivially copyable (if the element type is trivially copyable, the array +// is device copyable by default). +template +struct is_device_copyable< + T[N], std::enable_if_t::value>> + : is_device_copyable {}; + namespace detail { template struct IsDeprecatedDeviceCopyable : std::false_type {}; From 80c8fc68fab4416445b6a7e81e4941d8d04d8112 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Fri, 3 Feb 2023 03:29:12 -0800 Subject: [PATCH 04/15] Fix clang-format issues. Signed-off-by: Maronas, Marcos --- .../implicit_device_copyable_types.cpp | 278 +++++++++--------- 1 file changed, 139 insertions(+), 139 deletions(-) diff --git a/sycl/test/basic_tests/implicit_device_copyable_types.cpp b/sycl/test/basic_tests/implicit_device_copyable_types.cpp index 3861d8bcb4107..9cd6c12c2ebf5 100644 --- a/sycl/test/basic_tests/implicit_device_copyable_types.cpp +++ b/sycl/test/basic_tests/implicit_device_copyable_types.cpp @@ -13,149 +13,149 @@ struct ACopyable { template <> struct sycl::is_device_copyable : std::true_type {}; int main() { - sycl::queue q; - { - std::pair pair_arr[5]; - std::pair pair; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::pair p0 = pair_arr[0]; - std::pair p = pair; - }); - }).wait_and_throw(); - } - - { - std::pair pair_arr[5]; - std::pair pair; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::pair p0 = pair_arr[0]; - std::pair p = pair; - }); - }).wait_and_throw(); - } - - { - std::tuple tuple_arr[5]; - std::tuple tuple; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::tuple t0 = tuple_arr[0]; - std::tuple t = tuple; - }); - }).wait_and_throw(); - } - - { - std::tuple tuple_arr[5]; - std::tuple tuple; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::tuple t0 = tuple_arr[0]; - std::tuple t = tuple; - }); - }).wait_and_throw(); - } - - { - std::variant variant_arr[5]; - std::variant variant; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::variant v0 = variant_arr[0]; - std::variant v = variant; - }); - }).wait_and_throw(); - } - - { - std::variant variant_arr[5]; - std::variant variant; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - // std::variant with complex types relies on virtual functions, so - // they cannot be used within sycl kernels - auto size = sizeof(variant_arr[0]); - size = sizeof(variant); - }); - }).wait_and_throw(); - } - - { - std::array arr_arr[5]; - std::array arr; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::array arr0 = arr_arr[0]; - std::array a = arr; - }); - }).wait_and_throw(); - } - - { - std::array arr_arr[5]; - std::array arr; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::array arr0 = arr_arr[0]; - std::array a = arr; - }); - }).wait_and_throw(); - } - - { - sycl::queue q{}; - std::optional opt_arr[5]; - std::optional opt; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::optional o0 = opt_arr[0]; - std::optional o = opt; - }); - }).wait_and_throw(); - } - - { - sycl::queue q{}; - std::optional opt_arr[5]; - std::optional opt; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::optional o0 = opt_arr[0]; - std::optional o = opt; - }); - }).wait_and_throw(); - } - - { - std::string_view strv_arr[5]; - std::string_view strv; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::string_view str0 = strv_arr[0]; - std::string_view str = strv; - }); - }).wait_and_throw(); - } + sycl::queue q; + { + std::pair pair_arr[5]; + std::pair pair; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::pair p0 = pair_arr[0]; + std::pair p = pair; + }); + }).wait_and_throw(); + } + + { + std::pair pair_arr[5]; + std::pair pair; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::pair p0 = pair_arr[0]; + std::pair p = pair; + }); + }).wait_and_throw(); + } + + { + std::tuple tuple_arr[5]; + std::tuple tuple; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::tuple t0 = tuple_arr[0]; + std::tuple t = tuple; + }); + }).wait_and_throw(); + } + + { + std::tuple tuple_arr[5]; + std::tuple tuple; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::tuple t0 = tuple_arr[0]; + std::tuple t = tuple; + }); + }).wait_and_throw(); + } + + { + std::variant variant_arr[5]; + std::variant variant; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::variant v0 = variant_arr[0]; + std::variant v = variant; + }); + }).wait_and_throw(); + } + + { + std::variant variant_arr[5]; + std::variant variant; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + // std::variant with complex types relies on virtual functions, so + // they cannot be used within sycl kernels + auto size = sizeof(variant_arr[0]); + size = sizeof(variant); + }); + }).wait_and_throw(); + } + + { + std::array arr_arr[5]; + std::array arr; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::array arr0 = arr_arr[0]; + std::array a = arr; + }); + }).wait_and_throw(); + } + + { + std::array arr_arr[5]; + std::array arr; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::array arr0 = arr_arr[0]; + std::array a = arr; + }); + }).wait_and_throw(); + } + + { + sycl::queue q{}; + std::optional opt_arr[5]; + std::optional opt; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::optional o0 = opt_arr[0]; + std::optional o = opt; + }); + }).wait_and_throw(); + } + + { + sycl::queue q{}; + std::optional opt_arr[5]; + std::optional opt; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::optional o0 = opt_arr[0]; + std::optional o = opt; + }); + }).wait_and_throw(); + } + + { + std::string_view strv_arr[5]; + std::string_view strv; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + std::string_view str0 = strv_arr[0]; + std::string_view str = strv; + }); + }).wait_and_throw(); + } #if __cpp_lib_span >= 202002 - { - std::vector v(5); - std::span s{v.begin(), 4}; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { int x = s[0]; }); - }).wait_and_throw(); - } + { + std::vector v(5); + std::span s{v.begin(), 4}; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { int x = s[0]; }); + }).wait_and_throw(); + } #endif - { - std::vector v(5); - sycl::span s{v.data(), 4}; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { int x = s[0]; }); - }).wait_and_throw(); - } + { + std::vector v(5); + sycl::span s{v.data(), 4}; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { int x = s[0]; }); + }).wait_and_throw(); + } return 0; } From e36f4b8f387bf107b9598b9a39c3c82dc567a693 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Fri, 3 Feb 2023 03:43:20 -0800 Subject: [PATCH 05/15] Fix more clang-format issues. Signed-off-by: Maronas, Marcos --- sycl/test/basic_tests/implicit_device_copyable_types.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test/basic_tests/implicit_device_copyable_types.cpp b/sycl/test/basic_tests/implicit_device_copyable_types.cpp index 9cd6c12c2ebf5..f9b71791dd323 100644 --- a/sycl/test/basic_tests/implicit_device_copyable_types.cpp +++ b/sycl/test/basic_tests/implicit_device_copyable_types.cpp @@ -157,5 +157,5 @@ int main() { }).wait_and_throw(); } - return 0; + return 0; } From 7fcac676c773bb8e840e5fe4997f2387c596ee9a Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Wed, 8 Feb 2023 08:59:23 -0800 Subject: [PATCH 06/15] Addressing code review comments. Signed-off-by: Maronas, Marcos --- sycl/include/sycl/types.hpp | 28 ++-- .../implicit_device_copyable_types.cpp | 146 ++---------------- 2 files changed, 30 insertions(+), 144 deletions(-) diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp index f53d07ce7cbe1..fe8e02e6e5363 100644 --- a/sycl/include/sycl/types.hpp +++ b/sycl/include/sycl/types.hpp @@ -2394,6 +2394,12 @@ __SYCL_DECLARE_FLOAT_VECTOR_CONVERTERS(double) template struct is_device_copyable : std::false_type {}; +// NOTE: this specialization is a candidate for all T such that T is trivially +// copyable, including std::array, std::optional, std::variant, +// sycl::marray and T[N]. Thus, specializations for all these mentioned +// types are guarded by `std::enable_if_t>` +// so that they are candidates only for non-trivially-copyable types. +// Otherwise, there are several candidates and the compiler can't decide. template struct is_device_copyable< T, std::enable_if_t::value>> @@ -2407,8 +2413,6 @@ template struct is_device_copyable> : std::true_type {}; // std::array is implicitly device copyable type if T is device copyable -// and it is not trivially copyable (if it is trivially copyable it is device -// copyable by default) template struct is_device_copyable< std::array, @@ -2416,8 +2420,6 @@ struct is_device_copyable< : is_device_copyable {}; // std::optional is implicitly device copyable type if T is device copyable -// and it is not trivially copyable (if it is trivially copyable it is device -// copyable by default) template struct is_device_copyable< std::optional, @@ -2445,14 +2447,12 @@ struct is_device_copyable> template <> struct is_device_copyable> : std::true_type {}; // std::variant is implicitly device copyable type if each type T of -// Ts... is device copyable, and it is not trivially copyable (if it is -// trivially copyable it is device copyable by default) -template -struct is_device_copyable, - std::enable_if_t>::value>> - : detail::bool_constant::value && - is_device_copyable>::value> {}; +// Ts... is device copyable +template +struct is_device_copyable< + std::variant, + std::enable_if_t>::value>> + : is_device_copyable {}; // marray is device copyable if element type is device copyable and it is also // not trivially copyable (if the element type is trivially copyable, the marray @@ -2463,9 +2463,7 @@ struct is_device_copyable< !std::is_trivially_copyable::value>> : std::true_type {}; -// array is device copyable if element type is device copyable and it is also -// not trivially copyable (if the element type is trivially copyable, the array -// is device copyable by default). +// array is device copyable if element type is device copyable template struct is_device_copyable< T[N], std::enable_if_t::value>> diff --git a/sycl/test/basic_tests/implicit_device_copyable_types.cpp b/sycl/test/basic_tests/implicit_device_copyable_types.cpp index f9b71791dd323..2556e91a1d3f3 100644 --- a/sycl/test/basic_tests/implicit_device_copyable_types.cpp +++ b/sycl/test/basic_tests/implicit_device_copyable_types.cpp @@ -13,62 +13,24 @@ struct ACopyable { template <> struct sycl::is_device_copyable : std::true_type {}; int main() { - sycl::queue q; - { - std::pair pair_arr[5]; - std::pair pair; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::pair p0 = pair_arr[0]; - std::pair p = pair; - }); - }).wait_and_throw(); - } - - { - std::pair pair_arr[5]; - std::pair pair; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::pair p0 = pair_arr[0]; - std::pair p = pair; - }); - }).wait_and_throw(); - } - - { - std::tuple tuple_arr[5]; - std::tuple tuple; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::tuple t0 = tuple_arr[0]; - std::tuple t = tuple; - }); - }).wait_and_throw(); - } - - { - std::tuple tuple_arr[5]; - std::tuple tuple; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::tuple t0 = tuple_arr[0]; - std::tuple t = tuple; - }); - }).wait_and_throw(); - } - - { - std::variant variant_arr[5]; - std::variant variant; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::variant v0 = variant_arr[0]; - std::variant v = variant; - }); - }).wait_and_throw(); - } + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v); +#if __cpp_lib_span >= 202002 + static_assert(sycl::is_device_copyable_v>); +#endif + static_assert(sycl::is_device_copyable_v>); +#if COMPILE_ONLY + sycl::queue q; { std::variant variant_arr[5]; std::variant variant; @@ -81,81 +43,7 @@ int main() { }); }).wait_and_throw(); } - - { - std::array arr_arr[5]; - std::array arr; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::array arr0 = arr_arr[0]; - std::array a = arr; - }); - }).wait_and_throw(); - } - - { - std::array arr_arr[5]; - std::array arr; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::array arr0 = arr_arr[0]; - std::array a = arr; - }); - }).wait_and_throw(); - } - - { - sycl::queue q{}; - std::optional opt_arr[5]; - std::optional opt; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::optional o0 = opt_arr[0]; - std::optional o = opt; - }); - }).wait_and_throw(); - } - - { - sycl::queue q{}; - std::optional opt_arr[5]; - std::optional opt; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::optional o0 = opt_arr[0]; - std::optional o = opt; - }); - }).wait_and_throw(); - } - - { - std::string_view strv_arr[5]; - std::string_view strv; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - std::string_view str0 = strv_arr[0]; - std::string_view str = strv; - }); - }).wait_and_throw(); - } - -#if __cpp_lib_span >= 202002 - { - std::vector v(5); - std::span s{v.begin(), 4}; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { int x = s[0]; }); - }).wait_and_throw(); - } #endif - { - std::vector v(5); - sycl::span s{v.data(), 4}; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { int x = s[0]; }); - }).wait_and_throw(); - } - return 0; } From 61c6e1290abbca6c47918a2487afc31c8810733b Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Thu, 9 Feb 2023 05:58:57 -0800 Subject: [PATCH 07/15] Extends is_device_copyable to consider cv-qualifiers. Signed-off-by: Maronas, Marcos --- .../sycl/ext/oneapi/properties/properties.hpp | 7 ---- sycl/include/sycl/types.hpp | 35 +++++++++++-------- .../properties/properties_device_copyable.cpp | 6 ++-- 3 files changed, 24 insertions(+), 24 deletions(-) diff --git a/sycl/include/sycl/ext/oneapi/properties/properties.hpp b/sycl/include/sycl/ext/oneapi/properties/properties.hpp index 587c59a8b28c3..5b83abf6cef5a 100644 --- a/sycl/include/sycl/ext/oneapi/properties/properties.hpp +++ b/sycl/include/sycl/ext/oneapi/properties/properties.hpp @@ -232,12 +232,5 @@ struct is_device_copyable< std::enable_if_t>::value>> : is_device_copyable {}; -template -struct is_device_copyable< - const ext::oneapi::experimental::properties, - std::enable_if_t>::value>> - : is_device_copyable {}; - } // __SYCL_INLINE_VER_NAMESPACE(_V1) } // namespace sycl diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp index fe8e02e6e5363..2c4cb380e6f6c 100644 --- a/sycl/include/sycl/types.hpp +++ b/sycl/include/sycl/types.hpp @@ -2469,22 +2469,29 @@ struct is_device_copyable< T[N], std::enable_if_t::value>> : is_device_copyable {}; +template +struct is_device_copyable< + T, std::enable_if_t::value && + (std::is_const_v || std::is_volatile_v)>> + : is_device_copyable> {}; + namespace detail { -template -struct IsDeprecatedDeviceCopyable : std::false_type {}; + template + struct IsDeprecatedDeviceCopyable : std::false_type {}; -// TODO: using C++ attribute [[deprecated]] or the macro __SYCL2020_DEPRECATED -// does not produce expected warning message for the type 'T'. -template -struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") - IsDeprecatedDeviceCopyable< - T, std::enable_if_t::value && - std::is_trivially_destructible::value && - !is_device_copyable::value>> : std::true_type {}; - -template -struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") - IsDeprecatedDeviceCopyable : IsDeprecatedDeviceCopyable {}; + // TODO: using C++ attribute [[deprecated]] or the macro __SYCL2020_DEPRECATED + // does not produce expected warning message for the type 'T'. + template + struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") + IsDeprecatedDeviceCopyable< + T, std::enable_if_t::value && + std::is_trivially_destructible::value && + !is_device_copyable::value>> : std::true_type { + }; + + template + struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") + IsDeprecatedDeviceCopyable : IsDeprecatedDeviceCopyable {}; #ifdef __SYCL_DEVICE_ONLY__ // Checks that the fields of the type T with indices 0 to (NumFieldsToCheck - 1) diff --git a/sycl/test/extensions/properties/properties_device_copyable.cpp b/sycl/test/extensions/properties/properties_device_copyable.cpp index 7ed16ba681c3d..479ab39568f07 100644 --- a/sycl/test/extensions/properties/properties_device_copyable.cpp +++ b/sycl/test/extensions/properties/properties_device_copyable.cpp @@ -13,9 +13,9 @@ namespace sycl { template <> struct is_device_copyable : std::true_type {}; -template <> -struct is_device_copyable - : std::true_type {}; +// template <> +// struct is_device_copyable +// : std::true_type {}; } // namespace sycl int main() { From 918c61a9d0a0b9a6393a6e76ebec22d8c809fc71 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Thu, 9 Feb 2023 05:59:16 -0800 Subject: [PATCH 08/15] Extends testing with cv-qualifiers. Signed-off-by: Maronas, Marcos --- .../implicit_device_copyable_types.cpp | 105 +++++++++++++++++- 1 file changed, 104 insertions(+), 1 deletion(-) diff --git a/sycl/test/basic_tests/implicit_device_copyable_types.cpp b/sycl/test/basic_tests/implicit_device_copyable_types.cpp index 2556e91a1d3f3..11db6d926370d 100644 --- a/sycl/test/basic_tests/implicit_device_copyable_types.cpp +++ b/sycl/test/basic_tests/implicit_device_copyable_types.cpp @@ -27,7 +27,74 @@ int main() { #if __cpp_lib_span >= 202002 static_assert(sycl::is_device_copyable_v>); #endif - static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + + // const + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v); +#if __cpp_lib_span >= 202002 + static_assert(sycl::is_device_copyable_v>); +#endif + static_assert(sycl::is_device_copyable_v>); + + // volatile + static_assert(sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v); +#if __cpp_lib_span >= 202002 + static_assert(sycl::is_device_copyable_v>); +#endif + static_assert(sycl::is_device_copyable_v>); + + // const volatile + static_assert( + sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v< + const volatile std::tuple>); + static_assert(sycl::is_device_copyable_v< + const volatile std::variant>); + static_assert( + sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v>); + static_assert( + sycl::is_device_copyable_v>); + static_assert(sycl::is_device_copyable_v); +#if __cpp_lib_span >= 202002 + static_assert(sycl::is_device_copyable_v>); +#endif + static_assert(sycl::is_device_copyable_v>); #if COMPILE_ONLY sycl::queue q; @@ -43,6 +110,42 @@ int main() { }); }).wait_and_throw(); } + { + const std::variant variant_arr[5]; + const std::variant variant; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + // std::variant with complex types relies on virtual functions, so + // they cannot be used within sycl kernels + auto size = sizeof(variant_arr[0]); + size = sizeof(variant); + }); + }).wait_and_throw(); + } + { + volatile std::variant variant_arr[5]; + volatile std::variant variant; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + // std::variant with complex types relies on virtual functions, so + // they cannot be used within sycl kernels + auto size = sizeof(variant_arr[0]); + size = sizeof(variant); + }); + }).wait_and_throw(); + } + { + const volatile std::variant variant_arr[5]; + const volatile std::variant variant; + q.submit([&](sycl::handler &cgh) { + cgh.single_task([=]() { + // std::variant with complex types relies on virtual functions, so + // they cannot be used within sycl kernels + auto size = sizeof(variant_arr[0]); + size = sizeof(variant); + }); + }).wait_and_throw(); + } #endif return 0; From 8804479e0973b916249dae01486b15aa12de3cfc Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Thu, 9 Feb 2023 06:07:18 -0800 Subject: [PATCH 09/15] Undoing unnecessary comment. Signed-off-by: Maronas, Marcos --- .../extensions/properties/properties_device_copyable.cpp | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/sycl/test/extensions/properties/properties_device_copyable.cpp b/sycl/test/extensions/properties/properties_device_copyable.cpp index 479ab39568f07..7ed16ba681c3d 100644 --- a/sycl/test/extensions/properties/properties_device_copyable.cpp +++ b/sycl/test/extensions/properties/properties_device_copyable.cpp @@ -13,9 +13,9 @@ namespace sycl { template <> struct is_device_copyable : std::true_type {}; -// template <> -// struct is_device_copyable -// : std::true_type {}; +template <> +struct is_device_copyable + : std::true_type {}; } // namespace sycl int main() { From 290c72ec33f034af0274c3cf232a5bdf2fe0b7ba Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Thu, 9 Feb 2023 06:23:01 -0800 Subject: [PATCH 10/15] Fixes clang-format issues. Signed-off-by: Maronas, Marcos --- sycl/include/sycl/types.hpp | 33 ++++++++++++++++----------------- 1 file changed, 16 insertions(+), 17 deletions(-) diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp index 2c4cb380e6f6c..0a4c4241d09e9 100644 --- a/sycl/include/sycl/types.hpp +++ b/sycl/include/sycl/types.hpp @@ -2476,26 +2476,25 @@ struct is_device_copyable< : is_device_copyable> {}; namespace detail { - template - struct IsDeprecatedDeviceCopyable : std::false_type {}; +template +struct IsDeprecatedDeviceCopyable : std::false_type {}; - // TODO: using C++ attribute [[deprecated]] or the macro __SYCL2020_DEPRECATED - // does not produce expected warning message for the type 'T'. - template - struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") - IsDeprecatedDeviceCopyable< - T, std::enable_if_t::value && - std::is_trivially_destructible::value && - !is_device_copyable::value>> : std::true_type { - }; +// TODO: using C++ attribute [[deprecated]] or the macro __SYCL2020_DEPRECATED +// does not produce expected warning message for the type 'T'. +template +struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") + IsDeprecatedDeviceCopyable< + T, std::enable_if_t::value && + std::is_trivially_destructible::value && + !is_device_copyable::value>> : std::true_type {}; - template - struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") - IsDeprecatedDeviceCopyable : IsDeprecatedDeviceCopyable {}; +template +struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020") + IsDeprecatedDeviceCopyable : IsDeprecatedDeviceCopyable {}; #ifdef __SYCL_DEVICE_ONLY__ -// Checks that the fields of the type T with indices 0 to (NumFieldsToCheck - 1) -// are device copyable. +// Checks that the fields of the type T with indices 0 to (NumFieldsToCheck - +// 1) are device copyable. template struct CheckFieldsAreDeviceCopyable : CheckFieldsAreDeviceCopyable { @@ -2514,7 +2513,7 @@ struct CheckBasesAreDeviceCopyable : CheckBasesAreDeviceCopyable { using BaseT = decltype(__builtin_base_type(T, NumBasesToCheck - 1)); static_assert(is_device_copyable::value || - detail::IsDeprecatedDeviceCopyable::value, + "The specified type is not device copyable"); }; From 10913deefabb35c588698c7d7a4b77672da5eb92 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Thu, 9 Feb 2023 07:46:52 -0800 Subject: [PATCH 11/15] Recovers accidentally removed line. Signed-off-by: Maronas, Marcos --- sycl/include/sycl/types.hpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp index 0a4c4241d09e9..1655a11bd9350 100644 --- a/sycl/include/sycl/types.hpp +++ b/sycl/include/sycl/types.hpp @@ -2513,7 +2513,7 @@ struct CheckBasesAreDeviceCopyable : CheckBasesAreDeviceCopyable { using BaseT = decltype(__builtin_base_type(T, NumBasesToCheck - 1)); static_assert(is_device_copyable::value || - + detail::IsDeprecatedDeviceCopyable::value, "The specified type is not device copyable"); }; From ad1a7106dd17235514d6da6db2c6252a69b6a492 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Fri, 10 Feb 2023 02:18:27 -0800 Subject: [PATCH 12/15] Corrects test behaviour. Signed-off-by: Maronas, Marcos --- .../implicit_device_copyable_types.cpp | 43 ++----------------- 1 file changed, 3 insertions(+), 40 deletions(-) diff --git a/sycl/test/basic_tests/implicit_device_copyable_types.cpp b/sycl/test/basic_tests/implicit_device_copyable_types.cpp index 11db6d926370d..fdd4545abc861 100644 --- a/sycl/test/basic_tests/implicit_device_copyable_types.cpp +++ b/sycl/test/basic_tests/implicit_device_copyable_types.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify -Xclang -fsycl-is-device -DCOMPILE_ONLY %s #include #include @@ -99,52 +99,15 @@ int main() { #if COMPILE_ONLY sycl::queue q; { - std::variant variant_arr[5]; std::variant variant; q.submit([&](sycl::handler &cgh) { cgh.single_task([=]() { // std::variant with complex types relies on virtual functions, so // they cannot be used within sycl kernels - auto size = sizeof(variant_arr[0]); - size = sizeof(variant); - }); - }).wait_and_throw(); - } - { - const std::variant variant_arr[5]; - const std::variant variant; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - // std::variant with complex types relies on virtual functions, so - // they cannot be used within sycl kernels - auto size = sizeof(variant_arr[0]); - size = sizeof(variant); - }); - }).wait_and_throw(); - } - { - volatile std::variant variant_arr[5]; - volatile std::variant variant; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - // std::variant with complex types relies on virtual functions, so - // they cannot be used within sycl kernels - auto size = sizeof(variant_arr[0]); - size = sizeof(variant); - }); - }).wait_and_throw(); - } - { - const volatile std::variant variant_arr[5]; - const volatile std::variant variant; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - // std::variant with complex types relies on virtual functions, so - // they cannot be used within sycl kernels - auto size = sizeof(variant_arr[0]); - size = sizeof(variant); + auto v = variant; }); }).wait_and_throw(); + //expected-error@variant:* {{SYCL kernel cannot call through a function pointer}} } #endif From d8e5c39c01df87632a21ed17368b9eb0bbc30c85 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Fri, 10 Feb 2023 06:20:19 -0800 Subject: [PATCH 13/15] Fixes issues with MSVC. Signed-off-by: Maronas, Marcos --- sycl/include/sycl/types.hpp | 9 +++++++-- 1 file changed, 7 insertions(+), 2 deletions(-) diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp index 1655a11bd9350..80a7cb5699af5 100644 --- a/sycl/include/sycl/types.hpp +++ b/sycl/include/sycl/types.hpp @@ -58,6 +58,7 @@ #include #include #include +#include #include #ifndef __SYCL_DEVICE_ONLY__ #include @@ -2429,7 +2430,9 @@ struct is_device_copyable< // std::pair is implicitly device copyable type if T1 and T2 are device // copyable template -struct is_device_copyable> +struct is_device_copyable< + std::pair, + std::enable_if_t>::value>> : detail::bool_constant::value && is_device_copyable::value> {}; @@ -2439,7 +2442,9 @@ template <> struct is_device_copyable> : std::true_type {}; // std::tuple is implicitly device copyable type if each type T of Ts... // is device copyable. template -struct is_device_copyable> +struct is_device_copyable< + std::tuple, + std::enable_if_t>::value>> : detail::bool_constant::value && is_device_copyable>::value> {}; From 7e612012979ca4e5895c31dfbb9b1e550433ad9b Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Fri, 10 Feb 2023 07:11:39 -0800 Subject: [PATCH 14/15] Removes incorrect test. Signed-off-by: Maronas, Marcos --- .../implicit_device_copyable_types.cpp | 15 --------------- 1 file changed, 15 deletions(-) diff --git a/sycl/test/basic_tests/implicit_device_copyable_types.cpp b/sycl/test/basic_tests/implicit_device_copyable_types.cpp index fdd4545abc861..0bda989649056 100644 --- a/sycl/test/basic_tests/implicit_device_copyable_types.cpp +++ b/sycl/test/basic_tests/implicit_device_copyable_types.cpp @@ -96,20 +96,5 @@ int main() { #endif static_assert(sycl::is_device_copyable_v>); -#if COMPILE_ONLY - sycl::queue q; - { - std::variant variant; - q.submit([&](sycl::handler &cgh) { - cgh.single_task([=]() { - // std::variant with complex types relies on virtual functions, so - // they cannot be used within sycl kernels - auto v = variant; - }); - }).wait_and_throw(); - //expected-error@variant:* {{SYCL kernel cannot call through a function pointer}} - } -#endif - return 0; } From c82955e58d152069053453eec1ad346589109f33 Mon Sep 17 00:00:00 2001 From: "Maronas, Marcos" Date: Mon, 13 Feb 2023 01:19:58 -0800 Subject: [PATCH 15/15] Removes unnecessary flags from test. Signed-off-by: Maronas, Marcos --- sycl/test/basic_tests/implicit_device_copyable_types.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test/basic_tests/implicit_device_copyable_types.cpp b/sycl/test/basic_tests/implicit_device_copyable_types.cpp index 0bda989649056..b1f72115f42af 100644 --- a/sycl/test/basic_tests/implicit_device_copyable_types.cpp +++ b/sycl/test/basic_tests/implicit_device_copyable_types.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify -Xclang -fsycl-is-device -DCOMPILE_ONLY %s +// RUN: %clangxx -fsycl -fsyntax-only %s #include #include