Skip to content

[SYCL] Extends support for SYCL 2020 implicitly device copyable types #8195

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
bader merged 16 commits into from
Feb 15, 2023
Merged

[SYCL] Extends support for SYCL 2020 implicitly device copyable types #8195

Show file tree
Hide file tree
Changes from 12 commits
Commits
Show all changes
16 commits
Select commit Hold shift + click to select a range
cbf6abe
Extend support for implicitly device copyable types
maarquitos14 Feb 1, 2023
cd4b93d
Adds tests for implicit device copyable types.
maarquitos14 Feb 1, 2023
6b02640
Extend support for SYCL2020 implicitly device copyable types.
maarquitos14 Feb 3, 2023
80c8fc6
Fix clang-format issues.
maarquitos14 Feb 3, 2023
e36f4b8
Fix more clang-format issues.
maarquitos14 Feb 3, 2023
7fcac67
Addressing code review comments.
maarquitos14 Feb 8, 2023
61c6e12
Extends is_device_copyable to consider cv-qualifiers.
maarquitos14 Feb 9, 2023
918c61a
Extends testing with cv-qualifiers.
maarquitos14 Feb 9, 2023
8804479
Undoing unnecessary comment.
maarquitos14 Feb 9, 2023
290c72e
Fixes clang-format issues.
maarquitos14 Feb 9, 2023
982208a
Merge remote-tracking branch 'intel/origin/sycl' into maronas/implici…
maarquitos14 Feb 9, 2023
10913de
Recovers accidentally removed line.
maarquitos14 Feb 9, 2023
ad1a710
Corrects test behaviour.
maarquitos14 Feb 10, 2023
d8e5c39
Fixes issues with MSVC.
maarquitos14 Feb 10, 2023
7e61201
Removes incorrect test.
maarquitos14 Feb 10, 2023
c82955e
Removes unnecessary flags from test.
maarquitos14 Feb 13, 2023
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
7 changes: 0 additions & 7 deletions sycl/include/sycl/ext/oneapi/properties/properties.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -232,12 +232,5 @@ struct is_device_copyable<
std::enable_if_t<!std::is_trivially_copyable<
ext::oneapi::experimental::properties<PropertiesT>>::value>>
: is_device_copyable<PropertiesT> {};
template <typename PropertiesT>
struct is_device_copyable<
const ext::oneapi::experimental::properties<PropertiesT>,
std::enable_if_t<!std::is_trivially_copyable<
const ext::oneapi::experimental::properties<PropertiesT>>::value>>
: is_device_copyable<PropertiesT> {};

} // __SYCL_INLINE_VER_NAMESPACE(_V1)
} // namespace sycl
63 changes: 61 additions & 2 deletions sycl/include/sycl/types.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -58,6 +58,7 @@
#include <array>
#include <cmath>
#include <cstring>
#include <variant>
#ifndef __SYCL_DEVICE_ONLY__
#include <cfenv>
#endif
Expand Down Expand Up @@ -2393,6 +2394,12 @@ __SYCL_DECLARE_FLOAT_VECTOR_CONVERTERS(double)
template <typename T, typename = void>
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<T, N>, std::optional<T>, std::variant<T>,
// sycl::marray<T> and T[N]. Thus, specializations for all these mentioned
// types are guarded by `std::enable_if_t<!std::is_trivially_copyable<...>>`
// so that they are candidates only for non-trivially-copyable types.
// Otherwise, there are several candidates and the compiler can't decide.
template <typename T>
struct is_device_copyable<
T, std::enable_if_t<std::is_trivially_copyable<T>::value>>
Expand All @@ -2401,6 +2408,31 @@ struct is_device_copyable<
template <typename T>
inline constexpr bool is_device_copyable_v = is_device_copyable<T>::value;

// std::array<T, 0> is implicitly device copyable type.
template <typename T>
struct is_device_copyable<std::array<T, 0>> : std::true_type {};

// std::array<T, N> is implicitly device copyable type if T is device copyable
template <typename T, std::size_t N>
struct is_device_copyable<
std::array<T, N>,
std::enable_if_t<!std::is_trivially_copyable<std::array<T, N>>::value>>
: is_device_copyable<T> {};

// std::optional<T> is implicitly device copyable type if T is device copyable
template <typename T>
struct is_device_copyable<
std::optional<T>,
std::enable_if_t<!std::is_trivially_copyable<std::optional<T>>::value>>
: is_device_copyable<T> {};

// std::pair<T1, T2> is implicitly device copyable type if T1 and T2 are device
// copyable
template <typename T1, typename T2>
struct is_device_copyable<std::pair<T1, T2>>
: detail::bool_constant<is_device_copyable<T1>::value &&
is_device_copyable<T2>::value> {};

// std::tuple<> is implicitly device copyable type.
template <> struct is_device_copyable<std::tuple<>> : std::true_type {};

