diff --git a/sycl/include/CL/sycl/handler.hpp b/sycl/include/CL/sycl/handler.hpp index da60f8258a0bb..982250a52879b 100644 --- a/sycl/include/CL/sycl/handler.hpp +++ b/sycl/include/CL/sycl/handler.hpp @@ -796,13 +796,42 @@ class __SYCL_EXPORT handler { } } + template + using remove_cv_ref_t = + typename std::remove_cv>::type; + + template + using is_same_type = std::is_same, remove_cv_ref_t>; + + template struct ShouldEnableSetArg { + static constexpr bool value = + std::is_trivially_copyable::value +#if CL_SYCL_LANGUAGE_VERSION && CL_SYCL_LANGUAGE_VERSION <= 121 + && std::is_standard_layout::value +#endif + || is_same_type::value // Sampler + || (!is_same_type::value && + std::is_pointer>::value) // USM + || is_same_type::value; // Interop + }; + /// Sets argument for OpenCL interoperability kernels. /// /// Registers Arg passed as argument # ArgIndex. /// /// \param ArgIndex is a positional number of argument to be set. /// \param Arg is an argument value to be set. - template void set_arg(int ArgIndex, T &&Arg) { + template + typename std::enable_if::value, void>::type + set_arg(int ArgIndex, T &&Arg) { + setArgHelper(ArgIndex, std::move(Arg)); + } + + template + void + set_arg(int ArgIndex, + accessor Arg) { setArgHelper(ArgIndex, std::move(Arg)); } diff --git a/sycl/test/basic_tests/set_arg_error.cpp b/sycl/test/basic_tests/set_arg_error.cpp new file mode 100644 index 0000000000000..4472b50bcd669 --- /dev/null +++ b/sycl/test/basic_tests/set_arg_error.cpp @@ -0,0 +1,47 @@ +// RUN: %clangxx -fsycl -Xclang -verify %s -I %sycl_include -Xclang -verify-ignore-unexpected=note,warning -fsyntax-only + +#include + +struct TriviallyCopyable { + int a; + int b; +}; + +struct NonTriviallyCopyable { + NonTriviallyCopyable() = default; + NonTriviallyCopyable(NonTriviallyCopyable const &) {} + int a; + int b; +}; + +struct NonStdLayout { + int a; + +private: + int b; +}; + +int main() { + constexpr size_t size = 1; + cl::sycl::buffer buf(size); + cl::sycl::queue q; + q.submit([&](cl::sycl::handler &h) { + auto global_acc = buf.get_access(h); + cl::sycl::sampler samp(cl::sycl::coordinate_normalization_mode::normalized, + cl::sycl::addressing_mode::clamp, + cl::sycl::filtering_mode::nearest); + cl::sycl::accessor + local_acc({size}, h); + h.set_arg(0, local_acc); + h.set_arg(1, global_acc); + h.set_arg(2, samp); + h.set_arg(3, TriviallyCopyable{}); + h.set_arg( // expected-error {{no matching member function for call to 'set_arg'}} + 4, NonTriviallyCopyable{}); +#if CL_SYCL_LANGUAGE_VERSION && CL_SYCL_LANGUAGE_VERSION <= 121 + h.set_arg( // expected-error {{no matching member function for call to 'set_arg'}} + 5, NonStdLayout{}); +#endif + }); +}