diff --git a/sycl/include/sycl/ext/intel/esimd/detail/simd_obj_impl.hpp b/sycl/include/sycl/ext/intel/esimd/detail/simd_obj_impl.hpp index 7d7681b6dba20..bd74e70fcf145 100644 --- a/sycl/include/sycl/ext/intel/esimd/detail/simd_obj_impl.hpp +++ b/sycl/include/sycl/ext/intel/esimd/detail/simd_obj_impl.hpp @@ -15,7 +15,6 @@ #include #include #include -#include #include #include @@ -1038,8 +1037,8 @@ class [[__sycl_detail__::__uses_aspects__( /** @tparam SimdT The argument object type(auto-deduced). */ \ /** @param RHS The argument object. */ \ template == \ - is_simd_type_v)&&COND>> \ + class = std::enable_if_t< \ + (is_simd_type_v == is_simd_type_v) && COND>> \ Derived &operator OPASSIGN( \ const __ESIMD_DNS::simd_obj_impl &RHS) { \ auto Res = *this BINOP RHS; \ @@ -1056,10 +1055,9 @@ class [[__sycl_detail__::__uses_aspects__( /** @param RHS The argument object. */ \ template == \ - is_simd_type_v)&&(RegionT1::length == length) && \ - COND>> \ + class = std::enable_if_t<(is_simd_type_v == \ + is_simd_type_v) && \ + (RegionT1::length == length) && COND>> \ Derived &operator OPASSIGN( \ const __ESIMD_NS::simd_view &RHS) { \ auto Res = *this BINOP RHS.read(); \ @@ -1132,9 +1130,6 @@ class [[__sycl_detail__::__uses_aspects__( #undef __ESIMD_ARITH_OP_FILTER #undef __ESIMD_DEF_SIMD_OBJ_IMPL_OPASSIGN - // Getter for the test proxy member, if enabled - __ESIMD_DECLARE_TEST_PROXY_ACCESS - private: // The underlying data for this vector. raw_vector_type M_data; @@ -1159,9 +1154,6 @@ class [[__sycl_detail__::__uses_aspects__( PropertyListT = {}) SYCL_ESIMD_FUNCTION; protected: - // The test proxy if enabled - __ESIMD_DECLARE_TEST_PROXY - void set(const raw_vector_type &Val) { #ifndef __SYCL_DEVICE_ONLY__ M_data = Val; diff --git a/sycl/include/sycl/ext/intel/esimd/detail/simd_view_impl.hpp b/sycl/include/sycl/ext/intel/esimd/detail/simd_view_impl.hpp index 454e10719e50f..9860d23dcbcc1 100644 --- a/sycl/include/sycl/ext/intel/esimd/detail/simd_view_impl.hpp +++ b/sycl/include/sycl/ext/intel/esimd/detail/simd_view_impl.hpp @@ -12,7 +12,6 @@ #include #include -#include #include namespace sycl { @@ -300,10 +299,9 @@ class [[__sycl_detail__::__uses_aspects__( /* OPASSIGN simd_obj_impl */ \ template == \ - is_simd_type_v)&&(N1 == length) && \ - COND>> \ + class = std::enable_if_t<(is_simd_type_v == \ + is_simd_type_v) && \ + (N1 == length) && COND>> \ Derived &operator OPASSIGN(const simd_obj_impl &RHS) { \ auto Res = read() BINOP RHS; \ write(Res); \ @@ -316,9 +314,8 @@ class [[__sycl_detail__::__uses_aspects__( typename __ESIMD_NS::shape_type::element_type, \ class T = element_type, class SimdT = BaseTy, \ class = std::enable_if_t< \ - (is_simd_type_v == is_simd_type_v)&&( \ - length == __ESIMD_NS::shape_type::length) && \ - COND>> \ + (is_simd_type_v == is_simd_type_v) && \ + (length == __ESIMD_NS::shape_type::length) && COND>> \ Derived &operator OPASSIGN(const simd_view_impl &RHS) { \ *this OPASSIGN RHS.read(); \ return cast_this_to_derived(); \ @@ -404,14 +401,10 @@ class [[__sycl_detail__::__uses_aspects__( /// region viewed by this object. /// @param Other The source rvalue object. /// @return This object cast to the derived class. - Derived &operator=(Derived &&Other) { - __esimd_move_test_proxy(Other); - return write(Other.read()); - } + Derived &operator=(Derived &&Other) { return write(Other.read()); } /// Move assignment operator. Updates the target region viewed by this object. simd_view_impl &operator=(simd_view_impl &&Other) { - __esimd_move_test_proxy(Other); return write(Other.read()); } @@ -425,8 +418,8 @@ class [[__sycl_detail__::__uses_aspects__( /// @return This object cast to the derived class. template == - is_simd_type_v)&&(length == - SimdT::length)>> + is_simd_type_v) && + (length == SimdT::length)>> Derived &operator=(const simd_obj_impl &Other) { return write(convert_vector( Other.data())); @@ -587,16 +580,10 @@ class [[__sycl_detail__::__uses_aspects__( /// @cond EXCLUDE public: - // Getter for the test proxy member, if enabled - __ESIMD_DECLARE_TEST_PROXY_ACCESS - protected: // The reference to the base object, which must be a simd object BaseTy &M_base; - // The test proxy if enabled - __ESIMD_DECLARE_TEST_PROXY - // The region applied on the base object. Its type could be // - region1d_t // - region2d_t diff --git a/sycl/include/sycl/ext/intel/esimd/detail/test_proxy.hpp b/sycl/include/sycl/ext/intel/esimd/detail/test_proxy.hpp deleted file mode 100644 index a42571a673216..0000000000000 --- a/sycl/include/sycl/ext/intel/esimd/detail/test_proxy.hpp +++ /dev/null @@ -1,111 +0,0 @@ -//==-------------- test_proxy.hpp - DPC++ Explicit SIMD API ----------------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// -// Test proxy to differentiate move and copy constructor calls -//===----------------------------------------------------------------------===// - -#pragma once - -/// @cond ESIMD_DETAIL - -// The test proxy if solely for the test purposes, so it's off by default, with -// no any code generated. It is enabled only if the __ESIMD_ENABLE_TEST_PROXY -// macro is defined. -// It's expected for the proxy class to be available in device code, so that it -// could be incorporated into the ESIMD API classes. Though there is no reason -// to limit it to the __SYCL_DEVICE_ONLY__. -#ifndef __ESIMD_ENABLE_TEST_PROXY - -// No code generation by default -#define __ESIMD_DECLARE_TEST_PROXY -#define __ESIMD_DECLARE_TEST_PROXY_ACCESS -#define __esimd_move_test_proxy(other) - -#else - -// Declare the class attribute -// -// We are using non static data member initialization approach to force -// the value required. Initialization will take place even if no -// default/copy/move constructor of the test_proxy class was explcitly -// called by any of the user-defined constructors of the proxy target -#define __ESIMD_DECLARE_TEST_PROXY \ - esimd::detail::test::test_proxy M_testProxy = \ - esimd::detail::test::test_proxy(); - -// Declare the getter to access the proxy from the tests -#define __ESIMD_DECLARE_TEST_PROXY_ACCESS \ - const auto &get_test_proxy() const { return M_testProxy; } - -// Test proxy will be handled in a proper way by default/implicit move -// constructors and move operators. -// Still the user-defined constructors or move operators should explicitly state -// what to do with each of class atributes, so a proper wrapper required -// -// We are using a simple do-while trick to make sure no code breakage could -// possibly occur in case macro becomes multistatement (PRE10-C in SEI CERT C) -#define __esimd_move_test_proxy(other) \ - do { \ - M_testProxy = std::move(other.M_testProxy); \ - } while (false) - -namespace sycl { -inline namespace _V1 { -namespace ext::intel::esimd::detail::test { - -// The test_proxy class. -// Being intended solely for the test purposes, it is enabled only if the -// __ESIMD_ENABLE_TEST_PROXY macro is defined, which is off by default. -// -// This is a helper class for tests to differentiate between the copy -// constructor/assignment and the move constructor/assignment calls, -// as the copy constructor works as the default fallback for every case with -// move constructor disabled or not provided -// -// It is expected for the class with the test proxy (the class under test) to: -// - provide the get_test_proxy() method -// - properly handle moving the test_proxy member in user-defined move -// constructors and user-defined assignment operators -// -// Therefore the following expression is expected to return `true` only if the -// move constructor or move operator was called for the instance of the class -// under test: -// instance.get_test_proxy().was_move_destination() -// -class test_proxy { - // Define the default value to use for every constructor - bool M_move_destination = false; - -public: - test_proxy() { __esimd_dbg_print(test_proxy()); } - - test_proxy(const test_proxy &) { - __esimd_dbg_print(test_proxy(const test_proxy &other)); - } - test_proxy(test_proxy &&) { - __esimd_dbg_print(test_proxy(test_proxy && other)); - M_move_destination = true; - } - test_proxy &operator=(const test_proxy &) { - __esimd_dbg_print(test_proxy::operator=(const test_proxy &other)); - return *this; - } - test_proxy &operator=(test_proxy &&) { - __esimd_dbg_print(test_proxy::operator=(test_proxy &&other)); - M_move_destination = true; - return *this; - } - bool was_move_destination() const { return M_move_destination; } -}; - -} // namespace ext::intel::esimd::detail::test -} // namespace _V1 -} // namespace sycl - -#endif // __ESIMD_ENABLE_TEST_PROXY - -/// @endcond ESIMD_DETAIL diff --git a/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move.hpp b/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move.hpp deleted file mode 100644 index 6805bc54b8416..0000000000000 --- a/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move.hpp +++ /dev/null @@ -1,209 +0,0 @@ -//===-- ctor_move.hpp - Functions for tests on simd vector constructor -// definition. -------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// -/// -/// \file -/// This file provides functions for tests on simd move constructor. -/// -//===----------------------------------------------------------------------===// - -#pragma once -#define ESIMD_TESTS_DISABLE_DEPRECATED_TEST_DESCRIPTION_FOR_LOGS - -#include "common.hpp" - -#include -#include - -namespace esimd = sycl::ext::intel::esimd; - -namespace esimd_test::api::functional::ctors { - -// Uses the initializer C++ context to call simd move constructor -struct initializer { - static std::string get_description() { return "initializer"; } - - template - static void run(SimdT &&source, const ActionT &action) { - static_assert( - type_traits::is_nonconst_rvalue_reference_v); - - const auto instance = SimdT(std::move(source)); - action(instance); - } -}; - -// Uses the variable declaration C++ context to call simd move constructor -struct var_decl { - static std::string get_description() { return "variable declaration"; } - - template - static void run(SimdT &&source, const ActionT &action) { - static_assert( - type_traits::is_nonconst_rvalue_reference_v); - - const auto instance(std::move(source)); - action(instance); - } -}; - -// Uses the rvalue in expression C++ context to call simd move constructor -struct rval_in_expr { - static std::string get_description() { return "rvalue in expression"; } - - template - static void run(SimdT &&source, const ActionT &action) { - static_assert( - type_traits::is_nonconst_rvalue_reference_v); - - SimdT instance; - instance = SimdT(std::move(source)); - action(instance); - } -}; - -// Uses the function argument C++ context to call simd move constructor -class const_ref { -public: - static std::string get_description() { return "const reference"; } - - template - static void run(SimdT &&source, const ActionT &action) { - static_assert( - type_traits::is_nonconst_rvalue_reference_v); - - action(SimdT(std::move(source))); - } -}; - -// The core test functionality. -// Runs a TestCaseT, specific for each C++ context, for a simd -// instance -template class run_test { - static constexpr int NumElems = SizeT::value; - using KernelName = Kernel; - using TestDescriptionT = ctors::TestDescription; - -public: - bool operator()(sycl::queue &queue, const std::string &data_type) { - bool passed = true; - bool was_moved = false; - log::trace(data_type); - - if (should_skip_test_with(queue.get_device())) { - return true; - } - - const shared_allocator data_allocator(queue); - const shared_allocator flags_allocator(queue); - const auto reference = generate_ref_data(); - - shared_vector input(reference.cbegin(), reference.cend(), - data_allocator); - shared_vector result(reference.size(), data_allocator); - - // We need a special handling for case of simd, as we need to check - // more than a single data value; therefore we need to loop over the - // reference data to run test - if constexpr (NumElems == 1) { - const auto n_checks = input.size(); - const sycl::range<1> range(n_checks); - - // We need a separate flag per each check to have a parallel_for possible, - // because any concurrent access to the same memory location is UB, even - // in case we are updating variable to the same value in multiple threads - shared_vector flags_storage(n_checks, flags_allocator); - - // Run check for each of the reference elements using a single work-item - // per single element - queue.submit([&](sycl::handler &cgh) { - const DataT *const ptr_in = input.data(); - const auto ptr_out = result.data(); - const auto ptr_flags = flags_storage.data(); - - cgh.parallel_for( - range, [=](sycl::id<1> id) SYCL_ESIMD_KERNEL { - const auto work_item_index = id[0]; - // Access a separate memory areas from each of the work items - const DataT *const in = ptr_in + work_item_index; - const auto out = ptr_out + work_item_index; - const auto was_moved_flag = ptr_flags + work_item_index; - *was_moved_flag = run_check(in, out); - }); - }); - queue.wait_and_throw(); - - // Oversafe: verify the proper signature was called for every check - was_moved = std::all_of(flags_storage.cbegin(), flags_storage.cend(), - [](int flag) { return flag; }); - } else { - assert((input.size() == NumElems) && - "Unexpected size of the input vector"); - - shared_vector flags_storage(1, flags_allocator); - - queue.submit([&](sycl::handler &cgh) { - const DataT *const in = input.data(); - const auto out = result.data(); - const auto was_moved_flag = flags_storage.data(); - - cgh.single_task( - [=]() SYCL_ESIMD_KERNEL { *was_moved_flag = run_check(in, out); }); - }); - queue.wait_and_throw(); - - was_moved = flags_storage[0]; - } - - if (!was_moved) { - passed = false; - log::fail(TestDescriptionT(data_type), - "A copy constructor instead of a move constructor was used."); - } else { - for (size_t i = 0; i < reference.size(); ++i) { - const auto &retrieved = result[i]; - const auto &expected = reference[i]; - - if (!are_bitwise_equal(retrieved, expected)) { - passed = false; - log::fail(TestDescriptionT(data_type), "Unexpected value at index ", - i, ", retrieved: ", retrieved, ", expected: ", expected); - } - } - } - - return passed; - } - -private: - // The core check logic. - // Uses USM pointers for input data and to store the data from the new simd - // instance, so that we could check it later - // Returns the flag that should be true only if the move constructor was - // actually called, to differentiate with the copy constructor calls - static bool run_check(const DataT *const in, DataT *const out) { - bool was_moved = false; - - // Prepare the source simd to move - esimd::simd source; - source.copy_from(in); - - // Action to run over the simd move constructor result - const auto action = [&](const esimd::simd &instance) { - was_moved = instance.get_test_proxy().was_move_destination(); - instance.copy_to(out); - }; - // Call the move constructor in the specific context and run action - // directly over the simd move constructor result - TestCaseT::template run(std::move(source), action); - - return was_moved; - } -}; - -} // namespace esimd_test::api::functional::ctors diff --git a/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move_core.cpp b/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move_core.cpp deleted file mode 100644 index baceb2d7ff8a6..0000000000000 --- a/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move_core.cpp +++ /dev/null @@ -1,49 +0,0 @@ -//==------- ctor_move_core.cpp - DPC++ ESIMD on-device test ---------------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// -// RUN: %{build} -fsycl-device-code-split=per_kernel -o %t.out -// RUN: %{run} %t.out -// XFAIL: * -// TODO Remove XFAIL once the simd vector provides move constructor -// -// Test for esimd move constructor -// The test creates source simd instance with reference data and invokes move -// constructor in different C++ contexts to create a new simd instance from the -// source simd instance. It is expected for a new simd instance to store -// bitwise same data as the one passed as the source simd constructor. - -// The following issues for simd observed: -// - freeze with T in {char, unsigned char, singned char}; -// - runtime failure with T in {short, unsigned short}. -// TODO Remove once the freeze is fixed -#define SKIP_VECTOR_LEN_32 - -// The test proxy is used to verify the move constructor was called actually. -#define __ESIMD_ENABLE_TEST_PROXY - -#include "ctor_move.hpp" - -using namespace esimd_test::api::functional; - -int main(int, char **) { - bool passed = true; - const auto types = get_tested_types(); - const auto sizes = get_all_sizes(); - const auto contexts = - unnamed_type_pack::generate(); - - sycl::queue queue(esimd_test::ESIMDSelector, - esimd_test::createExceptionHandler()); - - // Run test for all combinations possible - passed &= - for_all_combinations(types, sizes, contexts, queue); - - std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n"); - return passed ? 0 : 1; -} diff --git a/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move_fp_extra.cpp b/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move_fp_extra.cpp deleted file mode 100644 index b5695783a303b..0000000000000 --- a/sycl/test-e2e/ESIMD/api/functional/ctors/ctor_move_fp_extra.cpp +++ /dev/null @@ -1,43 +0,0 @@ -//==------- ctor_move_fp_extra.cpp - DPC++ ESIMD on-device test -----------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// -// RUN: %{build} -fsycl-device-code-split=per_kernel -o %t.out -// RUN: %{run} %t.out -// XFAIL: * -// TODO Remove XFAIL once the simd vector provides move constructor -// -// Test for esimd move constructor -// The test creates source simd instance with reference data and invokes move -// constructor in different C++ contexts to create a new simd instance from the -// source simd instance. It is expected for a new simd instance to store -// bitwise same data as the one passed as the source simd constructor. - -// The test proxy is used to verify the move constructor was called actually. -#define __ESIMD_ENABLE_TEST_PROXY - -#include "ctor_move.hpp" - -using namespace esimd_test::api::functional; - -int main(int, char **) { - bool passed = true; - const auto types = get_tested_types(); - const auto sizes = get_all_sizes(); - const auto contexts = - unnamed_type_pack::generate(); - - sycl::queue queue(esimd_test::ESIMDSelector, - esimd_test::createExceptionHandler()); - - // Run test for all combinations possible - passed &= - for_all_combinations(types, sizes, contexts, queue); - - std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n"); - return passed ? 0 : 1; -} diff --git a/sycl/test-e2e/ESIMD/api/functional/operators/operator_assignment_move_and_copy_core.cpp b/sycl/test-e2e/ESIMD/api/functional/operators/operator_assignment_copy_core.cpp similarity index 66% rename from sycl/test-e2e/ESIMD/api/functional/operators/operator_assignment_move_and_copy_core.cpp rename to sycl/test-e2e/ESIMD/api/functional/operators/operator_assignment_copy_core.cpp index 5c85ab28ea944..230c64d650777 100644 --- a/sycl/test-e2e/ESIMD/api/functional/operators/operator_assignment_move_and_copy_core.cpp +++ b/sycl/test-e2e/ESIMD/api/functional/operators/operator_assignment_copy_core.cpp @@ -8,11 +8,9 @@ //===----------------------------------------------------------------------===// // RUN: %{build} -fsycl-device-code-split=per_kernel -o %t.out // RUN: %{run} %t.out -// XFAIL: * -// TODO Remove XFAIL once the simd vector provides move assignment operator // -// Test for simd move and copy assignment operators. -// The test creates source simd instance with reference data and invokes move or +// Test for simd copy assignment operators. +// The test creates source simd instance with reference data and invokes // copy assignment operator from source to destination simd instance. It is // expected for destination simd instance to store a bitwise same data as the // reference one. @@ -22,23 +20,6 @@ using namespace sycl::ext::intel::esimd; using namespace esimd_test::api::functional; -// Descriptor class for the case of calling move assignment operator. -struct move_assignment { - static std::string get_description() { return "move assignment operator"; } - - static constexpr bool is_move_expected() { return true; } - - template - static bool run(const DataT *const ref_data, DataT *const out) { - simd source_simd; - source_simd.copy_from(ref_data); - simd simd_obj; - simd_obj = std::move(source_simd); - simd_obj.copy_to(out); - return simd_obj.get_test_proxy().was_move_destination(); - } -}; - // Descriptor class for the case of calling copy assignment operator. struct copy_assignment { static std::string get_description() { return "copy assignment operator"; } @@ -52,7 +33,7 @@ struct copy_assignment { simd simd_obj; simd_obj = source_simd; simd_obj.copy_to(out); - return simd_obj.get_test_proxy().was_move_destination(); + return false; } }; @@ -62,11 +43,10 @@ int main(int, char **) { bool passed = true; - const auto types = get_tested_types(); + const auto types = get_tested_types(); const auto all_sizes = get_all_sizes(); - const auto context = - unnamed_type_pack::generate(); + const auto context = unnamed_type_pack::generate(); passed &= for_all_combinations(types, all_sizes, context, queue);