Expand All @@ -2411,6 +2443,17 @@ struct is_device_copyable<std::tuple<T, Ts...>>
: detail::bool_constant<is_device_copyable<T>::value &&
is_device_copyable<std::tuple<Ts...>>::value> {};

// std::variant<> is implicitly device copyable type
template <> struct is_device_copyable<std::variant<>> : std::true_type {};

// std::variant<Ts...> is implicitly device copyable type if each type T of
// Ts... is device copyable
template <typename... Ts>
struct is_device_copyable<
std::variant<Ts...>,
std::enable_if_t<!std::is_trivially_copyable<std::variant<Ts...>>::value>>
: is_device_copyable<Ts...> {};

// 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).
Expand All @@ -2420,6 +2463,18 @@ struct is_device_copyable<
!std::is_trivially_copyable<T>::value>>
: std::true_type {};

// array is device copyable if element type is device copyable
template <typename T, std::size_t N>
struct is_device_copyable<
T[N], std::enable_if_t<!std::is_trivially_copyable<T>::value>>
: is_device_copyable<T> {};

template <typename T>
struct is_device_copyable<
T, std::enable_if_t<!std::is_trivially_copyable<T>::value &&
(std::is_const_v<T> || std::is_volatile_v<T>)>>
: is_device_copyable<std::remove_cv_t<T>> {};

