[SYCL] Consistent is_device_copyable for sycl::accessor

[inteltdelisle]

Previously, is_device_copyable_v evaluated to different values for device and host code. Added explicit instantiation to consistently return false. However, the spec requires that accessor be capturable by kernels, so added an extra condition to CheckFieldsAreDeviceCopyable specifically for accessors.
Also added tests for other basic sycl types.

[koparasy]

A high level comment. What makes accessors change their behavior depending on host|device compilation?

[inteltdelisle]

A high level comment. What makes accessors change their behavior depending on host|device compilation?

By default is_device_copyable simply falls back on std::is_trivially_copyable, which in parts look at what the ctors are doing. sycl::accessor ctors are implemented differently depending on whether ___SYCL_DEVICE_ONLY__ is defined and that is what was causing the difference.

[slawekptak]

@inteltdelisle thanks for the fix. Could you please describe a bit of a background for this change - was some specific test failing? Thanks!
Adding @gmlueck

[inteltdelisle]

This fix was of result of a bug report submitted by @koparasy, who noted that the following code was inconsistent between device and host compilation:

#include <sycl/sycl.hpp>

using AccT = sycl::accessor<int, 1, sycl::access_mode::read_write>;
using LAccT = sycl::local_accessor<int, 1>;

static_assert(
   sycl::is_device_copyable_v<AccT>,
   "DEVICE pass: accessor IS device-copyable (trivially copyable here)");
static_assert(sycl::is_device_copyable_v<LAccT>,
             "DEVICE pass: local_accessor IS device-copyable here");
void f() {}

This code compiled for the device but not the host. The additional tests are simply there to verify the behavior of other basic sycl types.

Worth noting that I did not add a test to verify the behavior of is_device_copyable<buffer<...>>, I am not sure what the desired behavior is for that type.

[gmlueck]

I wonder if we should change the implementation of the accessor instead of specializing is_device_copyable. For example, adding an empty user-provided destructor will cause the types to be non-trivially copyable.

This seems better than specializing the trait because it also makes is_trivially_copyable consistent between host and device.

[inteltdelisle]

I'll confirm locally whether that works. I was hesitant do change the accessor type itself, since that will have a less targeted impact. On the other hand the spec does state "Any type that is trivially copyable (as defined by the C++ core language) is implicitly device copyable." so that solution match the spec more closely.

[inteltdelisle]

@gmlueck Adding the dtors instead works locally. I can push that change if it's a preferred solution.

[gmlueck]

Spec-wise, I think it's better. Maybe @intel/llvm-reviewers-runtime can weigh in from the implementation side.

[uditagarwal97]

@inteltdelisle Changing a type's trivial-copyability is an ABI breaking change. I am on board with @gmlueck's suggestion, but I think we should add the user-provided destructor under PREVIEW_BREAKING_CHANGES flag.

[inteltdelisle]

Updated

[koparasy]

@gmlueck Does SYCL require std::is_trivially_copyable_v<accessor> itself to evaluate consistently with sycl::is_device_copyable_v<accessor>? If not, wouldn't an explicit specialization of is_device_copyable be preferable to deliberately making the destructor non-trivial, since the latter changes additional C++ type properties unrelated to device copyability?

I am also wondering whether we actually want to break code that may rely on std::is_trivially_copyable_v<accessor>, even if the current behavior is somewhat obscure and differs between host and device compilation.

[gmlueck]

In general, I think it's bad for traits to evaluate differently in the host vs. the device compilation passes because this can lead to constexpr expressions having different values in the two passes. I think this could lead to quite confusing behavior (e.g. SFINAE constraints evaluating different in the two passes). Therefore, I think std::is_trivially_copyable_v<accessor> should evaluate the same in both passes.

TBH, I think the SYCL spec should be changed to guarantee that all standard traits and all SYCL traits evaluate the same in all compiler passes. This is a battle for another day, though.

I agree that changing the trait in the device compiler could potentially break an application (thought I think it's pretty unlikely). Therefore, I agree that it's better to make this change at the next major release.

[inteltdelisle]

@gmlueck I'm not super familiar with the release procedure yet, what does making this change at the next major release entail precisely? Do we merge this change in behind the PREVIEW_BREAKING_CHANGES?

[gmlueck]

Yes, that's right. You would merge the change as it is now in this PR behind the PREVIEW_BREAKING_CHANGES macro. At the next major release, we search for all these macros and enable the code they protect by default. Merging the changes now under the macro provides two benefits:

• It allows user to see what potential breaking changes might be coming in the next major release by compiling with -fpreview-breaking-changes.
• It lets us address issues like this early, rather than waiting until right before the next major release.