From 514789815070253711a7069efd194b4754a2dc71 Mon Sep 17 00:00:00 2001 From: "Klochkov, Denis" Date: Mon, 25 May 2026 17:14:03 +0200 Subject: [PATCH 1/6] [SYCL][TEST] add unit and e2e tests of multi ptr --- sycl/include/sycl/multi_ptr.hpp | 6 +- sycl/test-e2e/multi_ptr/prefetch.cpp | 245 ++++++++++ .../test/multi_ptr/restricted_conversions.cpp | 88 ++++ sycl/unittests/CMakeLists.txt | 1 + sycl/unittests/multi_ptr/Accessors.cpp | 130 ++++++ sycl/unittests/multi_ptr/CMakeLists.txt | 9 + sycl/unittests/multi_ptr/Constructors.cpp | 194 ++++++++ sycl/unittests/multi_ptr/Conversion.cpp | 432 ++++++++++++++++++ .../multi_ptr/NegativeConversions.cpp | 143 ++++++ sycl/unittests/multi_ptr/Operators.cpp | 247 ++++++++++ .../multi_ptr/VoidSpecialization.cpp | 358 +++++++++++++++ 11 files changed, 1852 insertions(+), 1 deletion(-) create mode 100644 sycl/test-e2e/multi_ptr/prefetch.cpp create mode 100644 sycl/test/multi_ptr/restricted_conversions.cpp create mode 100644 sycl/unittests/multi_ptr/Accessors.cpp create mode 100644 sycl/unittests/multi_ptr/CMakeLists.txt create mode 100644 sycl/unittests/multi_ptr/Constructors.cpp create mode 100644 sycl/unittests/multi_ptr/Conversion.cpp create mode 100644 sycl/unittests/multi_ptr/NegativeConversions.cpp create mode 100644 sycl/unittests/multi_ptr/Operators.cpp create mode 100644 sycl/unittests/multi_ptr/VoidSpecialization.cpp diff --git a/sycl/include/sycl/multi_ptr.hpp b/sycl/include/sycl/multi_ptr.hpp index 97ea1d7b3e864..400fccf3fb7d3 100644 --- a/sycl/include/sycl/multi_ptr.hpp +++ b/sycl/include/sycl/multi_ptr.hpp @@ -825,7 +825,11 @@ class multi_ptr { : multi_ptr(Accessor.get_pointer()) {} // Only if Space == local_space || generic_space - template + template < + int dimensions, access::address_space _Space = Space, + typename = typename std::enable_if_t< + _Space == Space && (Space == access::address_space::generic_space || + Space == access::address_space::local_space)>> multi_ptr(local_accessor Accessor) : multi_ptr(Accessor.get_pointer()) {} diff --git a/sycl/test-e2e/multi_ptr/prefetch.cpp b/sycl/test-e2e/multi_ptr/prefetch.cpp new file mode 100644 index 0000000000000..c169aac4ff47e --- /dev/null +++ b/sycl/test-e2e/multi_ptr/prefetch.cpp @@ -0,0 +1,245 @@ +// RUN: %{build} -o %t.out +// RUN: %{run} %t.out + +//===------------------------- prefetch.cpp --------------------------------===// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#include + +#include +#include + +using namespace sycl; + +template +void testPrefetchWithDecoration() { + constexpr size_t Size = 1024; + std::vector HostData(Size); + for (size_t i = 0; i < Size; ++i) { + HostData[i] = static_cast(i); + } + + queue Q; + buffer Buf(HostData.data(), range<1>(Size)); + + Q.submit([&](handler &CGH) { + auto Acc = Buf.get_access(CGH); + + CGH.parallel_for(range<1>(Size), [=](id<1> Index) { + auto Ptr = Acc.template get_multi_ptr(); + + // Test prefetch with different element counts + if (Index[0] == 0) { + Ptr.prefetch(1); + Ptr.prefetch(16); + Ptr.prefetch(64); + Ptr.prefetch(256); + } + + // Test prefetch at different offsets + if (Index[0] < Size - 100) { + auto OffsetPtr = Ptr + Index[0]; + OffsetPtr.prefetch(10); + } + + // Actual computation to ensure prefetch is useful + int Sum = 0; + for (size_t i = 0; i < 10 && Index[0] + i < Size; ++i) { + Sum += Ptr[Index[0] + i]; + } + Acc[Index] = Sum; + }); + }); + + Q.wait(); + + // Verify results + auto HostAcc = Buf.get_host_access(); + for (size_t i = 0; i < Size; ++i) { + int Expected = 0; + for (size_t j = 0; j < 10 && i + j < Size; ++j) { + Expected += static_cast(i + j); + } + assert(HostAcc[i] == Expected && "Prefetch test failed"); + } +} + +void testPrefetchWithGlobalPointer() { + constexpr size_t Size = 512; + std::vector HostData(Size); + for (size_t i = 0; i < Size; ++i) { + HostData[i] = static_cast(i) * 0.5f; + } + + queue Q; + buffer Buf(HostData.data(), range<1>(Size)); + + Q.submit([&](handler &CGH) { + auto Acc = Buf.get_access(CGH); + + CGH.parallel_for(range<1>(Size), [=](id<1> Index) { + using global_ptr = multi_ptr; + + global_ptr Ptr = address_space_cast(&Acc[0]); + + // Prefetch future data + if (Index[0] < Size - 50) { + auto FuturePtr = Ptr + Index[0] + 10; + FuturePtr.prefetch(20); + } + + // Process data + float Sum = 0.0f; + for (size_t i = 0; i < 5 && Index[0] + i < Size; ++i) { + Sum += Ptr[Index[0] + i]; + } + Acc[Index] = Sum; + }); + }); + + Q.wait(); + + // Verify results + auto HostAcc = Buf.get_host_access(); + for (size_t i = 0; i < Size; ++i) { + float Expected = 0.0f; + for (size_t j = 0; j < 5 && i + j < Size; ++j) { + Expected += static_cast(i + j) * 0.5f; + } + assert(std::abs(HostAcc[i] - Expected) < 0.001f && "Global prefetch test failed"); + } +} + +void testPrefetchWithLargeData() { + constexpr size_t Size = 4096; + std::vector HostData(Size); + for (size_t i = 0; i < Size; ++i) { + HostData[i] = static_cast(i); + } + + queue Q; + buffer Buf(HostData.data(), range<1>(Size)); + + Q.submit([&](handler &CGH) { + auto Acc = Buf.get_access(CGH); + + CGH.parallel_for(range<1>(Size / 8), [=](id<1> Index) { + auto Ptr = Acc.template get_multi_ptr(); + size_t BaseIndex = Index[0] * 8; + + // Prefetch a chunk of data + auto ChunkPtr = Ptr + BaseIndex; + ChunkPtr.prefetch(64); + + // Process the prefetched chunk + double Sum = 0.0; + for (size_t i = 0; i < 8; ++i) { + Sum += ChunkPtr[i]; + } + // Just to use the Sum (avoid optimization removal) + if (Sum < 0) { + ChunkPtr[0] = Sum; + } + }); + }); + + Q.wait(); +} + +void testPrefetchAtBoundaries() { + constexpr size_t Size = 256; + std::vector HostData(Size, 42); + + queue Q; + buffer Buf(HostData.data(), range<1>(Size)); + + Q.submit([&](handler &CGH) { + auto Acc = Buf.get_access(CGH); + + CGH.single_task([=]() { + auto Ptr = Acc.template get_multi_ptr(); + + // Prefetch at start + Ptr.prefetch(1); + Ptr.prefetch(10); + + // Prefetch at end (careful not to go beyond buffer) + auto EndPtr = Ptr + (Size - 10); + EndPtr.prefetch(10); + + // Prefetch zero elements (edge case) + Ptr.prefetch(0); + }); + }); + + Q.wait(); +} + +void testPrefetchWithStructs() { + struct TestStruct { + int A; + float B; + double C; + }; + + constexpr size_t Size = 128; + std::vector HostData(Size); + for (size_t i = 0; i < Size; ++i) { + HostData[i] = {static_cast(i), static_cast(i) * 1.5f, static_cast(i) * 2.5}; + } + + queue Q; + buffer Buf(HostData.data(), range<1>(Size)); + + Q.submit([&](handler &CGH) { + auto Acc = Buf.get_access(CGH); + + CGH.parallel_for(range<1>(Size / 2), [=](id<1> Index) { + auto Ptr = Acc.template get_multi_ptr(); + size_t BaseIdx = Index[0] * 2; + + // Prefetch structures + auto StructPtr = Ptr + BaseIdx; + StructPtr.prefetch(8); + + // Access the data + int Sum = 0; + for (size_t i = 0; i < 2 && BaseIdx + i < Size; ++i) { + Sum += StructPtr[i].A; + } + // Use sum to prevent optimization + if (Sum < 0) { + StructPtr[0].A = Sum; + } + }); + }); + + Q.wait(); +} + +int main() { + // Test prefetch with decorated pointers + testPrefetchWithDecoration(); + testPrefetchWithDecoration(); + + // Test prefetch with explicit global pointers + testPrefetchWithGlobalPointer(); + + // Test prefetch with large datasets + testPrefetchWithLargeData(); + + // Test prefetch at buffer boundaries + testPrefetchAtBoundaries(); + + // Test prefetch with complex data structures + testPrefetchWithStructs(); + + std::cout << "All prefetch tests passed!" << std::endl; + return 0; +} diff --git a/sycl/test/multi_ptr/restricted_conversions.cpp b/sycl/test/multi_ptr/restricted_conversions.cpp new file mode 100644 index 0000000000000..7ca01667dbfe2 --- /dev/null +++ b/sycl/test/multi_ptr/restricted_conversions.cpp @@ -0,0 +1,88 @@ +#include +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s -Xclang -verify-ignore-unexpected=note + +template +using multi_ptr_t = sycl::multi_ptr; + +using legacy_ptr = multi_ptr_t; +using non_legacy_ptr = + multi_ptr_t; +using private_ptr = multi_ptr_t; +using const_void_ptr = + multi_ptr_t; +using void_ptr = multi_ptr_t; +using global_ptr = multi_ptr_t; +using local_ptr = multi_ptr_t; +// expected-warning@+2 {{constant_space' is deprecated: sycl::access::address_space::constant_space is deprecated since SYCL 2020}} +using constant_ptr = + multi_ptr_t; +using generic_ptr = + multi_ptr_t; + +legacy_ptr leg_ptr{nullptr}; +non_legacy_ptr nonleg_ptr{nullptr}; + +// expected-error@+1 {{no matching constructor for initialization of}} +non_legacy_ptr nonleg_ptr1{leg_ptr}; +// expected-error@+1 {{no viable conversion from}} +non_legacy_ptr nonleg_ptr2 = leg_ptr; +// TODO: is constructor legal? +// expected-warning@+1 {{'operator int *' is deprecated: Conversion to pointer type is deprecated since SYCL 2020. Please use get() instead}} +legacy_ptr leg_ptr1{nonleg_ptr}; +// expected-error@+1 {{no viable conversion from 'multi_ptr_t<}} +legacy_ptr leg_ptr2 = nonleg_ptr; + +const_void_ptr const_void{nullptr}; + +// expected-error@+1 {{no matching constructor for initialization of 'void_ptr'}} +void_ptr void_ptr1{const_void}; +// expected-error@+1 {{no viable conversion from 'multi_ptr_t<}} +void_ptr void_ptr2 = const_void; + +global_ptr global{nullptr}; + +private_ptr private_ptr_instance{nullptr}; + +// expected-error@+1 {{no matching constructor for initialization of 'local_ptr'}} +local_ptr local{global}; + +// expected-error@+1 {{no viable conversion from 'multi_ptr_t<}} +local_ptr local1 = global; + +// expected-error@+1 {{no matching constructor for initialization of 'global_ptr'}} +global_ptr global1{local}; + +// expected-error@+1 {{no viable conversion from 'multi_ptr_t<}} +global_ptr global2 = local; + +// expected-error@+1 {{no matching constructor for initialization of 'local_ptr'}} +local_ptr local_from_private{private_ptr_instance}; + +// expected-error@+1 {{no viable conversion from 'multi_ptr_t<}} +local_ptr local_from_private_implicit = private_ptr_instance; + +// expected-error@+1 {{no matching constructor for initialization of 'global_ptr'}} +global_ptr global_from_private{private_ptr_instance}; + +// expected-error@+1 {{no matching constructor for initialization of 'private_ptr'}} +private_ptr private_from_local{local}; + +// expected-error@+1 {{no matching constructor for initialization of 'private_ptr'}} +private_ptr private_from_global{global}; + +// expected-warning@+1 2 {{'operator int *' is deprecated: Conversion to pointer type is deprecated since SYCL 2020. Please use get() instead.}} +bool private_equals_local = private_ptr_instance == local; + +// expected-warning@+1 2 {{'operator int *' is deprecated: Conversion to pointer type is deprecated since SYCL 2020. Please use get() instead.}} +bool private_less_than_local = private_ptr_instance < local; + diff --git a/sycl/unittests/CMakeLists.txt b/sycl/unittests/CMakeLists.txt index 0370caa29e4ee..757184e150456 100644 --- a/sycl/unittests/CMakeLists.txt +++ b/sycl/unittests/CMakeLists.txt @@ -51,6 +51,7 @@ add_subdirectory(accessor) add_subdirectory(handler) add_subdirectory(builtins) add_subdirectory(buffer/l0_specific) +add_subdirectory(multi_ptr) # Enable compression unit-tests only if zstd is present. if (LLVM_ENABLE_ZSTD) diff --git a/sycl/unittests/multi_ptr/Accessors.cpp b/sycl/unittests/multi_ptr/Accessors.cpp new file mode 100644 index 0000000000000..972c96efeebeb --- /dev/null +++ b/sycl/unittests/multi_ptr/Accessors.cpp @@ -0,0 +1,130 @@ +#include + +#include + +#include + +namespace { + +using sycl::access::address_space; +using sycl::access::decorated; + +template +using multi_ptr_t = sycl::multi_ptr; + +} // namespace + +TEST(MultiPtrAccessors, AccessorDevice) { + using rw_acc = sycl::accessor; + using atomic_acc = + sycl::accessor; + using ro_acc = + sycl::accessor; + using global_multi_ptr = + sycl::multi_ptr; + using generic_multi_ptr = + sycl::multi_ptr; + using const_multi_ptr = + sycl::multi_ptr; + + using float_global_ptr = + sycl::multi_ptr; + + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_constructible_v)); + EXPECT_FALSE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); +} + +TEST(MultiPtrAccessors, LocalAccessor) { + using local_acc = sycl::local_accessor; + using local_multi_ptr = + sycl::multi_ptr; + using generic_multi_ptr = + sycl::multi_ptr; + using const_multi_ptr = + sycl::multi_ptr; + using float_local_ptr = + sycl::multi_ptr; + + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_constructible_v< + sycl::multi_ptr, + local_acc>)); + EXPECT_FALSE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); +} + +TEST(MultiPtrAccessors, AddrSpaceCast) { + int Value = 42; + const int ConstValue = 17; + + auto PrivateUndecorated = + sycl::address_space_cast( + &Value); + auto GlobalDecorated = + sycl::address_space_cast( + &Value); + auto LocalUndecorated = + sycl::address_space_cast( + &Value); + auto GenericDecorated = + sycl::address_space_cast( + &Value); + auto ConstPrivateDecorated = + sycl::address_space_cast( + &ConstValue); + + EXPECT_TRUE((std::is_same_v< + decltype(PrivateUndecorated), + multi_ptr_t>)); + EXPECT_TRUE((std::is_same_v< + decltype(GlobalDecorated), + multi_ptr_t>)); + EXPECT_TRUE((std::is_same_v< + decltype(LocalUndecorated), + multi_ptr_t>)); + EXPECT_TRUE( + (std::is_same_v< + decltype(GenericDecorated), + multi_ptr_t>)); + EXPECT_TRUE( + (std::is_same_v>)); + + EXPECT_EQ(PrivateUndecorated.get_raw(), &Value); + EXPECT_EQ(GlobalDecorated.get_raw(), &Value); + EXPECT_EQ(LocalUndecorated.get_raw(), &Value); + EXPECT_EQ(GenericDecorated.get_raw(), &Value); + EXPECT_EQ(ConstPrivateDecorated.get_raw(), &ConstValue); + + EXPECT_FALSE((std::is_same_v< + decltype(PrivateUndecorated), + multi_ptr_t>)); + EXPECT_FALSE( + (std::is_same_v< + decltype(PrivateUndecorated), + multi_ptr_t>)); + EXPECT_FALSE( + (std::is_same_v< + decltype(ConstPrivateDecorated), + multi_ptr_t>)); + + auto NullPrivate = + sycl::address_space_cast( + static_cast(nullptr)); + auto NullGlobalConst = + sycl::address_space_cast( + static_cast(nullptr)); + + EXPECT_EQ(NullPrivate, nullptr); + EXPECT_EQ(NullGlobalConst, nullptr); +} diff --git a/sycl/unittests/multi_ptr/CMakeLists.txt b/sycl/unittests/multi_ptr/CMakeLists.txt new file mode 100644 index 0000000000000..c2f2c88c500d9 --- /dev/null +++ b/sycl/unittests/multi_ptr/CMakeLists.txt @@ -0,0 +1,9 @@ +add_sycl_unittest(MultiPtrTests OBJECT + Constructors.cpp + Operators.cpp + Conversion.cpp + Accessors.cpp + NegativeConversions.cpp + VoidSpecialization.cpp +) + diff --git a/sycl/unittests/multi_ptr/Constructors.cpp b/sycl/unittests/multi_ptr/Constructors.cpp new file mode 100644 index 0000000000000..18ac8d4c85ffb --- /dev/null +++ b/sycl/unittests/multi_ptr/Constructors.cpp @@ -0,0 +1,194 @@ +//===---------------------------- Constructors.cpp ------------===// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#include + +#include + +#include +#include + +namespace { + +using sycl::access::address_space; +using sycl::access::decorated; + +template +using private_multi_ptr = + sycl::multi_ptr; + +template +using local_multi_ptr = + sycl::multi_ptr; + +template +private_multi_ptr +makePrivatePtr(std::add_pointer_t Ptr) { + return sycl::address_space_cast( + Ptr); +} + +template void checkCopyConstruction() { + int Value = 17; + auto Original = makePrivatePtr(&Value); + private_multi_ptr Copy{Original}; + + EXPECT_EQ(Copy, Original); + EXPECT_NE(Copy, nullptr); +} + +template void checkMoveConstruction() { + int Value = 23; + auto Source = makePrivatePtr(&Value); + auto Expected = makePrivatePtr(&Value); + private_multi_ptr Moved{std::move(Source)}; + + EXPECT_EQ(Moved, Expected); + EXPECT_NE(Moved, nullptr); +} + +TEST(MultiPtrConstructors, ValidConstructorTraitsAreSatisfied) { + EXPECT_TRUE( + (std::is_default_constructible_v>)); + EXPECT_TRUE(( + std::is_default_constructible_v>)); + EXPECT_TRUE((std::is_constructible_v, + std::nullptr_t>)); + EXPECT_TRUE((std::is_constructible_v, + std::nullptr_t>)); + EXPECT_TRUE( + (std::is_copy_constructible_v>)); + EXPECT_TRUE( + (std::is_copy_constructible_v>)); + EXPECT_TRUE( + (std::is_move_constructible_v>)); + EXPECT_TRUE( + (std::is_move_constructible_v>)); + EXPECT_TRUE( + (std::is_constructible_v, + private_multi_ptr>)); + EXPECT_TRUE( + (std::is_constructible_v, + private_multi_ptr>)); + EXPECT_TRUE( + (std::is_constructible_v, + private_multi_ptr>)); + EXPECT_TRUE((std::is_constructible_v, + private_multi_ptr>)); + EXPECT_TRUE( + (std::is_constructible_v, + private_multi_ptr>)); + EXPECT_TRUE( + (std::is_constructible_v, + private_multi_ptr>)); +} + +TEST(MultiPtrConstructors, InvalidConstructorTraitsAreRejected) { + EXPECT_FALSE( + (std::is_constructible_v, + private_multi_ptr>)); + EXPECT_FALSE( + (std::is_constructible_v, + private_multi_ptr>)); + EXPECT_FALSE( + (std::is_constructible_v, + private_multi_ptr>)); + EXPECT_FALSE( + (std::is_constructible_v, + private_multi_ptr>)); + EXPECT_FALSE( + (std::is_constructible_v, + local_multi_ptr>)); + EXPECT_FALSE( + (std::is_constructible_v, + local_multi_ptr>)); +} + +TEST(MultiPtrConstructors, DefaultConstructedPointersAreNull) { + private_multi_ptr Undecorated; + private_multi_ptr Decorated; + + EXPECT_EQ(Undecorated, nullptr); + EXPECT_EQ(Decorated, nullptr); +} + +TEST(MultiPtrConstructors, NullptrConstructedPointersAreNull) { + private_multi_ptr Undecorated{nullptr}; + private_multi_ptr Decorated{nullptr}; + + EXPECT_EQ(Undecorated, nullptr); + EXPECT_EQ(Decorated, nullptr); +} + +TEST(MultiPtrConstructors, CopyConstructionPreservesPointerValue) { + checkCopyConstruction(); + checkCopyConstruction(); +} + +TEST(MultiPtrConstructors, MoveConstructionPreservesPointerValue) { + checkMoveConstruction(); + checkMoveConstruction(); +} + +TEST(MultiPtrConstructors, ConvertingConstructionAddsConstQualification) { + int Value = 31; + + private_multi_ptr MutableUndecorated = + makePrivatePtr(&Value); + private_multi_ptr MutableDecorated = + makePrivatePtr(&Value); + + private_multi_ptr ConstUndecorated{ + MutableUndecorated}; + private_multi_ptr ConstDecorated{MutableDecorated}; + auto ExpectedUndecorated = makePrivatePtr(&Value); + auto ExpectedDecorated = makePrivatePtr(&Value); + + EXPECT_EQ(ConstUndecorated, ExpectedUndecorated); + EXPECT_EQ(ConstDecorated, ExpectedDecorated); +} + +TEST(MultiPtrConstructors, + ConvertingConstructionBetweenDecorationModesPreservesPointerValue) { + int Value = 47; + + private_multi_ptr DecoratedPtr = + makePrivatePtr(&Value); + private_multi_ptr UndecoratedPtr = + makePrivatePtr(&Value); + + private_multi_ptr FromDecorated{DecoratedPtr}; + private_multi_ptr FromUndecorated{UndecoratedPtr}; + auto ExpectedUndecorated = makePrivatePtr(&Value); + auto ExpectedDecorated = makePrivatePtr(&Value); + + EXPECT_EQ(FromDecorated, ExpectedUndecorated); + EXPECT_EQ(FromUndecorated, ExpectedDecorated); +} + +TEST(MultiPtrConstructors, + ConvertingConstructionCanAddConstAndChangeDecorationTogether) { + int Value = 59; + + private_multi_ptr MutableUndecorated = + makePrivatePtr(&Value); + private_multi_ptr MutableDecorated = + makePrivatePtr(&Value); + + private_multi_ptr ConstDecorated{ + MutableUndecorated}; + private_multi_ptr ConstUndecorated{ + MutableDecorated}; + auto ExpectedDecorated = makePrivatePtr(&Value); + auto ExpectedUndecorated = makePrivatePtr(&Value); + + EXPECT_EQ(ConstDecorated, ExpectedDecorated); + EXPECT_EQ(ConstUndecorated, ExpectedUndecorated); +} + +} // namespace diff --git a/sycl/unittests/multi_ptr/Conversion.cpp b/sycl/unittests/multi_ptr/Conversion.cpp new file mode 100644 index 0000000000000..fe9cf71411ea8 --- /dev/null +++ b/sycl/unittests/multi_ptr/Conversion.cpp @@ -0,0 +1,432 @@ +#include + +#include + +#include +#include +#include + +namespace { + +using sycl::access::address_space; +using sycl::access::decorated; + +struct Record { + int Value; + int Tag; +}; + +template +using multi_ptr_t = sycl::multi_ptr; + +template +using access_element_t = + std::conditional_t; + +template +using access_storage_t = std::remove_const_t>; + +template +multi_ptr_t +makePtr(std::add_pointer_t Ptr) { + if constexpr (IsDecorated == decorated::legacy) { + return multi_ptr_t{Ptr}; + } else { + return sycl::address_space_cast(Ptr); + } +} + +template constexpr decorated toggledDecoration() { + static_assert(IsDecorated != decorated::legacy); + return IsDecorated == decorated::no ? decorated::yes : decorated::no; +} + +template +void checkPointerAccessAndUnderlyingPointers() { + using element_t = access_element_t; + using ptr_t = multi_ptr_t; + + access_storage_t Values[] = {{11, 111}, {22, 222}}; + auto *RawPtr = static_cast>(&Values[0]); + ptr_t Ptr = makePtr(RawPtr); + + EXPECT_EQ(Ptr->Value, 11); + EXPECT_EQ(Ptr->Tag, 111); + EXPECT_EQ(Ptr[1].Value, 22); + EXPECT_EQ(Ptr[1].Tag, 222); + + decltype(Ptr.get()) PointerValue = Ptr; + EXPECT_EQ(PointerValue, Ptr.get()); + EXPECT_EQ(Ptr.get_raw(), RawPtr); + + if constexpr (IsDecorated == decorated::legacy) { + EXPECT_EQ(Ptr.get_decorated(), Ptr.get()); + } else { + auto DecoratedPtr = makePtr(RawPtr); + EXPECT_EQ(Ptr.get_decorated(), DecoratedPtr.get_decorated()); + } +} + +template +void checkGenericExplicitMutableConversions() { + using source_t = + multi_ptr_t; + using private_t = + multi_ptr_t; + using global_t = + multi_ptr_t; + using local_t = + multi_ptr_t; + using const_private_t = + multi_ptr_t; + using const_global_t = + multi_ptr_t; + using const_local_t = + multi_ptr_t; + + Record Value{31, 131}; + source_t Source = + makePtr(&Value); + + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + + private_t Private{Source}; + global_t Global{Source}; + local_t Local{Source}; + const_private_t ConstPrivate{Source}; + const_global_t ConstGlobal{Source}; + const_local_t ConstLocal{Source}; + + EXPECT_EQ(Private.get_raw(), &Value); + EXPECT_EQ(Global.get_raw(), &Value); + EXPECT_EQ(Local.get_raw(), &Value); + EXPECT_EQ(ConstPrivate.get_raw(), &Value); + EXPECT_EQ(ConstGlobal.get_raw(), &Value); + EXPECT_EQ(ConstLocal.get_raw(), &Value); +} + +template +void checkGenericExplicitConstConversions() { + using source_t = + multi_ptr_t; + using private_t = + multi_ptr_t; + using global_t = + multi_ptr_t; + using local_t = + multi_ptr_t; + + const Record Value{32, 132}; + source_t Source = makePtr(&Value); + + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + + private_t Private{Source}; + global_t Global{Source}; + local_t Local{Source}; + + EXPECT_EQ(Private.get_raw(), &Value); + EXPECT_EQ(Global.get_raw(), &Value); + EXPECT_EQ(Local.get_raw(), &Value); +} + +template +void checkNonLegacyMutableSameSpaceConversions() { + static_assert(SourceDecoration != decorated::legacy); + + using source_t = multi_ptr_t; + using typed_no_t = multi_ptr_t; + using typed_yes_t = multi_ptr_t; + using typed_legacy_t = multi_ptr_t; + using toggled_t = + multi_ptr_t()>; + using void_same_t = multi_ptr_t; + using void_toggled_t = + multi_ptr_t()>; + using void_no_t = multi_ptr_t; + using void_yes_t = multi_ptr_t; + using void_legacy_t = multi_ptr_t; + using const_no_t = multi_ptr_t; + using const_yes_t = multi_ptr_t; + using const_legacy_t = multi_ptr_t; + + Record Value{41, 141}; + source_t Source = makePtr(&Value); + + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + + toggled_t Toggled = Source; + void_no_t VoidNo = Source; + void_yes_t VoidYes = Source; + void_legacy_t VoidLegacy = Source; + const_no_t ConstNo = Source; + const_yes_t ConstYes = Source; + const_legacy_t ConstLegacy = Source; + void_same_t VoidSame = Source; + void_toggled_t VoidToggled = VoidSame; + typed_no_t FromVoidNo = static_cast(VoidNo); + typed_yes_t FromVoidYes = static_cast(VoidYes); + typed_legacy_t FromVoidLegacy = static_cast(VoidLegacy); + toggled_t FromVoidToggled = static_cast(VoidToggled); + source_t FromVoid = static_cast(VoidSame); + + EXPECT_EQ(Toggled.get_raw(), &Value); + EXPECT_EQ(ConstNo.get_raw(), &Value); + EXPECT_EQ(ConstYes.get_raw(), &Value); + EXPECT_EQ(ConstLegacy.get_raw(), &Value); + EXPECT_EQ(FromVoidNo.get_raw(), &Value); + EXPECT_EQ(FromVoidYes.get_raw(), &Value); + EXPECT_EQ(FromVoidLegacy.get_raw(), &Value); + EXPECT_EQ(FromVoidToggled.get_raw(), &Value); + EXPECT_EQ(FromVoid.get_raw(), &Value); +} + +template +void checkNonLegacyConstSameSpaceConversions() { + static_assert(SourceDecoration != decorated::legacy); + + using source_t = multi_ptr_t; + using typed_no_t = multi_ptr_t; + using typed_yes_t = multi_ptr_t; + using typed_legacy_t = multi_ptr_t; + using toggled_t = + multi_ptr_t()>; + using const_void_same_t = multi_ptr_t; + using const_void_toggled_t = + multi_ptr_t()>; + using const_void_no_t = multi_ptr_t; + using const_void_yes_t = multi_ptr_t; + using const_void_legacy_t = multi_ptr_t; + + const Record Value{42, 142}; + source_t Source = makePtr(&Value); + + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + + toggled_t Toggled = Source; + const_void_no_t ConstVoidNo = Source; + const_void_yes_t ConstVoidYes = Source; + const_void_legacy_t ConstVoidLegacy = Source; + const_void_same_t ConstVoidSame = Source; + const_void_toggled_t ConstVoidToggled = ConstVoidSame; + typed_no_t FromConstVoidNo = static_cast(ConstVoidNo); + typed_yes_t FromConstVoidYes = static_cast(ConstVoidYes); + typed_legacy_t FromConstVoidLegacy = static_cast(ConstVoidLegacy); + toggled_t FromConstVoidToggled = static_cast(ConstVoidToggled); + source_t FromConstVoid = static_cast(ConstVoidSame); + + EXPECT_EQ(Toggled.get_raw(), &Value); + EXPECT_EQ(FromConstVoidNo.get_raw(), &Value); + EXPECT_EQ(FromConstVoidYes.get_raw(), &Value); + EXPECT_EQ(FromConstVoidLegacy.get_raw(), &Value); + EXPECT_EQ(FromConstVoidToggled.get_raw(), &Value); + EXPECT_EQ(FromConstVoid.get_raw(), &Value); +} + +template void checkLegacyConversions() { + using mutable_t = multi_ptr_t; + using const_t = multi_ptr_t; + using void_t = multi_ptr_t; + using const_void_t = multi_ptr_t; + + Record MutableValue{51, 151}; + const Record ConstValue{52, 152}; + + mutable_t Mutable = makePtr(&MutableValue); + const_t Const = makePtr(&ConstValue); + + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + + void_t VoidPtr = Mutable; + const_t ConstFromMutable = Mutable; + const_void_t ConstVoidPtr = Const; + const_void_t ConstVoidFromVoid = VoidPtr; + mutable_t MutableRoundTrip = static_cast(VoidPtr); + const_t ConstRoundTrip = static_cast(ConstVoidPtr); + + EXPECT_EQ(VoidPtr.get_raw(), static_cast(&MutableValue)); + EXPECT_EQ(ConstFromMutable.get_raw(), &MutableValue); + EXPECT_EQ(ConstVoidPtr.get_raw(), static_cast(&ConstValue)); + EXPECT_EQ(ConstVoidFromVoid.get_raw(), static_cast(&MutableValue)); + EXPECT_EQ(MutableRoundTrip.get_raw(), &MutableValue); + EXPECT_EQ(ConstRoundTrip.get_raw(), &ConstValue); +} + +template void checkFactoryFunctions() { + using ptr_t = multi_ptr_t; + + int Value = 77; + auto CastResult = + sycl::address_space_cast( + &Value); + using cast_result_t = decltype(CastResult); + EXPECT_TRUE((std::is_same_v)); + EXPECT_EQ(CastResult.get_raw(), &Value); + + auto NullCast = + sycl::address_space_cast( + static_cast(nullptr)); + EXPECT_EQ(NullCast, nullptr); + + auto MakeResult = + sycl::make_ptr( + CastResult.get()); + using make_result_t = decltype(MakeResult); + EXPECT_TRUE((std::is_same_v)); + EXPECT_EQ(MakeResult.get_raw(), &Value); + + using underlying_pointer_t = decltype(CastResult.get()); + underlying_pointer_t NullPointer = nullptr; + auto NullMake = + sycl::make_ptr( + NullPointer); + EXPECT_EQ(NullMake, nullptr); +} + +void checkDefaultLegacyMakePtr() { + int Value = 88; + auto LegacyPtr = + sycl::make_ptr(&Value); + EXPECT_TRUE((std::is_same_v>)); + EXPECT_EQ(LegacyPtr.get_raw(), &Value); +} + +TEST(MultiPtrConversion, + PointerAccessPointerConversionAndGetDecoratedCoverAllEnumValues) { + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); +} + +TEST(MultiPtrConversion, + GenericSpaceExplicitConversionsSupportAllNonLegacyDecorationPairs) { + checkGenericExplicitMutableConversions(); + checkGenericExplicitMutableConversions(); + checkGenericExplicitMutableConversions(); + checkGenericExplicitMutableConversions(); + + checkGenericExplicitConstConversions(); + checkGenericExplicitConstConversions(); + checkGenericExplicitConstConversions(); + checkGenericExplicitConstConversions(); +} + +TEST(MultiPtrConversion, + NonLegacySameSpaceConversionsCoverVoidConstAndDecorationChanges) { + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); +} + +TEST(MultiPtrConversion, LegacyConversionsCoverAllAddressSpaces) { + checkLegacyConversions(); + checkLegacyConversions(); + checkLegacyConversions(); + checkLegacyConversions(); + checkLegacyConversions(); +} + +TEST(MultiPtrConversion, FactoryFunctionsPreservePrivatePointersAndNullptr) { + checkFactoryFunctions(); + checkFactoryFunctions(); + checkFactoryFunctions(); + checkDefaultLegacyMakePtr(); +} + +} // namespace + diff --git a/sycl/unittests/multi_ptr/NegativeConversions.cpp b/sycl/unittests/multi_ptr/NegativeConversions.cpp new file mode 100644 index 0000000000000..6f2b42047368d --- /dev/null +++ b/sycl/unittests/multi_ptr/NegativeConversions.cpp @@ -0,0 +1,143 @@ +//===------------------ NegativeConversions.cpp ---------------===// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#include + +#include + +#include + +namespace { + +using sycl::access::address_space; +using sycl::access::decorated; + +template +using multi_ptr_t = sycl::multi_ptr; + +TEST(MultiPtrNegativeConversions, CannotRemoveConstQualification) { + using const_ptr = + multi_ptr_t; + using mutable_ptr = + multi_ptr_t; + + EXPECT_FALSE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_assignable_v)); +} + +TEST(MultiPtrNegativeConversions, CannotConvertBetweenIncompatibleTypes) { + using int_ptr = multi_ptr_t; + using float_ptr = + multi_ptr_t; + using double_ptr = + multi_ptr_t; + + EXPECT_FALSE((std::is_constructible_v)); + EXPECT_FALSE((std::is_constructible_v)); + EXPECT_FALSE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); +} + + +TEST(MultiPtrNegativeConversions, CannotConvertToGenericFromNonGeneric) { + using private_ptr = + multi_ptr_t; + using local_ptr = multi_ptr_t; + using global_ptr = + multi_ptr_t; + using generic_ptr = + multi_ptr_t; + + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); +} + +TEST(MultiPtrNegativeConversions, + CannotImplicitlyConvertFromGenericToSpecificSpace) { + using generic_ptr = + multi_ptr_t; + using private_ptr = + multi_ptr_t; + using local_ptr = multi_ptr_t; + using global_ptr = + multi_ptr_t; + + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); +} + +TEST(MultiPtrNegativeConversions, CannotConvertGenericToConstantSpace) { + using generic_ptr = + multi_ptr_t; + using constant_ptr = + multi_ptr_t; + + generic_ptr gen_ptr; + constant_ptr const_ptr; + + generic_ptr gen_ptr2{const_ptr}; + constant_ptr const_ptr2{gen_ptr}; + + EXPECT_EQ(gen_ptr2.get(), const_ptr.get()); + EXPECT_EQ(const_ptr2.get(), gen_ptr.get()); +} + +TEST(MultiPtrNegativeConversions, CannotConvertVoidToTypedWithoutExplicitCast) { + using void_ptr = + multi_ptr_t; + using int_ptr = multi_ptr_t; + + EXPECT_FALSE((std::is_convertible_v)); +} + +TEST(MultiPtrNegativeConversions, CannotConstructMutableFromConstTypedVoid) { + using const_void_ptr = + multi_ptr_t; + using int_ptr = multi_ptr_t; + + EXPECT_FALSE((std::is_constructible_v)); +} + +TEST(MultiPtrNegativeConversions, CannotAssignBetweenIncompatibleSpaces) { + using private_ptr = + multi_ptr_t; + using local_ptr = multi_ptr_t; + + EXPECT_FALSE((std::is_assignable_v)); + EXPECT_FALSE((std::is_assignable_v)); +} + +TEST(MultiPtrNegativeConversions, GenericCannotAssignFromConstantSpace) { + using generic_ptr = + multi_ptr_t; + using constant_ptr = + multi_ptr_t; + + EXPECT_FALSE((std::is_assignable_v)); +} + +TEST(MultiPtrNegativeConversions, ConstantSpaceNotSupportedForVoidInNonLegacy) { + using const_void_constant_legacy = + multi_ptr_t; + EXPECT_TRUE((std::is_default_constructible_v)); +} + + +TEST(MultiPtrNegativeConversions, CannotComparePointersOfDifferentTypes) { + using int_ptr = multi_ptr_t; + using float_ptr = + multi_ptr_t; + + EXPECT_FALSE((std::is_invocable_v, int_ptr, float_ptr>)); +} + +} // namespace diff --git a/sycl/unittests/multi_ptr/Operators.cpp b/sycl/unittests/multi_ptr/Operators.cpp new file mode 100644 index 0000000000000..2468e9e262ef7 --- /dev/null +++ b/sycl/unittests/multi_ptr/Operators.cpp @@ -0,0 +1,247 @@ +//===---------------------------- Operators.cpp ------------===// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#include + +#include + +#include +#include +#include + +namespace { + +using sycl::access::address_space; +using sycl::access::decorated; + +template +using private_multi_ptr = + sycl::multi_ptr; + +template +private_multi_ptr +makePrivatePtr(std::add_pointer_t Ptr) { + return sycl::address_space_cast( + Ptr); +} + +struct Record { + int Value; + int Tag; +}; + +template void checkOperatorTraits() { + using ptr_t = private_multi_ptr; + using diff_t = typename ptr_t::difference_type; + + EXPECT_TRUE( + (std::is_same_v()), + typename ptr_t::reference>)); + EXPECT_TRUE((std::is_same_v()[diff_t{0}]), + typename ptr_t::reference>)); + EXPECT_TRUE((std::is_same_v().operator->()), + typename ptr_t::pointer>)); + + EXPECT_TRUE((std::is_same_v()), ptr_t &>)); + EXPECT_TRUE((std::is_same_v()++), ptr_t>)); + EXPECT_TRUE((std::is_same_v()), ptr_t &>)); + EXPECT_TRUE((std::is_same_v()--), ptr_t>)); + EXPECT_TRUE((std::is_same_v() += diff_t{0}), + ptr_t &>)); + EXPECT_TRUE((std::is_same_v() -= diff_t{0}), + ptr_t &>)); + EXPECT_TRUE((std::is_same_v() + diff_t{0}), + ptr_t>)); + EXPECT_TRUE((std::is_same_v() - diff_t{0}), + ptr_t>)); + + EXPECT_TRUE((std::is_same_v() == + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() != + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() < + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() > + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() <= + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() >= + std::declval()), + bool>)); + + EXPECT_TRUE((std::is_same_v() == nullptr), + bool>)); + EXPECT_TRUE((std::is_same_v() != nullptr), + bool>)); + EXPECT_TRUE((std::is_same_v() < nullptr), + bool>)); + EXPECT_TRUE((std::is_same_v() > nullptr), + bool>)); + EXPECT_TRUE((std::is_same_v() <= nullptr), + bool>)); + EXPECT_TRUE((std::is_same_v() >= nullptr), + bool>)); + EXPECT_TRUE((std::is_same_v()), + bool>)); + EXPECT_TRUE((std::is_same_v()), + bool>)); + EXPECT_TRUE((std::is_same_v()), + bool>)); + EXPECT_TRUE((std::is_same_v std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v()), + bool>)); + EXPECT_TRUE((std::is_same_v= std::declval()), + bool>)); +} + +template void checkAccessOperators() { + Record Values[] = {{10, 100}, {20, 200}, {30, 300}}; + auto Ptr = makePrivatePtr(&Values[0]); + + EXPECT_EQ((*Ptr).Value, 10); + EXPECT_EQ((*Ptr).Tag, 100); + EXPECT_EQ(Ptr->Value, 10); + EXPECT_EQ(Ptr->Tag, 100); + EXPECT_EQ(Ptr[1].Value, 20); + EXPECT_EQ(Ptr[1].Tag, 200); + EXPECT_EQ(Ptr[2].Value, 30); + EXPECT_EQ(Ptr[2].Tag, 300); +} + +template void checkArithmeticOperators() { + Record Values[] = {{1, 10}, {2, 20}, {3, 30}, {4, 40}}; + using ptr_t = private_multi_ptr; + using diff_t = typename ptr_t::difference_type; + + auto Ptr = makePrivatePtr(&Values[0]); + auto Expected0 = makePrivatePtr(&Values[0]); + auto Expected1 = makePrivatePtr(&Values[1]); + auto Expected2 = makePrivatePtr(&Values[2]); + auto Expected3 = makePrivatePtr(&Values[3]); + + ptr_t &PrefixIncrement = ++Ptr; + EXPECT_EQ(std::addressof(PrefixIncrement), std::addressof(Ptr)); + EXPECT_EQ(Ptr, Expected1); + + ptr_t PostfixIncrement = Ptr++; + EXPECT_EQ(PostfixIncrement, Expected1); + EXPECT_EQ(Ptr, Expected2); + + ptr_t &PrefixDecrement = --Ptr; + EXPECT_EQ(std::addressof(PrefixDecrement), std::addressof(Ptr)); + EXPECT_EQ(Ptr, Expected1); + + ptr_t PostfixDecrement = Ptr--; + EXPECT_EQ(PostfixDecrement, Expected1); + EXPECT_EQ(Ptr, Expected0); + + ptr_t &PlusAssign = (Ptr += diff_t{3}); + EXPECT_EQ(std::addressof(PlusAssign), std::addressof(Ptr)); + EXPECT_EQ(Ptr, Expected3); + + ptr_t &MinusAssign = (Ptr -= diff_t{2}); + EXPECT_EQ(std::addressof(MinusAssign), std::addressof(Ptr)); + EXPECT_EQ(Ptr, Expected1); + + ptr_t PlusResult = Ptr + diff_t{2}; + EXPECT_EQ(PlusResult, Expected3); + EXPECT_EQ(Ptr, Expected1); + + ptr_t MinusResult = Ptr - diff_t{1}; + EXPECT_EQ(MinusResult, Expected0); + EXPECT_EQ(Ptr, Expected1); +} + +template void checkRelationalOperators() { + int Values[] = {5, 6, 7}; + + auto First = makePrivatePtr(&Values[0]); + auto SameAsFirst = makePrivatePtr(&Values[0]); + auto Second = makePrivatePtr(&Values[1]); + + EXPECT_TRUE(First == SameAsFirst); + EXPECT_FALSE(First != SameAsFirst); + EXPECT_FALSE(First < SameAsFirst); + EXPECT_FALSE(First > SameAsFirst); + EXPECT_TRUE(First <= SameAsFirst); + EXPECT_TRUE(First >= SameAsFirst); + + EXPECT_FALSE(First == Second); + EXPECT_TRUE(First != Second); + EXPECT_TRUE(First < Second); + EXPECT_FALSE(First > Second); + EXPECT_TRUE(First <= Second); + EXPECT_FALSE(First >= Second); + + EXPECT_FALSE(Second == First); + EXPECT_TRUE(Second != First); + EXPECT_FALSE(Second < First); + EXPECT_TRUE(Second > First); + EXPECT_FALSE(Second <= First); + EXPECT_TRUE(Second >= First); +} + +template void checkNullptrComparisonOperators() { + int Value = 9; + auto NullPtr = private_multi_ptr{nullptr}; + auto NonNullPtr = makePrivatePtr(&Value); + + EXPECT_TRUE(NullPtr == nullptr); + EXPECT_FALSE(NullPtr != nullptr); + EXPECT_FALSE(NullPtr < nullptr); + EXPECT_FALSE(NullPtr > nullptr); + EXPECT_TRUE(NullPtr <= nullptr); + EXPECT_TRUE(NullPtr >= nullptr); + + EXPECT_TRUE(nullptr == NullPtr); + EXPECT_FALSE(nullptr != NullPtr); + EXPECT_FALSE(nullptr < NullPtr); + EXPECT_FALSE(nullptr > NullPtr); + EXPECT_TRUE(nullptr <= NullPtr); + EXPECT_TRUE(nullptr >= NullPtr); + + EXPECT_FALSE(NonNullPtr == nullptr); + EXPECT_TRUE(NonNullPtr != nullptr); + EXPECT_FALSE(nullptr == NonNullPtr); + EXPECT_TRUE(nullptr != NonNullPtr); +} + +TEST(MultiPtrOperators, OperatorTraitsAreSatisfied) { + checkOperatorTraits(); + checkOperatorTraits(); +} + +TEST(MultiPtrOperators, AccessOperatorsProvidePointerLikeSemantics) { + checkAccessOperators(); + checkAccessOperators(); +} + +TEST(MultiPtrOperators, ArithmeticOperatorsProvideRandomAccessTraversal) { + checkArithmeticOperators(); + checkArithmeticOperators(); +} + +TEST(MultiPtrOperators, RelationalOperatorsCompareUnderlyingLocations) { + checkRelationalOperators(); + checkRelationalOperators(); +} + +TEST(MultiPtrOperators, NullptrComparisonOperatorsHandleNullValuesBothWays) { + checkNullptrComparisonOperators(); + checkNullptrComparisonOperators(); +} + +} // namespace + + diff --git a/sycl/unittests/multi_ptr/VoidSpecialization.cpp b/sycl/unittests/multi_ptr/VoidSpecialization.cpp new file mode 100644 index 0000000000000..3f070d8196e13 --- /dev/null +++ b/sycl/unittests/multi_ptr/VoidSpecialization.cpp @@ -0,0 +1,358 @@ +//===------------------- VoidSpecialization.cpp ----------------===// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#include + +#include + +#include + +namespace { + +using sycl::access::address_space; +using sycl::access::decorated; + +template +using multi_ptr_t = sycl::multi_ptr; + +template +multi_ptr_t +makePtr(std::add_pointer_t Ptr) { + if constexpr (IsDecorated == decorated::legacy) { + return multi_ptr_t{Ptr}; + } else { + return sycl::address_space_cast(Ptr); + } +} + +struct TestData { + int Value; + float Score; +}; + +template void checkVoidPointerConstruction() { + using void_ptr = multi_ptr_t; + using const_void_ptr = + multi_ptr_t; + + void_ptr DefaultVoid; + const_void_ptr DefaultConstVoid; + EXPECT_EQ(DefaultVoid, nullptr); + EXPECT_EQ(DefaultConstVoid, nullptr); + + void_ptr NullVoid{nullptr}; + const_void_ptr NullConstVoid{nullptr}; + EXPECT_EQ(NullVoid, nullptr); + EXPECT_EQ(NullConstVoid, nullptr); + + TestData Data{42, 3.14f}; + auto TypedPtr = + makePtr(&Data); + + void_ptr VoidFromTyped = TypedPtr; + EXPECT_NE(VoidFromTyped, nullptr); + // Convert back to typed to verify it points to the right data + auto TypedCheck = static_cast< + multi_ptr_t>( + VoidFromTyped); + EXPECT_EQ(TypedCheck->Value, 42); + + // For const void, need to go through const typed pointer first + auto ConstTypedPtr = static_cast< + multi_ptr_t>( + TypedPtr); + const_void_ptr ConstVoidFromTyped = ConstTypedPtr; + EXPECT_NE(ConstVoidFromTyped, nullptr); + // Convert back to const typed to verify + auto ConstTypedCheck = static_cast< + multi_ptr_t>( + ConstVoidFromTyped); + EXPECT_EQ(ConstTypedCheck->Value, 42); +} + +template void checkVoidPointerCopyAndMove() { + using void_ptr = multi_ptr_t; + + TestData Data{17, 2.71f}; + auto TypedPtr = + makePtr(&Data); + void_ptr VoidPtr = TypedPtr; + + void_ptr CopiedVoid{VoidPtr}; + EXPECT_EQ(CopiedVoid, VoidPtr); + // Verify by converting back to typed + auto TypedCheck = static_cast< + multi_ptr_t>( + CopiedVoid); + EXPECT_EQ(TypedCheck->Value, 17); + + void_ptr MovedVoid{std::move(VoidPtr)}; + auto TypedCheck2 = static_cast< + multi_ptr_t>( + MovedVoid); + EXPECT_EQ(TypedCheck2->Value, 17); +} + +template void checkVoidPointerAssignment() { + using void_ptr = multi_ptr_t; + + TestData Data1{10, 1.0f}; + TestData Data2{20, 2.0f}; + + auto TypedPtr1 = + makePtr(&Data1); + auto TypedPtr2 = + makePtr(&Data2); + + void_ptr VoidPtr1 = TypedPtr1; + void_ptr VoidPtr2 = TypedPtr2; + + EXPECT_NE(VoidPtr1, VoidPtr2); + + VoidPtr1 = VoidPtr2; + EXPECT_EQ(VoidPtr1, VoidPtr2); + // Verify by converting back + auto TypedCheck = static_cast< + multi_ptr_t>( + VoidPtr1); + EXPECT_EQ(TypedCheck->Value, 20); + + VoidPtr1 = nullptr; + EXPECT_EQ(VoidPtr1, nullptr); + EXPECT_NE(VoidPtr2, nullptr); +} + +template +void checkVoidPointerExplicitConversionToTyped() { + using void_ptr = multi_ptr_t; + using int_ptr = multi_ptr_t; + using data_ptr = + multi_ptr_t; + + TestData Data{99, 9.9f}; + auto TypedPtr = + makePtr(&Data); + void_ptr VoidPtr = TypedPtr; + + data_ptr RoundTrip = static_cast(VoidPtr); + EXPECT_EQ(RoundTrip->Value, 99); + EXPECT_EQ(RoundTrip->Score, 9.9f); + + int Value = 123; + auto IntTypedPtr = + makePtr(&Value); + void_ptr IntVoidPtr = IntTypedPtr; + int_ptr IntRoundTrip = static_cast(IntVoidPtr); + EXPECT_EQ(*IntRoundTrip, 123); +} + +template void checkConstVoidPointerConstConversions() { + using const_void_ptr = + multi_ptr_t; + using const_int_ptr = + multi_ptr_t; + + int Value = 456; + auto MutablePtr = + makePtr(&Value); + + // Convert mutable to const typed first, then to const void + const_int_ptr ConstIntPtr1 = MutablePtr; + const_void_ptr ConstVoidPtr = ConstIntPtr1; + EXPECT_NE(ConstVoidPtr.get(), nullptr); + + const_int_ptr ConstIntPtr = static_cast(ConstVoidPtr); + EXPECT_EQ(*ConstIntPtr, 456); + + const int ConstValue = 789; + auto ConstTypedPtr = + makePtr( + &ConstValue); + const_void_ptr ConstVoidPtr2 = ConstTypedPtr; + const_int_ptr ConstIntPtr2 = static_cast(ConstVoidPtr2); + EXPECT_EQ(*ConstIntPtr2, 789); +} + +template void checkVoidPointerComparisons() { + using void_ptr = multi_ptr_t; + + TestData Data1{1, 1.0f}; + TestData Data2{2, 2.0f}; + + auto TypedPtr1 = + makePtr(&Data1); + auto TypedPtr2 = + makePtr(&Data2); + + void_ptr VoidPtr1 = TypedPtr1; + void_ptr VoidPtr2 = TypedPtr2; + void_ptr VoidPtr1Copy = TypedPtr1; + + EXPECT_TRUE(VoidPtr1 == VoidPtr1Copy); + EXPECT_FALSE(VoidPtr1 != VoidPtr1Copy); + EXPECT_TRUE(VoidPtr1 != VoidPtr2); + EXPECT_FALSE(VoidPtr1 == VoidPtr2); + + void_ptr NullPtr = nullptr; + EXPECT_TRUE(NullPtr == nullptr); + EXPECT_FALSE(NullPtr != nullptr); + EXPECT_TRUE(nullptr == NullPtr); + EXPECT_FALSE(nullptr != NullPtr); + + EXPECT_FALSE(VoidPtr1 == nullptr); + EXPECT_TRUE(VoidPtr1 != nullptr); +} + +template void checkVoidPointerGetMethods() { + using void_ptr = multi_ptr_t; + + TestData Data{50, 5.0f}; + auto TypedPtr = + makePtr(&Data); + void_ptr VoidPtr = TypedPtr; + + // Check get() returns non-null pointer + EXPECT_NE(VoidPtr.get(), nullptr); + + if constexpr (IsDecorated == decorated::legacy) { + // Legacy has get_raw() method + EXPECT_EQ(VoidPtr.get_raw(), static_cast(&Data)); + EXPECT_EQ(VoidPtr.get(), VoidPtr.get_decorated()); + } else { + // Non-legacy void specialization doesn't have get_raw() + // Just verify get() works + EXPECT_NE(VoidPtr.get(), nullptr); + } +} + +template void checkVoidPointerDecorationConversion() { + using void_no = + multi_ptr_t; + using void_yes = + multi_ptr_t; + + TestData Data{100, 10.0f}; + auto TypedPtr = + makePtr(&Data); + + if constexpr (IsDecorated != decorated::legacy) { + auto VoidPtr = static_cast< + multi_ptr_t>(TypedPtr); + + if constexpr (IsDecorated == decorated::no) { + // Test implicit conversion to decorated::yes + void_yes ToDecorated = VoidPtr; + EXPECT_NE(ToDecorated.get(), nullptr); + + // Convert back to typed pointer to verify correctness + auto TypedCheck = static_cast< + multi_ptr_t>( + ToDecorated); + EXPECT_EQ(TypedCheck->Value, 100); + } else { + // Test implicit conversion to decorated::no + void_no ToUndecorated = VoidPtr; + EXPECT_NE(ToUndecorated.get(), nullptr); + + // Convert back to typed pointer to verify correctness + auto TypedCheck = static_cast< + multi_ptr_t>( + ToUndecorated); + EXPECT_EQ(TypedCheck->Value, 100); + } + } +} + +TEST(MultiPtrVoidSpecialization, + DefaultAndNullptrConstructionCreatesNullPointers) { + checkVoidPointerConstruction(); + checkVoidPointerConstruction(); + checkVoidPointerConstruction(); +} + +TEST(MultiPtrVoidSpecialization, CopyAndMoveConstructionPreservesPointerValue) { + checkVoidPointerCopyAndMove(); + checkVoidPointerCopyAndMove(); + checkVoidPointerCopyAndMove(); +} + +TEST(MultiPtrVoidSpecialization, AssignmentOperatorsWorkCorrectly) { + checkVoidPointerAssignment(); + checkVoidPointerAssignment(); + checkVoidPointerAssignment(); +} + +TEST(MultiPtrVoidSpecialization, ExplicitConversionToTypedPointerWorks) { + checkVoidPointerExplicitConversionToTyped(); + checkVoidPointerExplicitConversionToTyped(); + checkVoidPointerExplicitConversionToTyped(); +} + +TEST(MultiPtrVoidSpecialization, ConstVoidPointerHandlesConstCorrectly) { + checkConstVoidPointerConstConversions(); + checkConstVoidPointerConstConversions(); + checkConstVoidPointerConstConversions(); +} + +TEST(MultiPtrVoidSpecialization, ComparisonOperatorsWorkCorrectly) { + checkVoidPointerComparisons(); + checkVoidPointerComparisons(); + checkVoidPointerComparisons(); +} + +TEST(MultiPtrVoidSpecialization, GetMethodsReturnCorrectPointers) { + checkVoidPointerGetMethods(); + checkVoidPointerGetMethods(); + checkVoidPointerGetMethods(); +} + +TEST(MultiPtrVoidSpecialization, DecorationConversionsBetweenModesWork) { + checkVoidPointerDecorationConversion(); + checkVoidPointerDecorationConversion(); +} + +TEST(MultiPtrVoidSpecialization, VoidPointerWorksWithDifferentAddressSpaces) { + using global_void = + multi_ptr_t; + using local_void = + multi_ptr_t; + using generic_void = + multi_ptr_t; + + EXPECT_TRUE((std::is_default_constructible_v)); + EXPECT_TRUE((std::is_default_constructible_v)); + EXPECT_TRUE((std::is_default_constructible_v)); + + global_void GlobalNull; + local_void LocalNull; + generic_void GenericNull; + + EXPECT_EQ(GlobalNull, nullptr); + EXPECT_EQ(LocalNull, nullptr); + EXPECT_EQ(GenericNull, nullptr); +} + +TEST(MultiPtrVoidSpecialization, LegacyVoidPointerConstructsFromTypedPointer) { + using void_legacy = + multi_ptr_t; + using int_legacy = + multi_ptr_t; + + int Value = 999; + int_legacy IntPtr{&Value}; + + void_legacy VoidPtr{IntPtr}; + // Legacy void has get_raw() + EXPECT_EQ(VoidPtr.get_raw(), static_cast(&Value)); + + int_legacy RoundTrip = static_cast(VoidPtr); + EXPECT_EQ(RoundTrip.get_raw(), &Value); + EXPECT_EQ(*RoundTrip, 999); +} + +} // namespace From c513be085c44ad52feeab0dc8c739736c9d4e334 Mon Sep 17 00:00:00 2001 From: "Klochkov, Denis" Date: Mon, 25 May 2026 17:20:34 +0200 Subject: [PATCH 2/6] [SYCL][TEST] verify results in e2e tests --- sycl/test-e2e/multi_ptr/prefetch.cpp | 149 ++-- sycl/unittests/multi_ptr/Conversion.cpp | 710 +++++++++--------- .../multi_ptr/NegativeConversions.cpp | 2 - sycl/unittests/multi_ptr/Operators.cpp | 377 +++++----- 4 files changed, 638 insertions(+), 600 deletions(-) diff --git a/sycl/test-e2e/multi_ptr/prefetch.cpp b/sycl/test-e2e/multi_ptr/prefetch.cpp index c169aac4ff47e..83a42c5eef304 100644 --- a/sycl/test-e2e/multi_ptr/prefetch.cpp +++ b/sycl/test-e2e/multi_ptr/prefetch.cpp @@ -1,7 +1,8 @@ // RUN: %{build} -o %t.out // RUN: %{run} %t.out -//===------------------------- prefetch.cpp --------------------------------===// +//===------------------------- prefetch.cpp +//--------------------------------===// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. // See https://llvm.org/LICENSE.txt for license information. @@ -16,8 +17,7 @@ using namespace sycl; -template -void testPrefetchWithDecoration() { +template void testPrefetchWithDecoration() { constexpr size_t Size = 1024; std::vector HostData(Size); for (size_t i = 0; i < Size; ++i) { @@ -82,25 +82,29 @@ void testPrefetchWithGlobalPointer() { Q.submit([&](handler &CGH) { auto Acc = Buf.get_access(CGH); - CGH.parallel_for(range<1>(Size), [=](id<1> Index) { - using global_ptr = multi_ptr; - - global_ptr Ptr = address_space_cast(&Acc[0]); - - // Prefetch future data - if (Index[0] < Size - 50) { - auto FuturePtr = Ptr + Index[0] + 10; - FuturePtr.prefetch(20); - } - - // Process data - float Sum = 0.0f; - for (size_t i = 0; i < 5 && Index[0] + i < Size; ++i) { - Sum += Ptr[Index[0] + i]; - } - Acc[Index] = Sum; - }); + CGH.parallel_for( + range<1>(Size), [=](id<1> Index) { + using global_ptr = + multi_ptr; + + global_ptr Ptr = + address_space_cast(&Acc[0]); + + // Prefetch future data + if (Index[0] < Size - 50) { + auto FuturePtr = Ptr + Index[0] + 10; + FuturePtr.prefetch(20); + } + + // Process data + float Sum = 0.0f; + for (size_t i = 0; i < 5 && Index[0] + i < Size; ++i) { + Sum += Ptr[Index[0] + i]; + } + Acc[Index] = Sum; + }); }); Q.wait(); @@ -112,7 +116,8 @@ void testPrefetchWithGlobalPointer() { for (size_t j = 0; j < 5 && i + j < Size; ++j) { Expected += static_cast(i + j) * 0.5f; } - assert(std::abs(HostAcc[i] - Expected) < 0.001f && "Global prefetch test failed"); + assert(std::abs(HostAcc[i] - Expected) < 0.001f && + "Global prefetch test failed"); } } @@ -129,27 +134,34 @@ void testPrefetchWithLargeData() { Q.submit([&](handler &CGH) { auto Acc = Buf.get_access(CGH); - CGH.parallel_for(range<1>(Size / 8), [=](id<1> Index) { - auto Ptr = Acc.template get_multi_ptr(); - size_t BaseIndex = Index[0] * 8; - - // Prefetch a chunk of data - auto ChunkPtr = Ptr + BaseIndex; - ChunkPtr.prefetch(64); - - // Process the prefetched chunk - double Sum = 0.0; - for (size_t i = 0; i < 8; ++i) { - Sum += ChunkPtr[i]; - } - // Just to use the Sum (avoid optimization removal) - if (Sum < 0) { - ChunkPtr[0] = Sum; - } - }); + CGH.parallel_for( + range<1>(Size / 8), [=](id<1> Index) { + auto Ptr = Acc.template get_multi_ptr(); + size_t BaseIndex = Index[0] * 8; + + // Prefetch a chunk of data + auto ChunkPtr = Ptr + BaseIndex; + ChunkPtr.prefetch(64); + + // Process the prefetched chunk + double Sum = 0.0; + for (size_t i = 0; i < 8; ++i) { + Sum += ChunkPtr[i]; + } + // Just to use the Sum (avoid optimization removal) + if (Sum < 0) { + ChunkPtr[0] = Sum; + } + }); }); Q.wait(); + + auto HostAcc = Buf.get_host_access(); + for (size_t i = 0; i < Size; ++i) { + assert(HostAcc[i] == static_cast(i) && + "Large prefetch test failed"); + } } void testPrefetchAtBoundaries() { @@ -179,6 +191,11 @@ void testPrefetchAtBoundaries() { }); Q.wait(); + + auto HostAcc = Buf.get_host_access(); + for (size_t i = 0; i < Size; ++i) { + assert(HostAcc[i] == 42 && "Boundary prefetch test failed"); + } } void testPrefetchWithStructs() { @@ -191,7 +208,8 @@ void testPrefetchWithStructs() { constexpr size_t Size = 128; std::vector HostData(Size); for (size_t i = 0; i < Size; ++i) { - HostData[i] = {static_cast(i), static_cast(i) * 1.5f, static_cast(i) * 2.5}; + HostData[i] = {static_cast(i), static_cast(i) * 1.5f, + static_cast(i) * 2.5}; } queue Q; @@ -200,27 +218,38 @@ void testPrefetchWithStructs() { Q.submit([&](handler &CGH) { auto Acc = Buf.get_access(CGH); - CGH.parallel_for(range<1>(Size / 2), [=](id<1> Index) { - auto Ptr = Acc.template get_multi_ptr(); - size_t BaseIdx = Index[0] * 2; - - // Prefetch structures - auto StructPtr = Ptr + BaseIdx; - StructPtr.prefetch(8); - - // Access the data - int Sum = 0; - for (size_t i = 0; i < 2 && BaseIdx + i < Size; ++i) { - Sum += StructPtr[i].A; - } - // Use sum to prevent optimization - if (Sum < 0) { - StructPtr[0].A = Sum; - } - }); + CGH.parallel_for( + range<1>(Size / 2), [=](id<1> Index) { + auto Ptr = Acc.template get_multi_ptr(); + size_t BaseIdx = Index[0] * 2; + + // Prefetch structures + auto StructPtr = Ptr + BaseIdx; + StructPtr.prefetch(8); + + // Access the data + int Sum = 0; + for (size_t i = 0; i < 2 && BaseIdx + i < Size; ++i) { + Sum += StructPtr[i].A; + } + // Use sum to prevent optimization + if (Sum < 0) { + StructPtr[0].A = Sum; + } + }); }); Q.wait(); + + auto HostAcc = Buf.get_host_access(); + for (size_t i = 0; i < Size; ++i) { + assert(HostAcc[i].A == static_cast(i) && + "Struct prefetch integer field test failed"); + assert(std::abs(HostAcc[i].B - static_cast(i) * 1.5f) < 0.001f && + "Struct prefetch float field test failed"); + assert(std::abs(HostAcc[i].C - static_cast(i) * 2.5) < 0.001 && + "Struct prefetch double field test failed"); + } } int main() { diff --git a/sycl/unittests/multi_ptr/Conversion.cpp b/sycl/unittests/multi_ptr/Conversion.cpp index fe9cf71411ea8..555a76a191415 100644 --- a/sycl/unittests/multi_ptr/Conversion.cpp +++ b/sycl/unittests/multi_ptr/Conversion.cpp @@ -12,8 +12,8 @@ using sycl::access::address_space; using sycl::access::decorated; struct Record { - int Value; - int Tag; + int Value; + int Tag; }; template @@ -21,8 +21,8 @@ using multi_ptr_t = sycl::multi_ptr; template using access_element_t = - std::conditional_t; + std::conditional_t; template using access_storage_t = std::remove_const_t>; @@ -30,403 +30,405 @@ using access_storage_t = std::remove_const_t>; template multi_ptr_t makePtr(std::add_pointer_t Ptr) { - if constexpr (IsDecorated == decorated::legacy) { - return multi_ptr_t{Ptr}; - } else { - return sycl::address_space_cast(Ptr); - } + if constexpr (IsDecorated == decorated::legacy) { + return multi_ptr_t{Ptr}; + } else { + return sycl::address_space_cast(Ptr); + } } template constexpr decorated toggledDecoration() { - static_assert(IsDecorated != decorated::legacy); - return IsDecorated == decorated::no ? decorated::yes : decorated::no; + static_assert(IsDecorated != decorated::legacy); + return IsDecorated == decorated::no ? decorated::yes : decorated::no; } template void checkPointerAccessAndUnderlyingPointers() { - using element_t = access_element_t; - using ptr_t = multi_ptr_t; - - access_storage_t Values[] = {{11, 111}, {22, 222}}; - auto *RawPtr = static_cast>(&Values[0]); - ptr_t Ptr = makePtr(RawPtr); - - EXPECT_EQ(Ptr->Value, 11); - EXPECT_EQ(Ptr->Tag, 111); - EXPECT_EQ(Ptr[1].Value, 22); - EXPECT_EQ(Ptr[1].Tag, 222); - - decltype(Ptr.get()) PointerValue = Ptr; - EXPECT_EQ(PointerValue, Ptr.get()); - EXPECT_EQ(Ptr.get_raw(), RawPtr); - - if constexpr (IsDecorated == decorated::legacy) { - EXPECT_EQ(Ptr.get_decorated(), Ptr.get()); - } else { - auto DecoratedPtr = makePtr(RawPtr); - EXPECT_EQ(Ptr.get_decorated(), DecoratedPtr.get_decorated()); - } + using element_t = access_element_t; + using ptr_t = multi_ptr_t; + + access_storage_t Values[] = {{11, 111}, {22, 222}}; + auto *RawPtr = static_cast>(&Values[0]); + ptr_t Ptr = makePtr(RawPtr); + + EXPECT_EQ(Ptr->Value, 11); + EXPECT_EQ(Ptr->Tag, 111); + EXPECT_EQ(Ptr[1].Value, 22); + EXPECT_EQ(Ptr[1].Tag, 222); + + decltype(Ptr.get()) PointerValue = Ptr; + EXPECT_EQ(PointerValue, Ptr.get()); + EXPECT_EQ(Ptr.get_raw(), RawPtr); + + if constexpr (IsDecorated == decorated::legacy) { + EXPECT_EQ(Ptr.get_decorated(), Ptr.get()); + } else { + auto DecoratedPtr = makePtr(RawPtr); + EXPECT_EQ(Ptr.get_decorated(), DecoratedPtr.get_decorated()); + } } template void checkGenericExplicitMutableConversions() { - using source_t = - multi_ptr_t; - using private_t = - multi_ptr_t; - using global_t = - multi_ptr_t; - using local_t = - multi_ptr_t; - using const_private_t = - multi_ptr_t; - using const_global_t = - multi_ptr_t; - using const_local_t = - multi_ptr_t; - - Record Value{31, 131}; - source_t Source = - makePtr(&Value); - - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_FALSE((std::is_convertible_v)); - EXPECT_FALSE((std::is_convertible_v)); - EXPECT_FALSE((std::is_convertible_v)); - - private_t Private{Source}; - global_t Global{Source}; - local_t Local{Source}; - const_private_t ConstPrivate{Source}; - const_global_t ConstGlobal{Source}; - const_local_t ConstLocal{Source}; - - EXPECT_EQ(Private.get_raw(), &Value); - EXPECT_EQ(Global.get_raw(), &Value); - EXPECT_EQ(Local.get_raw(), &Value); - EXPECT_EQ(ConstPrivate.get_raw(), &Value); - EXPECT_EQ(ConstGlobal.get_raw(), &Value); - EXPECT_EQ(ConstLocal.get_raw(), &Value); + using source_t = + multi_ptr_t; + using private_t = + multi_ptr_t; + using global_t = + multi_ptr_t; + using local_t = + multi_ptr_t; + using const_private_t = + multi_ptr_t; + using const_global_t = + multi_ptr_t; + using const_local_t = + multi_ptr_t; + + Record Value{31, 131}; + source_t Source = + makePtr(&Value); + + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + + private_t Private{Source}; + global_t Global{Source}; + local_t Local{Source}; + const_private_t ConstPrivate{Source}; + const_global_t ConstGlobal{Source}; + const_local_t ConstLocal{Source}; + + EXPECT_EQ(Private.get_raw(), &Value); + EXPECT_EQ(Global.get_raw(), &Value); + EXPECT_EQ(Local.get_raw(), &Value); + EXPECT_EQ(ConstPrivate.get_raw(), &Value); + EXPECT_EQ(ConstGlobal.get_raw(), &Value); + EXPECT_EQ(ConstLocal.get_raw(), &Value); } template void checkGenericExplicitConstConversions() { - using source_t = - multi_ptr_t; - using private_t = - multi_ptr_t; - using global_t = - multi_ptr_t; - using local_t = - multi_ptr_t; - - const Record Value{32, 132}; - source_t Source = makePtr(&Value); - - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_FALSE((std::is_convertible_v)); - EXPECT_FALSE((std::is_convertible_v)); - EXPECT_FALSE((std::is_convertible_v)); - - private_t Private{Source}; - global_t Global{Source}; - local_t Local{Source}; - - EXPECT_EQ(Private.get_raw(), &Value); - EXPECT_EQ(Global.get_raw(), &Value); - EXPECT_EQ(Local.get_raw(), &Value); + using source_t = + multi_ptr_t; + using private_t = + multi_ptr_t; + using global_t = + multi_ptr_t; + using local_t = + multi_ptr_t; + + const Record Value{32, 132}; + source_t Source = + makePtr( + &Value); + + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + EXPECT_FALSE((std::is_convertible_v)); + + private_t Private{Source}; + global_t Global{Source}; + local_t Local{Source}; + + EXPECT_EQ(Private.get_raw(), &Value); + EXPECT_EQ(Global.get_raw(), &Value); + EXPECT_EQ(Local.get_raw(), &Value); } template void checkNonLegacyMutableSameSpaceConversions() { - static_assert(SourceDecoration != decorated::legacy); - - using source_t = multi_ptr_t; - using typed_no_t = multi_ptr_t; - using typed_yes_t = multi_ptr_t; - using typed_legacy_t = multi_ptr_t; - using toggled_t = - multi_ptr_t()>; - using void_same_t = multi_ptr_t; - using void_toggled_t = - multi_ptr_t()>; - using void_no_t = multi_ptr_t; - using void_yes_t = multi_ptr_t; - using void_legacy_t = multi_ptr_t; - using const_no_t = multi_ptr_t; - using const_yes_t = multi_ptr_t; - using const_legacy_t = multi_ptr_t; - - Record Value{41, 141}; - source_t Source = makePtr(&Value); - - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_FALSE((std::is_convertible_v)); - - toggled_t Toggled = Source; - void_no_t VoidNo = Source; - void_yes_t VoidYes = Source; - void_legacy_t VoidLegacy = Source; - const_no_t ConstNo = Source; - const_yes_t ConstYes = Source; - const_legacy_t ConstLegacy = Source; - void_same_t VoidSame = Source; - void_toggled_t VoidToggled = VoidSame; - typed_no_t FromVoidNo = static_cast(VoidNo); - typed_yes_t FromVoidYes = static_cast(VoidYes); - typed_legacy_t FromVoidLegacy = static_cast(VoidLegacy); - toggled_t FromVoidToggled = static_cast(VoidToggled); - source_t FromVoid = static_cast(VoidSame); - - EXPECT_EQ(Toggled.get_raw(), &Value); - EXPECT_EQ(ConstNo.get_raw(), &Value); - EXPECT_EQ(ConstYes.get_raw(), &Value); - EXPECT_EQ(ConstLegacy.get_raw(), &Value); - EXPECT_EQ(FromVoidNo.get_raw(), &Value); - EXPECT_EQ(FromVoidYes.get_raw(), &Value); - EXPECT_EQ(FromVoidLegacy.get_raw(), &Value); - EXPECT_EQ(FromVoidToggled.get_raw(), &Value); - EXPECT_EQ(FromVoid.get_raw(), &Value); + static_assert(SourceDecoration != decorated::legacy); + + using source_t = multi_ptr_t; + using typed_no_t = multi_ptr_t; + using typed_yes_t = multi_ptr_t; + using typed_legacy_t = multi_ptr_t; + using toggled_t = + multi_ptr_t()>; + using void_same_t = multi_ptr_t; + using void_toggled_t = + multi_ptr_t()>; + using void_no_t = multi_ptr_t; + using void_yes_t = multi_ptr_t; + using void_legacy_t = multi_ptr_t; + using const_no_t = multi_ptr_t; + using const_yes_t = multi_ptr_t; + using const_legacy_t = multi_ptr_t; + + Record Value{41, 141}; + source_t Source = makePtr(&Value); + + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + + toggled_t Toggled = Source; + void_no_t VoidNo = Source; + void_yes_t VoidYes = Source; + void_legacy_t VoidLegacy = Source; + const_no_t ConstNo = Source; + const_yes_t ConstYes = Source; + const_legacy_t ConstLegacy = Source; + void_same_t VoidSame = Source; + void_toggled_t VoidToggled = VoidSame; + typed_no_t FromVoidNo = static_cast(VoidNo); + typed_yes_t FromVoidYes = static_cast(VoidYes); + typed_legacy_t FromVoidLegacy = static_cast(VoidLegacy); + toggled_t FromVoidToggled = static_cast(VoidToggled); + source_t FromVoid = static_cast(VoidSame); + + EXPECT_EQ(Toggled.get_raw(), &Value); + EXPECT_EQ(ConstNo.get_raw(), &Value); + EXPECT_EQ(ConstYes.get_raw(), &Value); + EXPECT_EQ(ConstLegacy.get_raw(), &Value); + EXPECT_EQ(FromVoidNo.get_raw(), &Value); + EXPECT_EQ(FromVoidYes.get_raw(), &Value); + EXPECT_EQ(FromVoidLegacy.get_raw(), &Value); + EXPECT_EQ(FromVoidToggled.get_raw(), &Value); + EXPECT_EQ(FromVoid.get_raw(), &Value); } template void checkNonLegacyConstSameSpaceConversions() { - static_assert(SourceDecoration != decorated::legacy); - - using source_t = multi_ptr_t; - using typed_no_t = multi_ptr_t; - using typed_yes_t = multi_ptr_t; - using typed_legacy_t = multi_ptr_t; - using toggled_t = - multi_ptr_t()>; - using const_void_same_t = multi_ptr_t; - using const_void_toggled_t = - multi_ptr_t()>; - using const_void_no_t = multi_ptr_t; - using const_void_yes_t = multi_ptr_t; - using const_void_legacy_t = multi_ptr_t; - - const Record Value{42, 142}; - source_t Source = makePtr(&Value); - - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_FALSE((std::is_convertible_v)); - - toggled_t Toggled = Source; - const_void_no_t ConstVoidNo = Source; - const_void_yes_t ConstVoidYes = Source; - const_void_legacy_t ConstVoidLegacy = Source; - const_void_same_t ConstVoidSame = Source; - const_void_toggled_t ConstVoidToggled = ConstVoidSame; - typed_no_t FromConstVoidNo = static_cast(ConstVoidNo); - typed_yes_t FromConstVoidYes = static_cast(ConstVoidYes); - typed_legacy_t FromConstVoidLegacy = static_cast(ConstVoidLegacy); - toggled_t FromConstVoidToggled = static_cast(ConstVoidToggled); - source_t FromConstVoid = static_cast(ConstVoidSame); - - EXPECT_EQ(Toggled.get_raw(), &Value); - EXPECT_EQ(FromConstVoidNo.get_raw(), &Value); - EXPECT_EQ(FromConstVoidYes.get_raw(), &Value); - EXPECT_EQ(FromConstVoidLegacy.get_raw(), &Value); - EXPECT_EQ(FromConstVoidToggled.get_raw(), &Value); - EXPECT_EQ(FromConstVoid.get_raw(), &Value); + static_assert(SourceDecoration != decorated::legacy); + + using source_t = multi_ptr_t; + using typed_no_t = multi_ptr_t; + using typed_yes_t = multi_ptr_t; + using typed_legacy_t = multi_ptr_t; + using toggled_t = + multi_ptr_t()>; + using const_void_same_t = multi_ptr_t; + using const_void_toggled_t = + multi_ptr_t()>; + using const_void_no_t = multi_ptr_t; + using const_void_yes_t = multi_ptr_t; + using const_void_legacy_t = multi_ptr_t; + + const Record Value{42, 142}; + source_t Source = makePtr(&Value); + + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_FALSE((std::is_convertible_v)); + + toggled_t Toggled = Source; + const_void_no_t ConstVoidNo = Source; + const_void_yes_t ConstVoidYes = Source; + const_void_legacy_t ConstVoidLegacy = Source; + const_void_same_t ConstVoidSame = Source; + const_void_toggled_t ConstVoidToggled = ConstVoidSame; + typed_no_t FromConstVoidNo = static_cast(ConstVoidNo); + typed_yes_t FromConstVoidYes = static_cast(ConstVoidYes); + typed_legacy_t FromConstVoidLegacy = + static_cast(ConstVoidLegacy); + toggled_t FromConstVoidToggled = static_cast(ConstVoidToggled); + source_t FromConstVoid = static_cast(ConstVoidSame); + + EXPECT_EQ(Toggled.get_raw(), &Value); + EXPECT_EQ(FromConstVoidNo.get_raw(), &Value); + EXPECT_EQ(FromConstVoidYes.get_raw(), &Value); + EXPECT_EQ(FromConstVoidLegacy.get_raw(), &Value); + EXPECT_EQ(FromConstVoidToggled.get_raw(), &Value); + EXPECT_EQ(FromConstVoid.get_raw(), &Value); } template void checkLegacyConversions() { - using mutable_t = multi_ptr_t; - using const_t = multi_ptr_t; - using void_t = multi_ptr_t; - using const_void_t = multi_ptr_t; - - Record MutableValue{51, 151}; - const Record ConstValue{52, 152}; - - mutable_t Mutable = makePtr(&MutableValue); - const_t Const = makePtr(&ConstValue); - - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_convertible_v)); - EXPECT_TRUE((std::is_constructible_v)); - EXPECT_TRUE((std::is_constructible_v)); - - void_t VoidPtr = Mutable; - const_t ConstFromMutable = Mutable; - const_void_t ConstVoidPtr = Const; - const_void_t ConstVoidFromVoid = VoidPtr; - mutable_t MutableRoundTrip = static_cast(VoidPtr); - const_t ConstRoundTrip = static_cast(ConstVoidPtr); - - EXPECT_EQ(VoidPtr.get_raw(), static_cast(&MutableValue)); - EXPECT_EQ(ConstFromMutable.get_raw(), &MutableValue); - EXPECT_EQ(ConstVoidPtr.get_raw(), static_cast(&ConstValue)); - EXPECT_EQ(ConstVoidFromVoid.get_raw(), static_cast(&MutableValue)); - EXPECT_EQ(MutableRoundTrip.get_raw(), &MutableValue); - EXPECT_EQ(ConstRoundTrip.get_raw(), &ConstValue); + using mutable_t = multi_ptr_t; + using const_t = multi_ptr_t; + using void_t = multi_ptr_t; + using const_void_t = multi_ptr_t; + + Record MutableValue{51, 151}; + const Record ConstValue{52, 152}; + + mutable_t Mutable = makePtr(&MutableValue); + const_t Const = makePtr(&ConstValue); + + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_convertible_v)); + EXPECT_TRUE((std::is_constructible_v)); + EXPECT_TRUE((std::is_constructible_v)); + + void_t VoidPtr = Mutable; + const_t ConstFromMutable = Mutable; + const_void_t ConstVoidPtr = Const; + const_void_t ConstVoidFromVoid = VoidPtr; + mutable_t MutableRoundTrip = static_cast(VoidPtr); + const_t ConstRoundTrip = static_cast(ConstVoidPtr); + + EXPECT_EQ(VoidPtr.get_raw(), static_cast(&MutableValue)); + EXPECT_EQ(ConstFromMutable.get_raw(), &MutableValue); + EXPECT_EQ(ConstVoidPtr.get_raw(), static_cast(&ConstValue)); + EXPECT_EQ(ConstVoidFromVoid.get_raw(), + static_cast(&MutableValue)); + EXPECT_EQ(MutableRoundTrip.get_raw(), &MutableValue); + EXPECT_EQ(ConstRoundTrip.get_raw(), &ConstValue); } template void checkFactoryFunctions() { - using ptr_t = multi_ptr_t; - - int Value = 77; - auto CastResult = - sycl::address_space_cast( - &Value); - using cast_result_t = decltype(CastResult); - EXPECT_TRUE((std::is_same_v)); - EXPECT_EQ(CastResult.get_raw(), &Value); - - auto NullCast = - sycl::address_space_cast( - static_cast(nullptr)); - EXPECT_EQ(NullCast, nullptr); - - auto MakeResult = - sycl::make_ptr( - CastResult.get()); - using make_result_t = decltype(MakeResult); - EXPECT_TRUE((std::is_same_v)); - EXPECT_EQ(MakeResult.get_raw(), &Value); - - using underlying_pointer_t = decltype(CastResult.get()); - underlying_pointer_t NullPointer = nullptr; - auto NullMake = - sycl::make_ptr( - NullPointer); - EXPECT_EQ(NullMake, nullptr); + using ptr_t = multi_ptr_t; + + int Value = 77; + auto CastResult = + sycl::address_space_cast( + &Value); + using cast_result_t = decltype(CastResult); + EXPECT_TRUE((std::is_same_v)); + EXPECT_EQ(CastResult.get_raw(), &Value); + + auto NullCast = + sycl::address_space_cast( + static_cast(nullptr)); + EXPECT_EQ(NullCast, nullptr); + + auto MakeResult = + sycl::make_ptr( + CastResult.get()); + using make_result_t = decltype(MakeResult); + EXPECT_TRUE((std::is_same_v)); + EXPECT_EQ(MakeResult.get_raw(), &Value); + + using underlying_pointer_t = decltype(CastResult.get()); + underlying_pointer_t NullPointer = nullptr; + auto NullMake = + sycl::make_ptr( + NullPointer); + EXPECT_EQ(NullMake, nullptr); } void checkDefaultLegacyMakePtr() { - int Value = 88; - auto LegacyPtr = - sycl::make_ptr(&Value); - EXPECT_TRUE((std::is_same_v>)); - EXPECT_EQ(LegacyPtr.get_raw(), &Value); + int Value = 88; + auto LegacyPtr = sycl::make_ptr(&Value); + EXPECT_TRUE( + (std::is_same_v< + decltype(LegacyPtr), + multi_ptr_t>)); + EXPECT_EQ(LegacyPtr.get_raw(), &Value); } TEST(MultiPtrConversion, - PointerAccessPointerConversionAndGetDecoratedCoverAllEnumValues) { - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); - checkPointerAccessAndUnderlyingPointers(); + PointerAccessPointerConversionAndGetDecoratedCoverAllEnumValues) { + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); + checkPointerAccessAndUnderlyingPointers(); } TEST(MultiPtrConversion, - GenericSpaceExplicitConversionsSupportAllNonLegacyDecorationPairs) { - checkGenericExplicitMutableConversions(); - checkGenericExplicitMutableConversions(); - checkGenericExplicitMutableConversions(); - checkGenericExplicitMutableConversions(); - - checkGenericExplicitConstConversions(); - checkGenericExplicitConstConversions(); - checkGenericExplicitConstConversions(); - checkGenericExplicitConstConversions(); + GenericSpaceExplicitConversionsSupportAllNonLegacyDecorationPairs) { + checkGenericExplicitMutableConversions(); + checkGenericExplicitMutableConversions(); + checkGenericExplicitMutableConversions(); + checkGenericExplicitMutableConversions(); + + checkGenericExplicitConstConversions(); + checkGenericExplicitConstConversions(); + checkGenericExplicitConstConversions(); + checkGenericExplicitConstConversions(); } TEST(MultiPtrConversion, - NonLegacySameSpaceConversionsCoverVoidConstAndDecorationChanges) { - checkNonLegacyMutableSameSpaceConversions(); - checkNonLegacyMutableSameSpaceConversions(); - checkNonLegacyMutableSameSpaceConversions(); - checkNonLegacyMutableSameSpaceConversions(); - checkNonLegacyMutableSameSpaceConversions(); - checkNonLegacyMutableSameSpaceConversions(); - checkNonLegacyMutableSameSpaceConversions(); - checkNonLegacyMutableSameSpaceConversions(); - - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); - checkNonLegacyConstSameSpaceConversions(); + NonLegacySameSpaceConversionsCoverVoidConstAndDecorationChanges) { + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + checkNonLegacyMutableSameSpaceConversions(); + + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); + checkNonLegacyConstSameSpaceConversions(); } TEST(MultiPtrConversion, LegacyConversionsCoverAllAddressSpaces) { - checkLegacyConversions(); - checkLegacyConversions(); - checkLegacyConversions(); - checkLegacyConversions(); - checkLegacyConversions(); + checkLegacyConversions(); + checkLegacyConversions(); + checkLegacyConversions(); + checkLegacyConversions(); + checkLegacyConversions(); } TEST(MultiPtrConversion, FactoryFunctionsPreservePrivatePointersAndNullptr) { - checkFactoryFunctions(); - checkFactoryFunctions(); - checkFactoryFunctions(); - checkDefaultLegacyMakePtr(); + checkFactoryFunctions(); + checkFactoryFunctions(); + checkFactoryFunctions(); + checkDefaultLegacyMakePtr(); } } // namespace - diff --git a/sycl/unittests/multi_ptr/NegativeConversions.cpp b/sycl/unittests/multi_ptr/NegativeConversions.cpp index 6f2b42047368d..ecdfdfa682093 100644 --- a/sycl/unittests/multi_ptr/NegativeConversions.cpp +++ b/sycl/unittests/multi_ptr/NegativeConversions.cpp @@ -45,7 +45,6 @@ TEST(MultiPtrNegativeConversions, CannotConvertBetweenIncompatibleTypes) { EXPECT_FALSE((std::is_convertible_v)); } - TEST(MultiPtrNegativeConversions, CannotConvertToGenericFromNonGeneric) { using private_ptr = multi_ptr_t; @@ -131,7 +130,6 @@ TEST(MultiPtrNegativeConversions, ConstantSpaceNotSupportedForVoidInNonLegacy) { EXPECT_TRUE((std::is_default_constructible_v)); } - TEST(MultiPtrNegativeConversions, CannotComparePointersOfDifferentTypes) { using int_ptr = multi_ptr_t; using float_ptr = diff --git a/sycl/unittests/multi_ptr/Operators.cpp b/sycl/unittests/multi_ptr/Operators.cpp index 2468e9e262ef7..63a296053ac9c 100644 --- a/sycl/unittests/multi_ptr/Operators.cpp +++ b/sycl/unittests/multi_ptr/Operators.cpp @@ -21,227 +21,236 @@ using sycl::access::decorated; template using private_multi_ptr = - sycl::multi_ptr; + sycl::multi_ptr; template private_multi_ptr makePrivatePtr(std::add_pointer_t Ptr) { - return sycl::address_space_cast( - Ptr); + return sycl::address_space_cast( + Ptr); } struct Record { - int Value; - int Tag; + int Value; + int Tag; }; template void checkOperatorTraits() { - using ptr_t = private_multi_ptr; - using diff_t = typename ptr_t::difference_type; - - EXPECT_TRUE( - (std::is_same_v()), - typename ptr_t::reference>)); - EXPECT_TRUE((std::is_same_v()[diff_t{0}]), - typename ptr_t::reference>)); - EXPECT_TRUE((std::is_same_v().operator->()), - typename ptr_t::pointer>)); - - EXPECT_TRUE((std::is_same_v()), ptr_t &>)); - EXPECT_TRUE((std::is_same_v()++), ptr_t>)); - EXPECT_TRUE((std::is_same_v()), ptr_t &>)); - EXPECT_TRUE((std::is_same_v()--), ptr_t>)); - EXPECT_TRUE((std::is_same_v() += diff_t{0}), - ptr_t &>)); - EXPECT_TRUE((std::is_same_v() -= diff_t{0}), - ptr_t &>)); - EXPECT_TRUE((std::is_same_v() + diff_t{0}), - ptr_t>)); - EXPECT_TRUE((std::is_same_v() - diff_t{0}), - ptr_t>)); - - EXPECT_TRUE((std::is_same_v() == - std::declval()), - bool>)); - EXPECT_TRUE((std::is_same_v() != - std::declval()), - bool>)); - EXPECT_TRUE((std::is_same_v() < - std::declval()), - bool>)); - EXPECT_TRUE((std::is_same_v() > - std::declval()), - bool>)); - EXPECT_TRUE((std::is_same_v() <= - std::declval()), - bool>)); - EXPECT_TRUE((std::is_same_v() >= - std::declval()), - bool>)); - - EXPECT_TRUE((std::is_same_v() == nullptr), - bool>)); - EXPECT_TRUE((std::is_same_v() != nullptr), - bool>)); - EXPECT_TRUE((std::is_same_v() < nullptr), - bool>)); - EXPECT_TRUE((std::is_same_v() > nullptr), - bool>)); - EXPECT_TRUE((std::is_same_v() <= nullptr), - bool>)); - EXPECT_TRUE((std::is_same_v() >= nullptr), - bool>)); - EXPECT_TRUE((std::is_same_v()), - bool>)); - EXPECT_TRUE((std::is_same_v()), - bool>)); - EXPECT_TRUE((std::is_same_v()), - bool>)); - EXPECT_TRUE((std::is_same_v std::declval()), - bool>)); - EXPECT_TRUE((std::is_same_v()), - bool>)); - EXPECT_TRUE((std::is_same_v= std::declval()), - bool>)); + using ptr_t = private_multi_ptr; + using diff_t = typename ptr_t::difference_type; + + EXPECT_TRUE((std::is_same_v()), + typename ptr_t::reference>)); + EXPECT_TRUE( + (std::is_same_v()[diff_t{0}]), + typename ptr_t::reference>)); + EXPECT_TRUE( + (std::is_same_v().operator->()), + typename ptr_t::pointer>)); + + EXPECT_TRUE((std::is_same_v()), ptr_t &>)); + EXPECT_TRUE((std::is_same_v()++), ptr_t>)); + EXPECT_TRUE((std::is_same_v()), ptr_t &>)); + EXPECT_TRUE((std::is_same_v()--), ptr_t>)); + EXPECT_TRUE(( + std::is_same_v() += diff_t{0}), ptr_t &>)); + EXPECT_TRUE(( + std::is_same_v() -= diff_t{0}), ptr_t &>)); + EXPECT_TRUE( + (std::is_same_v() + diff_t{0}), + ptr_t>)); + EXPECT_TRUE( + (std::is_same_v() - diff_t{0}), + ptr_t>)); + + EXPECT_TRUE((std::is_same_v() == + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() != + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() < + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() > + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() <= + std::declval()), + bool>)); + EXPECT_TRUE((std::is_same_v() >= + std::declval()), + bool>)); + + EXPECT_TRUE( + (std::is_same_v() == nullptr), + bool>)); + EXPECT_TRUE( + (std::is_same_v() != nullptr), + bool>)); + EXPECT_TRUE(( + std::is_same_v() < nullptr), bool>)); + EXPECT_TRUE(( + std::is_same_v() > nullptr), bool>)); + EXPECT_TRUE( + (std::is_same_v() <= nullptr), + bool>)); + EXPECT_TRUE( + (std::is_same_v() >= nullptr), + bool>)); + EXPECT_TRUE( + (std::is_same_v()), + bool>)); + EXPECT_TRUE( + (std::is_same_v()), + bool>)); + EXPECT_TRUE(( + std::is_same_v()), bool>)); + EXPECT_TRUE(( + std::is_same_v std::declval()), bool>)); + EXPECT_TRUE( + (std::is_same_v()), + bool>)); + EXPECT_TRUE( + (std::is_same_v= std::declval()), + bool>)); } template void checkAccessOperators() { - Record Values[] = {{10, 100}, {20, 200}, {30, 300}}; - auto Ptr = makePrivatePtr(&Values[0]); - - EXPECT_EQ((*Ptr).Value, 10); - EXPECT_EQ((*Ptr).Tag, 100); - EXPECT_EQ(Ptr->Value, 10); - EXPECT_EQ(Ptr->Tag, 100); - EXPECT_EQ(Ptr[1].Value, 20); - EXPECT_EQ(Ptr[1].Tag, 200); - EXPECT_EQ(Ptr[2].Value, 30); - EXPECT_EQ(Ptr[2].Tag, 300); + Record Values[] = {{10, 100}, {20, 200}, {30, 300}}; + auto Ptr = makePrivatePtr(&Values[0]); + + EXPECT_EQ((*Ptr).Value, 10); + EXPECT_EQ((*Ptr).Tag, 100); + EXPECT_EQ(Ptr->Value, 10); + EXPECT_EQ(Ptr->Tag, 100); + EXPECT_EQ(Ptr[1].Value, 20); + EXPECT_EQ(Ptr[1].Tag, 200); + EXPECT_EQ(Ptr[2].Value, 30); + EXPECT_EQ(Ptr[2].Tag, 300); } template void checkArithmeticOperators() { - Record Values[] = {{1, 10}, {2, 20}, {3, 30}, {4, 40}}; - using ptr_t = private_multi_ptr; - using diff_t = typename ptr_t::difference_type; - - auto Ptr = makePrivatePtr(&Values[0]); - auto Expected0 = makePrivatePtr(&Values[0]); - auto Expected1 = makePrivatePtr(&Values[1]); - auto Expected2 = makePrivatePtr(&Values[2]); - auto Expected3 = makePrivatePtr(&Values[3]); - - ptr_t &PrefixIncrement = ++Ptr; - EXPECT_EQ(std::addressof(PrefixIncrement), std::addressof(Ptr)); - EXPECT_EQ(Ptr, Expected1); - - ptr_t PostfixIncrement = Ptr++; - EXPECT_EQ(PostfixIncrement, Expected1); - EXPECT_EQ(Ptr, Expected2); - - ptr_t &PrefixDecrement = --Ptr; - EXPECT_EQ(std::addressof(PrefixDecrement), std::addressof(Ptr)); - EXPECT_EQ(Ptr, Expected1); - - ptr_t PostfixDecrement = Ptr--; - EXPECT_EQ(PostfixDecrement, Expected1); - EXPECT_EQ(Ptr, Expected0); - - ptr_t &PlusAssign = (Ptr += diff_t{3}); - EXPECT_EQ(std::addressof(PlusAssign), std::addressof(Ptr)); - EXPECT_EQ(Ptr, Expected3); - - ptr_t &MinusAssign = (Ptr -= diff_t{2}); - EXPECT_EQ(std::addressof(MinusAssign), std::addressof(Ptr)); - EXPECT_EQ(Ptr, Expected1); - - ptr_t PlusResult = Ptr + diff_t{2}; - EXPECT_EQ(PlusResult, Expected3); - EXPECT_EQ(Ptr, Expected1); - - ptr_t MinusResult = Ptr - diff_t{1}; - EXPECT_EQ(MinusResult, Expected0); - EXPECT_EQ(Ptr, Expected1); + Record Values[] = {{1, 10}, {2, 20}, {3, 30}, {4, 40}}; + using ptr_t = private_multi_ptr; + using diff_t = typename ptr_t::difference_type; + + auto Ptr = makePrivatePtr(&Values[0]); + auto Expected0 = makePrivatePtr(&Values[0]); + auto Expected1 = makePrivatePtr(&Values[1]); + auto Expected2 = makePrivatePtr(&Values[2]); + auto Expected3 = makePrivatePtr(&Values[3]); + + ptr_t &PrefixIncrement = ++Ptr; + EXPECT_EQ(std::addressof(PrefixIncrement), std::addressof(Ptr)); + EXPECT_EQ(Ptr, Expected1); + + ptr_t PostfixIncrement = Ptr++; + EXPECT_EQ(PostfixIncrement, Expected1); + EXPECT_EQ(Ptr, Expected2); + + ptr_t &PrefixDecrement = --Ptr; + EXPECT_EQ(std::addressof(PrefixDecrement), std::addressof(Ptr)); + EXPECT_EQ(Ptr, Expected1); + + ptr_t PostfixDecrement = Ptr--; + EXPECT_EQ(PostfixDecrement, Expected1); + EXPECT_EQ(Ptr, Expected0); + + ptr_t &PlusAssign = (Ptr += diff_t{3}); + EXPECT_EQ(std::addressof(PlusAssign), std::addressof(Ptr)); + EXPECT_EQ(Ptr, Expected3); + + ptr_t &MinusAssign = (Ptr -= diff_t{2}); + EXPECT_EQ(std::addressof(MinusAssign), std::addressof(Ptr)); + EXPECT_EQ(Ptr, Expected1); + + ptr_t PlusResult = Ptr + diff_t{2}; + EXPECT_EQ(PlusResult, Expected3); + EXPECT_EQ(Ptr, Expected1); + + ptr_t MinusResult = Ptr - diff_t{1}; + EXPECT_EQ(MinusResult, Expected0); + EXPECT_EQ(Ptr, Expected1); } template void checkRelationalOperators() { - int Values[] = {5, 6, 7}; - - auto First = makePrivatePtr(&Values[0]); - auto SameAsFirst = makePrivatePtr(&Values[0]); - auto Second = makePrivatePtr(&Values[1]); - - EXPECT_TRUE(First == SameAsFirst); - EXPECT_FALSE(First != SameAsFirst); - EXPECT_FALSE(First < SameAsFirst); - EXPECT_FALSE(First > SameAsFirst); - EXPECT_TRUE(First <= SameAsFirst); - EXPECT_TRUE(First >= SameAsFirst); - - EXPECT_FALSE(First == Second); - EXPECT_TRUE(First != Second); - EXPECT_TRUE(First < Second); - EXPECT_FALSE(First > Second); - EXPECT_TRUE(First <= Second); - EXPECT_FALSE(First >= Second); - - EXPECT_FALSE(Second == First); - EXPECT_TRUE(Second != First); - EXPECT_FALSE(Second < First); - EXPECT_TRUE(Second > First); - EXPECT_FALSE(Second <= First); - EXPECT_TRUE(Second >= First); + int Values[] = {5, 6, 7}; + + auto First = makePrivatePtr(&Values[0]); + auto SameAsFirst = makePrivatePtr(&Values[0]); + auto Second = makePrivatePtr(&Values[1]); + + EXPECT_TRUE(First == SameAsFirst); + EXPECT_FALSE(First != SameAsFirst); + EXPECT_FALSE(First < SameAsFirst); + EXPECT_FALSE(First > SameAsFirst); + EXPECT_TRUE(First <= SameAsFirst); + EXPECT_TRUE(First >= SameAsFirst); + + EXPECT_FALSE(First == Second); + EXPECT_TRUE(First != Second); + EXPECT_TRUE(First < Second); + EXPECT_FALSE(First > Second); + EXPECT_TRUE(First <= Second); + EXPECT_FALSE(First >= Second); + + EXPECT_FALSE(Second == First); + EXPECT_TRUE(Second != First); + EXPECT_FALSE(Second < First); + EXPECT_TRUE(Second > First); + EXPECT_FALSE(Second <= First); + EXPECT_TRUE(Second >= First); } template void checkNullptrComparisonOperators() { - int Value = 9; - auto NullPtr = private_multi_ptr{nullptr}; - auto NonNullPtr = makePrivatePtr(&Value); - - EXPECT_TRUE(NullPtr == nullptr); - EXPECT_FALSE(NullPtr != nullptr); - EXPECT_FALSE(NullPtr < nullptr); - EXPECT_FALSE(NullPtr > nullptr); - EXPECT_TRUE(NullPtr <= nullptr); - EXPECT_TRUE(NullPtr >= nullptr); - - EXPECT_TRUE(nullptr == NullPtr); - EXPECT_FALSE(nullptr != NullPtr); - EXPECT_FALSE(nullptr < NullPtr); - EXPECT_FALSE(nullptr > NullPtr); - EXPECT_TRUE(nullptr <= NullPtr); - EXPECT_TRUE(nullptr >= NullPtr); - - EXPECT_FALSE(NonNullPtr == nullptr); - EXPECT_TRUE(NonNullPtr != nullptr); - EXPECT_FALSE(nullptr == NonNullPtr); - EXPECT_TRUE(nullptr != NonNullPtr); + int Value = 9; + auto NullPtr = private_multi_ptr{nullptr}; + auto NonNullPtr = makePrivatePtr(&Value); + + EXPECT_TRUE(NullPtr == nullptr); + EXPECT_FALSE(NullPtr != nullptr); + EXPECT_FALSE(NullPtr < nullptr); + EXPECT_FALSE(NullPtr > nullptr); + EXPECT_TRUE(NullPtr <= nullptr); + EXPECT_TRUE(NullPtr >= nullptr); + + EXPECT_TRUE(nullptr == NullPtr); + EXPECT_FALSE(nullptr != NullPtr); + EXPECT_FALSE(nullptr < NullPtr); + EXPECT_FALSE(nullptr > NullPtr); + EXPECT_TRUE(nullptr <= NullPtr); + EXPECT_TRUE(nullptr >= NullPtr); + + EXPECT_FALSE(NonNullPtr == nullptr); + EXPECT_TRUE(NonNullPtr != nullptr); + EXPECT_FALSE(nullptr == NonNullPtr); + EXPECT_TRUE(nullptr != NonNullPtr); } TEST(MultiPtrOperators, OperatorTraitsAreSatisfied) { - checkOperatorTraits(); - checkOperatorTraits(); + checkOperatorTraits(); + checkOperatorTraits(); } TEST(MultiPtrOperators, AccessOperatorsProvidePointerLikeSemantics) { - checkAccessOperators(); - checkAccessOperators(); + checkAccessOperators(); + checkAccessOperators(); } TEST(MultiPtrOperators, ArithmeticOperatorsProvideRandomAccessTraversal) { - checkArithmeticOperators(); - checkArithmeticOperators(); + checkArithmeticOperators(); + checkArithmeticOperators(); } TEST(MultiPtrOperators, RelationalOperatorsCompareUnderlyingLocations) { - checkRelationalOperators(); - checkRelationalOperators(); + checkRelationalOperators(); + checkRelationalOperators(); } TEST(MultiPtrOperators, NullptrComparisonOperatorsHandleNullValuesBothWays) { - checkNullptrComparisonOperators(); - checkNullptrComparisonOperators(); + checkNullptrComparisonOperators(); + checkNullptrComparisonOperators(); } } // namespace - - From a5a991c5db946dbe426141532e944f4d11f21fd2 Mon Sep 17 00:00:00 2001 From: "Klochkov, Denis" Date: Mon, 25 May 2026 17:38:49 +0200 Subject: [PATCH 3/6] [SYCL][TEST] do not use sycl header in tests --- sycl/test-e2e/multi_ptr/prefetch.cpp | 3 ++- sycl/unittests/multi_ptr/Accessors.cpp | 3 ++- sycl/unittests/multi_ptr/Constructors.cpp | 2 +- sycl/unittests/multi_ptr/Conversion.cpp | 2 +- sycl/unittests/multi_ptr/NegativeConversions.cpp | 2 +- sycl/unittests/multi_ptr/Operators.cpp | 2 +- sycl/unittests/multi_ptr/VoidSpecialization.cpp | 2 +- 7 files changed, 9 insertions(+), 7 deletions(-) diff --git a/sycl/test-e2e/multi_ptr/prefetch.cpp b/sycl/test-e2e/multi_ptr/prefetch.cpp index 83a42c5eef304..7aa5d5cf153a1 100644 --- a/sycl/test-e2e/multi_ptr/prefetch.cpp +++ b/sycl/test-e2e/multi_ptr/prefetch.cpp @@ -10,7 +10,8 @@ // //===----------------------------------------------------------------------===// -#include +#include +#include #include #include diff --git a/sycl/unittests/multi_ptr/Accessors.cpp b/sycl/unittests/multi_ptr/Accessors.cpp index 972c96efeebeb..c55bb14989f44 100644 --- a/sycl/unittests/multi_ptr/Accessors.cpp +++ b/sycl/unittests/multi_ptr/Accessors.cpp @@ -1,4 +1,5 @@ -#include +#include +#include #include diff --git a/sycl/unittests/multi_ptr/Constructors.cpp b/sycl/unittests/multi_ptr/Constructors.cpp index 18ac8d4c85ffb..f36a3064c207b 100644 --- a/sycl/unittests/multi_ptr/Constructors.cpp +++ b/sycl/unittests/multi_ptr/Constructors.cpp @@ -6,7 +6,7 @@ // //===----------------------------------------------------------------------===// -#include +#include #include diff --git a/sycl/unittests/multi_ptr/Conversion.cpp b/sycl/unittests/multi_ptr/Conversion.cpp index 555a76a191415..a59082ac5801a 100644 --- a/sycl/unittests/multi_ptr/Conversion.cpp +++ b/sycl/unittests/multi_ptr/Conversion.cpp @@ -1,4 +1,4 @@ -#include +#include #include diff --git a/sycl/unittests/multi_ptr/NegativeConversions.cpp b/sycl/unittests/multi_ptr/NegativeConversions.cpp index ecdfdfa682093..843855c9fecb9 100644 --- a/sycl/unittests/multi_ptr/NegativeConversions.cpp +++ b/sycl/unittests/multi_ptr/NegativeConversions.cpp @@ -6,7 +6,7 @@ // //===----------------------------------------------------------------------===// -#include +#include #include diff --git a/sycl/unittests/multi_ptr/Operators.cpp b/sycl/unittests/multi_ptr/Operators.cpp index 63a296053ac9c..e35dc066f2499 100644 --- a/sycl/unittests/multi_ptr/Operators.cpp +++ b/sycl/unittests/multi_ptr/Operators.cpp @@ -6,7 +6,7 @@ // //===----------------------------------------------------------------------===// -#include +#include #include diff --git a/sycl/unittests/multi_ptr/VoidSpecialization.cpp b/sycl/unittests/multi_ptr/VoidSpecialization.cpp index 3f070d8196e13..ffa7f55678cfd 100644 --- a/sycl/unittests/multi_ptr/VoidSpecialization.cpp +++ b/sycl/unittests/multi_ptr/VoidSpecialization.cpp @@ -6,7 +6,7 @@ // //===----------------------------------------------------------------------===// -#include +#include #include From e6b4ac64d811f62c6a345ff6c200b6d42d11717d Mon Sep 17 00:00:00 2001 From: dklochkov-intel Date: Tue, 26 May 2026 11:12:22 +0200 Subject: [PATCH 4/6] [SYCL][TEST] remove namespace in tests Co-authored-by: Copilot Autofix powered by AI <175728472+Copilot@users.noreply.github.com> --- sycl/unittests/multi_ptr/Accessors.cpp | 4 ---- 1 file changed, 4 deletions(-) diff --git a/sycl/unittests/multi_ptr/Accessors.cpp b/sycl/unittests/multi_ptr/Accessors.cpp index c55bb14989f44..38b38681c58ae 100644 --- a/sycl/unittests/multi_ptr/Accessors.cpp +++ b/sycl/unittests/multi_ptr/Accessors.cpp @@ -5,16 +5,12 @@ #include -namespace { - using sycl::access::address_space; using sycl::access::decorated; template using multi_ptr_t = sycl::multi_ptr; -} // namespace - TEST(MultiPtrAccessors, AccessorDevice) { using rw_acc = sycl::accessor; From 0a3693b7e479658ea66ba6d399d0042516b82105 Mon Sep 17 00:00:00 2001 From: "Klochkov, Denis" Date: Tue, 26 May 2026 11:29:36 +0200 Subject: [PATCH 5/6] [SYCL][TEST] fix PR issues --- sycl/test-e2e/multi_ptr/prefetch.cpp | 18 +++++++----------- .../test/multi_ptr/restricted_conversions.cpp | 1 - .../multi_ptr/NegativeConversions.cpp | 19 ++----------------- .../multi_ptr/VoidSpecialization.cpp | 8 ++------ 4 files changed, 11 insertions(+), 35 deletions(-) diff --git a/sycl/test-e2e/multi_ptr/prefetch.cpp b/sycl/test-e2e/multi_ptr/prefetch.cpp index 7aa5d5cf153a1..c29af8d3c0276 100644 --- a/sycl/test-e2e/multi_ptr/prefetch.cpp +++ b/sycl/test-e2e/multi_ptr/prefetch.cpp @@ -10,14 +10,17 @@ // //===----------------------------------------------------------------------===// +#include +#include +#include #include #include - -#include #include using namespace sycl; +template class PrefetchKernel; + template void testPrefetchWithDecoration() { constexpr size_t Size = 1024; std::vector HostData(Size); @@ -31,7 +34,8 @@ template void testPrefetchWithDecoration() { Q.submit([&](handler &CGH) { auto Acc = Buf.get_access(CGH); - CGH.parallel_for(range<1>(Size), [=](id<1> Index) { + CGH.parallel_for>(range<1>(Size), + [=](id<1> Index) { auto Ptr = Acc.template get_multi_ptr(); // Test prefetch with different element counts @@ -149,10 +153,6 @@ void testPrefetchWithLargeData() { for (size_t i = 0; i < 8; ++i) { Sum += ChunkPtr[i]; } - // Just to use the Sum (avoid optimization removal) - if (Sum < 0) { - ChunkPtr[0] = Sum; - } }); }); @@ -233,10 +233,6 @@ void testPrefetchWithStructs() { for (size_t i = 0; i < 2 && BaseIdx + i < Size; ++i) { Sum += StructPtr[i].A; } - // Use sum to prevent optimization - if (Sum < 0) { - StructPtr[0].A = Sum; - } }); }); diff --git a/sycl/test/multi_ptr/restricted_conversions.cpp b/sycl/test/multi_ptr/restricted_conversions.cpp index 7ca01667dbfe2..ee1e7391736c4 100644 --- a/sycl/test/multi_ptr/restricted_conversions.cpp +++ b/sycl/test/multi_ptr/restricted_conversions.cpp @@ -85,4 +85,3 @@ bool private_equals_local = private_ptr_instance == local; // expected-warning@+1 2 {{'operator int *' is deprecated: Conversion to pointer type is deprecated since SYCL 2020. Please use get() instead.}} bool private_less_than_local = private_ptr_instance < local; - diff --git a/sycl/unittests/multi_ptr/NegativeConversions.cpp b/sycl/unittests/multi_ptr/NegativeConversions.cpp index 843855c9fecb9..0d88942517b5a 100644 --- a/sycl/unittests/multi_ptr/NegativeConversions.cpp +++ b/sycl/unittests/multi_ptr/NegativeConversions.cpp @@ -6,6 +6,7 @@ // //===----------------------------------------------------------------------===// +#include #include #include @@ -74,22 +75,6 @@ TEST(MultiPtrNegativeConversions, EXPECT_FALSE((std::is_convertible_v)); } -TEST(MultiPtrNegativeConversions, CannotConvertGenericToConstantSpace) { - using generic_ptr = - multi_ptr_t; - using constant_ptr = - multi_ptr_t; - - generic_ptr gen_ptr; - constant_ptr const_ptr; - - generic_ptr gen_ptr2{const_ptr}; - constant_ptr const_ptr2{gen_ptr}; - - EXPECT_EQ(gen_ptr2.get(), const_ptr.get()); - EXPECT_EQ(const_ptr2.get(), gen_ptr.get()); -} - TEST(MultiPtrNegativeConversions, CannotConvertVoidToTypedWithoutExplicitCast) { using void_ptr = multi_ptr_t; @@ -124,7 +109,7 @@ TEST(MultiPtrNegativeConversions, GenericCannotAssignFromConstantSpace) { EXPECT_FALSE((std::is_assignable_v)); } -TEST(MultiPtrNegativeConversions, ConstantSpaceNotSupportedForVoidInNonLegacy) { +TEST(MultiPtrNegativeConversions, ConstantSpaceVoidInNonLegacy) { using const_void_constant_legacy = multi_ptr_t; EXPECT_TRUE((std::is_default_constructible_v)); diff --git a/sycl/unittests/multi_ptr/VoidSpecialization.cpp b/sycl/unittests/multi_ptr/VoidSpecialization.cpp index ffa7f55678cfd..254b20de98c82 100644 --- a/sycl/unittests/multi_ptr/VoidSpecialization.cpp +++ b/sycl/unittests/multi_ptr/VoidSpecialization.cpp @@ -7,6 +7,7 @@ //===----------------------------------------------------------------------===// #include +#include #include @@ -216,18 +217,13 @@ template void checkVoidPointerGetMethods() { makePtr(&Data); void_ptr VoidPtr = TypedPtr; - // Check get() returns non-null pointer EXPECT_NE(VoidPtr.get(), nullptr); if constexpr (IsDecorated == decorated::legacy) { - // Legacy has get_raw() method EXPECT_EQ(VoidPtr.get_raw(), static_cast(&Data)); EXPECT_EQ(VoidPtr.get(), VoidPtr.get_decorated()); - } else { - // Non-legacy void specialization doesn't have get_raw() - // Just verify get() works + } else EXPECT_NE(VoidPtr.get(), nullptr); - } } template void checkVoidPointerDecorationConversion() { From 27faad40ac9950824505c2de689955ef80c2f85f Mon Sep 17 00:00:00 2001 From: "Klochkov, Denis" Date: Tue, 26 May 2026 12:31:36 +0200 Subject: [PATCH 6/6] [SYCL][TESTE2E] fix formatting --- sycl/test-e2e/multi_ptr/prefetch.cpp | 50 ++++++++++++++-------------- 1 file changed, 25 insertions(+), 25 deletions(-) diff --git a/sycl/test-e2e/multi_ptr/prefetch.cpp b/sycl/test-e2e/multi_ptr/prefetch.cpp index c29af8d3c0276..221e10ca77a56 100644 --- a/sycl/test-e2e/multi_ptr/prefetch.cpp +++ b/sycl/test-e2e/multi_ptr/prefetch.cpp @@ -34,31 +34,31 @@ template void testPrefetchWithDecoration() { Q.submit([&](handler &CGH) { auto Acc = Buf.get_access(CGH); - CGH.parallel_for>(range<1>(Size), - [=](id<1> Index) { - auto Ptr = Acc.template get_multi_ptr(); - - // Test prefetch with different element counts - if (Index[0] == 0) { - Ptr.prefetch(1); - Ptr.prefetch(16); - Ptr.prefetch(64); - Ptr.prefetch(256); - } - - // Test prefetch at different offsets - if (Index[0] < Size - 100) { - auto OffsetPtr = Ptr + Index[0]; - OffsetPtr.prefetch(10); - } - - // Actual computation to ensure prefetch is useful - int Sum = 0; - for (size_t i = 0; i < 10 && Index[0] + i < Size; ++i) { - Sum += Ptr[Index[0] + i]; - } - Acc[Index] = Sum; - }); + CGH.parallel_for>( + range<1>(Size), [=](id<1> Index) { + auto Ptr = Acc.template get_multi_ptr(); + + // Test prefetch with different element counts + if (Index[0] == 0) { + Ptr.prefetch(1); + Ptr.prefetch(16); + Ptr.prefetch(64); + Ptr.prefetch(256); + } + + // Test prefetch at different offsets + if (Index[0] < Size - 100) { + auto OffsetPtr = Ptr + Index[0]; + OffsetPtr.prefetch(10); + } + + // Actual computation to ensure prefetch is useful + int Sum = 0; + for (size_t i = 0; i < 10 && Index[0] + i < Size; ++i) { + Sum += Ptr[Index[0] + i]; + } + Acc[Index] = Sum; + }); }); Q.wait();