diff --git a/sycl/include/CL/sycl/marray.hpp b/sycl/include/CL/sycl/marray.hpp index fa3ad0d1b0462..6bc2c7bfffe47 100644 --- a/sycl/include/CL/sycl/marray.hpp +++ b/sycl/include/CL/sycl/marray.hpp @@ -66,17 +66,9 @@ template class marray { typename = typename std::enable_if::type> constexpr marray(const ArgTN &... Args) : MData{Args...} {} - constexpr marray(const marray &Rhs) { - for (std::size_t I = 0; I < NumElements; ++I) { - MData[I] = Rhs.MData[I]; - } - } + constexpr marray(const marray &Rhs) = default; - constexpr marray(marray &&Rhs) { - for (std::size_t I = 0; I < NumElements; ++I) { - MData[I] = Rhs.MData[I]; - } - } + constexpr marray(marray &&Rhs) = default; // Available only when: NumElements == 1 template class marray { const_reference operator[](std::size_t index) const { return MData[index]; } - marray &operator=(const marray &Rhs) { - for (std::size_t I = 0; I < NumElements; ++I) { - MData[I] = Rhs.MData[I]; - } - return *this; - } + marray &operator=(const marray &Rhs) = default; // broadcasting operator marray &operator=(const Type &Rhs) { diff --git a/sycl/include/CL/sycl/types.hpp b/sycl/include/CL/sycl/types.hpp index 5f3de50acb96d..5bb77ec0b4372 100644 --- a/sycl/include/CL/sycl/types.hpp +++ b/sycl/include/CL/sycl/types.hpp @@ -2414,10 +2414,13 @@ struct is_device_copyable> : detail::bool_constant::value && is_device_copyable>::value> {}; -// marray is device copyable if element type 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). template -struct is_device_copyable, - std::enable_if_t::value>> +struct is_device_copyable< + sycl::marray, std::enable_if_t::value && + !std::is_trivially_copyable::value>> : std::true_type {}; // vec is device copyable on host, on device vec is trivially copyable diff --git a/sycl/test/basic_tests/marray/marray.cpp b/sycl/test/basic_tests/marray/marray.cpp index 148b6958691fc..1c5debd1fdd48 100755 --- a/sycl/test/basic_tests/marray/marray.cpp +++ b/sycl/test/basic_tests/marray/marray.cpp @@ -94,5 +94,31 @@ int main() { b___ = !mint3{0, 1, 2}; assert(b___[0] == true && b___[1] == false && b___[2] == false); + // check copyability + constexpr sycl::marray ma; + constexpr sycl::marray mb(ma); + constexpr sycl::marray mc = ma; + + // check trivially copyability + struct Copyable { + int a; + double b; + const char *name; + }; + + static_assert(std::is_trivially_copyable>::value, + "sycl::marray is not trivially copyable type"); + static_assert( + !std::is_trivially_copyable>::value, + "sycl::marray is trivially copyable type"); + + // check device copyability + static_assert(sycl::is_device_copyable, 5>>::value, + "sycl::marray, 5> is not device copyable type"); + static_assert(!sycl::is_device_copyable>::value, + "sycl::marray is device copyable type"); + + return 0; + return 0; } diff --git a/sycl/test/basic_tests/valid_kernel_args.cpp b/sycl/test/basic_tests/valid_kernel_args.cpp index e6bc307ddd0e7..1f05990b0ed49 100644 --- a/sycl/test/basic_tests/valid_kernel_args.cpp +++ b/sycl/test/basic_tests/valid_kernel_args.cpp @@ -23,6 +23,10 @@ struct SomeStructure { } v; }; +struct SomeMarrayStructure { + cl::sycl::marray points; +}; + #define CHECK_PASSING_TO_KERNEL_BY_VALUE(Type) \ static_assert(std::is_standard_layout::value, \ "Is not standard layouti type."); \ @@ -34,3 +38,4 @@ CHECK_PASSING_TO_KERNEL_BY_VALUE(int) CHECK_PASSING_TO_KERNEL_BY_VALUE(cl::sycl::cl_uchar4) CHECK_PASSING_TO_KERNEL_BY_VALUE(SomeStructure) #endif +CHECK_PASSING_TO_KERNEL_BY_VALUE(SomeMarrayStructure)