namespace detail {
template <typename T, typename = void>
struct IsDeprecatedDeviceCopyable : std::false_type {};
Expand All @@ -2433,9 +2488,13 @@ struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020")
std::is_trivially_destructible<T>::value &&
!is_device_copyable<T>::value>> : std::true_type {};

template <typename T, int N>
struct __SYCL2020_DEPRECATED("This type isn't device copyable in SYCL 2020")
IsDeprecatedDeviceCopyable<T[N]> : IsDeprecatedDeviceCopyable<T> {};

#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 <typename T, unsigned NumFieldsToCheck>
struct CheckFieldsAreDeviceCopyable
: CheckFieldsAreDeviceCopyable<T, NumFieldsToCheck - 1> {
Expand Down
152 changes: 152 additions & 0 deletions sycl/test/basic_tests/implicit_device_copyable_types.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,152 @@
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
Copy link
Contributor

@aelovikov-intel aelovikov-intel Feb 3, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I'd like to have more runtime checks that the values of those types actually match between host/device, i.e. check the content of pair.first and/or pair.second.

Copy link
Contributor

@aelovikov-intel aelovikov-intel Feb 3, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Just to clarify - this is my main concern, two other comments are more subjective in nature and thus optional to implement.

Copy link
Contributor

@AlexeySachkov AlexeySachkov Feb 6, 2023 •
edited
Loading

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I suggest that we perhaps have two test cases: one which simply contains a bunch of static_assert(sycl::is_device_copyable_v<some_type_here>) to check that we properly specialize the trait and another one, which checks runtime behavior, i.e. that we are actually able to properly copy values of those types to kernels and read them back.

Copy link
Contributor Author

@maarquitos14 maarquitos14 Feb 8, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I have submitted a PR to llvm-test-suite repo adding tests that check the runtime behavior for this.

Copy link
Contributor

@AlexeySachkov AlexeySachkov Feb 9, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

If this test is supposed to be compile-only, then I suggest to remove -o %t.out and add -fsyntax-only, see docs for some BKMs. You can also remove any code which submits kernels, because that will be checked in the E2E test you have


#include <sycl/sycl.hpp>
#include <variant>

struct ACopyable {
int i;
ACopyable() = default;
ACopyable(int _i) : i(_i) {}
ACopyable(const ACopyable &x) : i(x.i) {}
};

template <> struct sycl::is_device_copyable<ACopyable> : std::true_type {};

int main() {
static_assert(sycl::is_device_copyable_v<std::pair<int, float>>);
static_assert(sycl::is_device_copyable_v<std::pair<ACopyable, float>>);
static_assert(sycl::is_device_copyable_v<std::tuple<int, float, bool>>);
static_assert(sycl::is_device_copyable_v<std::tuple<ACopyable, float, bool>>);
static_assert(sycl::is_device_copyable_v<std::variant<int, float, bool>>);
static_assert(sycl::is_device_copyable_v<std::variant<ACopyable>>);
static_assert(sycl::is_device_copyable_v<std::array<int, 513>>);
static_assert(sycl::is_device_copyable_v<std::array<ACopyable, 513>>);
static_assert(sycl::is_device_copyable_v<std::optional<int>>);
static_assert(sycl::is_device_copyable_v<std::optional<ACopyable>>);
static_assert(sycl::is_device_copyable_v<std::string_view>);
#if __cpp_lib_span >= 202002
static_assert(sycl::is_device_copyable_v<std::span<int>>);
#endif
static_assert(sycl::is_device_copyable_v<const sycl::span<int>>);

// const
static_assert(sycl::is_device_copyable_v<const std::pair<int, float>>);
static_assert(sycl::is_device_copyable_v<const std::pair<ACopyable, float>>);
static_assert(sycl::is_device_copyable_v<const std::tuple<int, float, bool>>);
static_assert(
sycl::is_device_copyable_v<const std::tuple<ACopyable, float, bool>>);
static_assert(
sycl::is_device_copyable_v<const std::variant<int, float, bool>>);
static_assert(sycl::is_device_copyable_v<const std::variant<ACopyable>>);
static_assert(sycl::is_device_copyable_v<const std::array<int, 513>>);
static_assert(sycl::is_device_copyable_v<const std::array<ACopyable, 513>>);
static_assert(sycl::is_device_copyable_v<const std::optional<int>>);
static_assert(sycl::is_device_copyable_v<const std::optional<ACopyable>>);
static_assert(sycl::is_device_copyable_v<const std::string_view>);
#if __cpp_lib_span >= 202002
static_assert(sycl::is_device_copyable_v<const std::span<int>>);
#endif
static_assert(sycl::is_device_copyable_v<const sycl::span<int>>);

// volatile
static_assert(sycl::is_device_copyable_v<volatile std::pair<int, float>>);
static_assert(
sycl::is_device_copyable_v<volatile std::pair<ACopyable, float>>);
static_assert(
sycl::is_device_copyable_v<volatile std::tuple<int, float, bool>>);
static_assert(
sycl::is_device_copyable_v<volatile std::tuple<ACopyable, float, bool>>);
static_assert(
sycl::is_device_copyable_v<volatile std::variant<int, float, bool>>);
static_assert(sycl::is_device_copyable_v<volatile std::variant<ACopyable>>);
static_assert(sycl::is_device_copyable_v<volatile std::array<int, 513>>);
static_assert(
sycl::is_device_copyable_v<volatile std::array<ACopyable, 513>>);
static_assert(sycl::is_device_copyable_v<volatile std::optional<int>>);
static_assert(sycl::is_device_copyable_v<volatile std::optional<ACopyable>>);
static_assert(sycl::is_device_copyable_v<volatile std::string_view>);
#if __cpp_lib_span >= 202002
static_assert(sycl::is_device_copyable_v<volatile std::span<int>>);
#endif
static_assert(sycl::is_device_copyable_v<volatile sycl::span<int>>);

// const volatile
static_assert(
sycl::is_device_copyable_v<const volatile std::pair<int, float>>);
static_assert(
sycl::is_device_copyable_v<const volatile std::pair<ACopyable, float>>);
static_assert(
sycl::is_device_copyable_v<const volatile std::tuple<int, float, bool>>);
static_assert(sycl::is_device_copyable_v<
const volatile std::tuple<ACopyable, float, bool>>);
static_assert(sycl::is_device_copyable_v<
const volatile std::variant<int, float, bool>>);
static_assert(
sycl::is_device_copyable_v<const volatile std::variant<ACopyable>>);
static_assert(
sycl::is_device_copyable_v<const volatile std::array<int, 513>>);
static_assert(
sycl::is_device_copyable_v<const volatile std::array<ACopyable, 513>>);
static_assert(sycl::is_device_copyable_v<const volatile std::optional<int>>);
static_assert(
sycl::is_device_copyable_v<const volatile std::optional<ACopyable>>);
static_assert(sycl::is_device_copyable_v<const volatile std::string_view>);
#if __cpp_lib_span >= 202002
static_assert(sycl::is_device_copyable_v<const volatile std::span<int>>);
#endif
static_assert(sycl::is_device_copyable_v<const volatile sycl::span<int>>);

#if COMPILE_ONLY
sycl::queue q;
{
std::variant<ACopyable> variant_arr[5];
std::variant<ACopyable> 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
Copy link
Contributor

@aelovikov-intel aelovikov-intel Feb 3, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Do we verify it anywhere? If not, I'd suggest to modify this test as

// RUN: compile and run
// RUN: compile-only -fsyntax-only -Xclang <enable error/warning verification in clang> -DCOMPILE_ONLY
...
#if COMPILE_ONLY
  {
    std::variant<ACopyable>
    // try to pass it to device
    // expected-error: ...
  }
#endif

Copy link
Contributor Author

@maarquitos14 maarquitos14 Feb 10, 2023 •
edited
Loading

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I developed this test and tried both in Linux and Windows. Turns out there's no error in Windows (I guess MSVC doesn't require virtual functions to implement std::variant), so I don't think we should have this test. Also, I will update my comment to say that this only happens in some implementations of std::variant.

auto size = sizeof(variant_arr[0]);
size = sizeof(variant);
});
}).wait_and_throw();
}
{
const std::variant<ACopyable> variant_arr[5];
const std::variant<ACopyable> 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<ACopyable> variant_arr[5];
volatile std::variant<ACopyable> 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<ACopyable> variant_arr[5];
const volatile std::variant<ACopyable> 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;
}