From 943f97d200a23c80c53430c453d3d2ab967a5d80 Mon Sep 17 00:00:00 2001 From: "Hu, Peisen" Date: Wed, 31 Jan 2024 14:55:46 -0800 Subject: [PATCH 1/6] [SYCL] Add logics for aligned_alloc_xxx to return nullptr when the Alignment argument is not a power of 2 Signed-off-by: Hu, Peisen --- sycl/include/sycl/usm.hpp | 72 +++++++++++++++++------- sycl/test-e2e/USM/align.cpp | 108 ++++++++++++++++++++++++++++++++++++ 2 files changed, 160 insertions(+), 20 deletions(-) create mode 100755 sycl/test-e2e/USM/align.cpp diff --git a/sycl/include/sycl/usm.hpp b/sycl/include/sycl/usm.hpp index f9ab634edc81a..54fa09694c37e 100644 --- a/sycl/include/sycl/usm.hpp +++ b/sycl/include/sycl/usm.hpp @@ -179,9 +179,13 @@ T *aligned_alloc_device( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - return static_cast(aligned_alloc_device(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, - PropList, CodeLoc)); + if (Alignment && !(Alignment & (Alignment - 1))) { + return static_cast(aligned_alloc_device(max(Alignment, alignof(T)), + Count * sizeof(T), Dev, Ctxt, + PropList, CodeLoc)); + } else { + return nullptr; + } } template @@ -189,8 +193,12 @@ T *aligned_alloc_device( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - return aligned_alloc_device(Alignment, Count, Q.get_device(), - Q.get_context(), PropList, CodeLoc); + if (Alignment && !(Alignment & (Alignment - 1))) { + return aligned_alloc_device(Alignment, Count, Q.get_device(), + Q.get_context(), PropList, CodeLoc); + } else { + return nullptr; + } } template @@ -230,9 +238,13 @@ T *aligned_alloc_host( size_t Alignment, size_t Count, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - return static_cast(aligned_alloc_host(std ::max(Alignment, alignof(T)), - Count * sizeof(T), Ctxt, PropList, - CodeLoc)); + if (Alignment && !(Alignment & (Alignment - 1))) { + return static_cast(aligned_alloc_host(std ::max(Alignment, alignof(T)), + Count * sizeof(T), Ctxt, + PropList, CodeLoc)); + } else { + return nullptr; + } } template @@ -240,8 +252,12 @@ T *aligned_alloc_host( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - return aligned_alloc_host(Alignment, Count, Q.get_context(), PropList, - CodeLoc); + if (Alignment && !(Alignment & (Alignment - 1))) { + return aligned_alloc_host(Alignment, Count, Q.get_context(), PropList, + CodeLoc); + } else { + return nullptr; + } } template @@ -249,9 +265,13 @@ T *aligned_alloc_shared( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - return static_cast(aligned_alloc_shared(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, - PropList, CodeLoc)); + if (Alignment && !(Alignment & (Alignment - 1))) { + return static_cast(aligned_alloc_shared(max(Alignment, alignof(T)), + Count * sizeof(T), Dev, Ctxt, + PropList, CodeLoc)); + } else { + return nullptr; + } } template @@ -259,8 +279,12 @@ T *aligned_alloc_shared( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - return aligned_alloc_shared(Alignment, Count, Q.get_device(), - Q.get_context(), PropList, CodeLoc); + if (Alignment && !(Alignment & (Alignment - 1))) { + return aligned_alloc_shared(Alignment, Count, Q.get_device(), + Q.get_context(), PropList, CodeLoc); + } else { + return nullptr; + } } template @@ -286,9 +310,13 @@ T *aligned_alloc( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, usm::alloc Kind, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - return static_cast(aligned_alloc(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, Kind, - PropList, CodeLoc)); + if (Alignment && !(Alignment & (Alignment - 1))) { + return static_cast(aligned_alloc(max(Alignment, alignof(T)), + Count * sizeof(T), Dev, Ctxt, Kind, + PropList, CodeLoc)); + } else { + return nullptr; + } } template @@ -296,8 +324,12 @@ T *aligned_alloc( size_t Alignment, size_t Count, const queue &Q, usm::alloc Kind, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - return aligned_alloc(Alignment, Count, Q.get_device(), Q.get_context(), - Kind, PropList, CodeLoc); + if (Alignment && !(Alignment & (Alignment - 1))) { + return aligned_alloc(Alignment, Count, Q.get_device(), Q.get_context(), + Kind, PropList, CodeLoc); + } else { + return nullptr; + } } // Device copy enhancement APIs, prepare_for and release_from USM. diff --git a/sycl/test-e2e/USM/align.cpp b/sycl/test-e2e/USM/align.cpp new file mode 100755 index 0000000000000..796be73a82444 --- /dev/null +++ b/sycl/test-e2e/USM/align.cpp @@ -0,0 +1,108 @@ +// RUN: %{build} -o %t.out +// RUN: %{run} %t.out + +// UNSUPPORTED: gpu + +// E2E tests for annotated USM allocation functions with alignment arguments +// that are not powers of 2. Note this test does not work on gpu because some +// tests expect to return nullptr, e.g. when the alignment argument is not a +// power of 2, while the gpu runtime has different behavior + +#include + +#include +#include + +// clang-format on +using namespace sycl; +using namespace ext::oneapi::experimental; +using namespace ext::intel::experimental; +using alloc = usm::alloc; + +template void testAlign(sycl::queue &q, unsigned align) { + const sycl::context &Ctx = q.get_context(); + auto dev = q.get_device(); + + constexpr int N = 10; + assert(align > 0 || (align & (align - 1)) == 0); + + auto ADevice = [&](size_t align, auto... args) { + return aligned_alloc_device(align, N, args...); + }; + auto AHost = [&](size_t align, auto... args) { + return aligned_alloc_host(align, N, args...); + }; + auto AShared = [&](size_t align, auto... args) { + return aligned_alloc_shared(align, N, args...); + }; + auto AAnnotated = [&](size_t align, auto... args) { + return aligned_alloc(align, N, args...); + }; + + auto ATDevice = [&](size_t align, auto... args) { + return aligned_alloc_device(align, N, args...); + }; + auto ATHost = [&](size_t align, auto... args) { + return aligned_alloc_host(align, N, args...); + }; + auto ATShared = [&](size_t align, auto... args) { + return aligned_alloc_shared(align, N, args...); + }; + auto ATAnnotated = [&](size_t align, auto... args) { + return aligned_alloc(align, N, args...); + }; + + // Test cases that are expected to return null + auto check_null = [&q](auto AllocFn, int Line = __builtin_LINE(), + int Case = 0) { + decltype(AllocFn()) Ptr = AllocFn(); + auto v = reinterpret_cast(Ptr); + if (v != 0) { + free(Ptr, q); + std::cout << "Failed at line " << Line << ", case " << Case << std::endl; + assert(false && "The return is not null!"); + } + }; + + auto CheckNullAll = [&](auto Funcs, int Line = __builtin_LINE()) { + std::apply( + [&](auto... Fs) { + int Case = 0; + (void)std::initializer_list{ + (check_null(Fs, Line, Case++), 0)...}; + }, + Funcs); + }; + + CheckNullAll(std::tuple{ + // Case: aligned_alloc_xxx with no alignment property, and the alignment + // argument is not a power of 2, the result is nullptr + [&]() { return ADevice(3, q); }, [&]() { return ADevice(5, dev, Ctx); }, + [&]() { return AHost(7, q); }, [&]() { return AHost(9, Ctx); }, + [&]() { return AShared(114, q); }, + [&]() { return AShared(1023, dev, Ctx); }, + [&]() { return AAnnotated(15, q, alloc::device); }, + [&]() { return AAnnotated(17, dev, Ctx, alloc::host); } + // Case: aligned_alloc_xxx with no alignment property, and the + // alignment + // argument is not a power of 2, the result is nullptr + , + [&]() { return ATDevice(3, q); }, [&]() { return ATDevice(5, dev, Ctx); }, + [&]() { return ATHost(7, q); }, [&]() { return ATHost(9, Ctx); }, + [&]() { return ATShared(1919, q); }, + [&]() { return ATShared(0, dev, Ctx); }, + [&]() { return ATAnnotated(15, q, alloc::device); }, + [&]() { return ATAnnotated(17, dev, Ctx, alloc::host); }}); + + // aligned_alloc(17, N, q, ...); + // aligned_alloc_device(17, N, q); + // aligned_alloc(17, N, q, alloc::host); +} + +int main() { + sycl::queue q; + testAlign(q, 4); + testAlign(q, 128); + testAlign>(q, 4); + return 0; +} \ No newline at end of file From 40171c6e0deebecf101612d5276e273ca808021e Mon Sep 17 00:00:00 2001 From: "Hu, Peisen" Date: Wed, 31 Jan 2024 16:48:43 -0800 Subject: [PATCH 2/6] [SYCL] Excluding cases with Alignment=0 from nullptr-returning scenario Signed-off-by: Hu, Peisen --- sycl/include/sycl/usm.hpp | 16 ++++++++-------- sycl/test-e2e/USM/align.cpp | 2 +- 2 files changed, 9 insertions(+), 9 deletions(-) diff --git a/sycl/include/sycl/usm.hpp b/sycl/include/sycl/usm.hpp index 54fa09694c37e..4ee726f1ecb1c 100644 --- a/sycl/include/sycl/usm.hpp +++ b/sycl/include/sycl/usm.hpp @@ -179,7 +179,7 @@ T *aligned_alloc_device( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (Alignment && !(Alignment & (Alignment - 1))) { + if (!(Alignment & (Alignment - 1))) { return static_cast(aligned_alloc_device(max(Alignment, alignof(T)), Count * sizeof(T), Dev, Ctxt, PropList, CodeLoc)); @@ -193,7 +193,7 @@ T *aligned_alloc_device( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (Alignment && !(Alignment & (Alignment - 1))) { + if (!(Alignment & (Alignment - 1))) { return aligned_alloc_device(Alignment, Count, Q.get_device(), Q.get_context(), PropList, CodeLoc); } else { @@ -238,7 +238,7 @@ T *aligned_alloc_host( size_t Alignment, size_t Count, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (Alignment && !(Alignment & (Alignment - 1))) { + if (!(Alignment & (Alignment - 1))) { return static_cast(aligned_alloc_host(std ::max(Alignment, alignof(T)), Count * sizeof(T), Ctxt, PropList, CodeLoc)); @@ -252,7 +252,7 @@ T *aligned_alloc_host( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (Alignment && !(Alignment & (Alignment - 1))) { + if (!(Alignment & (Alignment - 1))) { return aligned_alloc_host(Alignment, Count, Q.get_context(), PropList, CodeLoc); } else { @@ -265,7 +265,7 @@ T *aligned_alloc_shared( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (Alignment && !(Alignment & (Alignment - 1))) { + if (!(Alignment & (Alignment - 1))) { return static_cast(aligned_alloc_shared(max(Alignment, alignof(T)), Count * sizeof(T), Dev, Ctxt, PropList, CodeLoc)); @@ -279,7 +279,7 @@ T *aligned_alloc_shared( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (Alignment && !(Alignment & (Alignment - 1))) { + if (!(Alignment & (Alignment - 1))) { return aligned_alloc_shared(Alignment, Count, Q.get_device(), Q.get_context(), PropList, CodeLoc); } else { @@ -310,7 +310,7 @@ T *aligned_alloc( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, usm::alloc Kind, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (Alignment && !(Alignment & (Alignment - 1))) { + if (!(Alignment & (Alignment - 1))) { return static_cast(aligned_alloc(max(Alignment, alignof(T)), Count * sizeof(T), Dev, Ctxt, Kind, PropList, CodeLoc)); @@ -324,7 +324,7 @@ T *aligned_alloc( size_t Alignment, size_t Count, const queue &Q, usm::alloc Kind, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (Alignment && !(Alignment & (Alignment - 1))) { + if (!(Alignment & (Alignment - 1))) { return aligned_alloc(Alignment, Count, Q.get_device(), Q.get_context(), Kind, PropList, CodeLoc); } else { diff --git a/sycl/test-e2e/USM/align.cpp b/sycl/test-e2e/USM/align.cpp index 796be73a82444..07da3a6e81ff4 100755 --- a/sycl/test-e2e/USM/align.cpp +++ b/sycl/test-e2e/USM/align.cpp @@ -90,7 +90,7 @@ template void testAlign(sycl::queue &q, unsigned align) { [&]() { return ATDevice(3, q); }, [&]() { return ATDevice(5, dev, Ctx); }, [&]() { return ATHost(7, q); }, [&]() { return ATHost(9, Ctx); }, [&]() { return ATShared(1919, q); }, - [&]() { return ATShared(0, dev, Ctx); }, + [&]() { return ATShared(11, dev, Ctx); }, [&]() { return ATAnnotated(15, q, alloc::device); }, [&]() { return ATAnnotated(17, dev, Ctx, alloc::host); }}); From 14d8de3b8a7ccb715ee345470c5c90614b8b0a16 Mon Sep 17 00:00:00 2001 From: "Hu, Peisen" Date: Thu, 1 Feb 2024 12:17:25 -0800 Subject: [PATCH 3/6] [SYCL] Addressing Comments in PR Signed-off-by: Hu, Peisen --- sycl/include/sycl/usm.hpp | 72 ++++++++++++++++++------------------- sycl/test-e2e/USM/align.cpp | 11 ++---- 2 files changed, 39 insertions(+), 44 deletions(-) diff --git a/sycl/include/sycl/usm.hpp b/sycl/include/sycl/usm.hpp index 4ee726f1ecb1c..92238b18bc997 100644 --- a/sycl/include/sycl/usm.hpp +++ b/sycl/include/sycl/usm.hpp @@ -154,6 +154,14 @@ __SYCL_EXPORT void *aligned_alloc( const property_list &propList, const detail::code_location &CodeLoc = detail::code_location::current()); +/// +// Helper function used to determine if the Alignment argument is a power of 2 +/// +inline size_t is_not_power_of_two (size_t Alignment) { + return (Alignment & (Alignment - 1)); +} + + /// // Template forms /// @@ -179,13 +187,12 @@ T *aligned_alloc_device( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (!(Alignment & (Alignment - 1))) { - return static_cast(aligned_alloc_device(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, - PropList, CodeLoc)); - } else { + if (is_not_power_of_two(Alignment)) { return nullptr; } + return static_cast(aligned_alloc_device(max(Alignment, alignof(T)), + Count * sizeof(T), Dev, Ctxt, + PropList, CodeLoc)); } template @@ -193,12 +200,11 @@ T *aligned_alloc_device( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (!(Alignment & (Alignment - 1))) { - return aligned_alloc_device(Alignment, Count, Q.get_device(), - Q.get_context(), PropList, CodeLoc); - } else { + if (is_not_power_of_two(Alignment)) { return nullptr; } + return aligned_alloc_device(Alignment, Count, Q.get_device(), + Q.get_context(), PropList, CodeLoc); } template @@ -238,13 +244,12 @@ T *aligned_alloc_host( size_t Alignment, size_t Count, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (!(Alignment & (Alignment - 1))) { - return static_cast(aligned_alloc_host(std ::max(Alignment, alignof(T)), - Count * sizeof(T), Ctxt, - PropList, CodeLoc)); - } else { + if (is_not_power_of_two(Alignment)) { return nullptr; } + return static_cast(aligned_alloc_host(std ::max(Alignment, alignof(T)), + Count * sizeof(T), Ctxt, + PropList, CodeLoc)); } template @@ -252,12 +257,11 @@ T *aligned_alloc_host( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (!(Alignment & (Alignment - 1))) { - return aligned_alloc_host(Alignment, Count, Q.get_context(), PropList, - CodeLoc); - } else { + if (is_not_power_of_two(Alignment)) { return nullptr; } + return aligned_alloc_host(Alignment, Count, Q.get_context(), PropList, + CodeLoc); } template @@ -265,13 +269,12 @@ T *aligned_alloc_shared( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (!(Alignment & (Alignment - 1))) { - return static_cast(aligned_alloc_shared(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, - PropList, CodeLoc)); - } else { + if (is_not_power_of_two(Alignment)) { return nullptr; } + return static_cast(aligned_alloc_shared(max(Alignment, alignof(T)), + Count * sizeof(T), Dev, Ctxt, + PropList, CodeLoc)); } template @@ -279,12 +282,11 @@ T *aligned_alloc_shared( size_t Alignment, size_t Count, const queue &Q, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (!(Alignment & (Alignment - 1))) { - return aligned_alloc_shared(Alignment, Count, Q.get_device(), - Q.get_context(), PropList, CodeLoc); - } else { + if (is_not_power_of_two(Alignment)) { return nullptr; } + return aligned_alloc_shared(Alignment, Count, Q.get_device(), + Q.get_context(), PropList, CodeLoc); } template @@ -310,13 +312,12 @@ T *aligned_alloc( size_t Alignment, size_t Count, const device &Dev, const context &Ctxt, usm::alloc Kind, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (!(Alignment & (Alignment - 1))) { - return static_cast(aligned_alloc(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, Kind, - PropList, CodeLoc)); - } else { + if (is_not_power_of_two(Alignment)) { return nullptr; } + return static_cast(aligned_alloc(max(Alignment, alignof(T)), + Count * sizeof(T), Dev, Ctxt, Kind, + PropList, CodeLoc)); } template @@ -324,12 +325,11 @@ T *aligned_alloc( size_t Alignment, size_t Count, const queue &Q, usm::alloc Kind, const property_list &PropList = {}, const detail::code_location &CodeLoc = detail::code_location::current()) { - if (!(Alignment & (Alignment - 1))) { - return aligned_alloc(Alignment, Count, Q.get_device(), Q.get_context(), - Kind, PropList, CodeLoc); - } else { + if (is_not_power_of_two(Alignment)) { return nullptr; } + return aligned_alloc(Alignment, Count, Q.get_device(), Q.get_context(), + Kind, PropList, CodeLoc); } // Device copy enhancement APIs, prepare_for and release_from USM. diff --git a/sycl/test-e2e/USM/align.cpp b/sycl/test-e2e/USM/align.cpp index 07da3a6e81ff4..644bc114f4275 100755 --- a/sycl/test-e2e/USM/align.cpp +++ b/sycl/test-e2e/USM/align.cpp @@ -13,7 +13,6 @@ #include #include -// clang-format on using namespace sycl; using namespace ext::oneapi::experimental; using namespace ext::intel::experimental; @@ -53,8 +52,8 @@ template void testAlign(sycl::queue &q, unsigned align) { }; // Test cases that are expected to return null - auto check_null = [&q](auto AllocFn, int Line = __builtin_LINE(), - int Case = 0) { + auto check_null = [&q](auto AllocFn, int Line, + int Case) { decltype(AllocFn()) Ptr = AllocFn(); auto v = reinterpret_cast(Ptr); if (v != 0) { @@ -93,10 +92,6 @@ template void testAlign(sycl::queue &q, unsigned align) { [&]() { return ATShared(11, dev, Ctx); }, [&]() { return ATAnnotated(15, q, alloc::device); }, [&]() { return ATAnnotated(17, dev, Ctx, alloc::host); }}); - - // aligned_alloc(17, N, q, ...); - // aligned_alloc_device(17, N, q); - // aligned_alloc(17, N, q, alloc::host); } int main() { @@ -105,4 +100,4 @@ int main() { testAlign(q, 128); testAlign>(q, 4); return 0; -} \ No newline at end of file +} From 04b71d2c6db88150d4770c6e4b5a9c0f065954e4 Mon Sep 17 00:00:00 2001 From: "Hu, Peisen" Date: Thu, 1 Feb 2024 13:18:15 -0800 Subject: [PATCH 4/6] [SYCL] Formatting files and removing outdated comments in align.cpp Signed-off-by: Hu, Peisen --- sycl/include/sycl/usm.hpp | 27 +++++++++++++-------------- sycl/test-e2e/USM/align.cpp | 9 ++------- 2 files changed, 15 insertions(+), 21 deletions(-) diff --git a/sycl/include/sycl/usm.hpp b/sycl/include/sycl/usm.hpp index 92238b18bc997..ff941e865dff8 100644 --- a/sycl/include/sycl/usm.hpp +++ b/sycl/include/sycl/usm.hpp @@ -157,11 +157,10 @@ __SYCL_EXPORT void *aligned_alloc( /// // Helper function used to determine if the Alignment argument is a power of 2 /// -inline size_t is_not_power_of_two (size_t Alignment) { +inline size_t is_not_power_of_two(size_t Alignment) { return (Alignment & (Alignment - 1)); } - /// // Template forms /// @@ -191,8 +190,8 @@ T *aligned_alloc_device( return nullptr; } return static_cast(aligned_alloc_device(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, - PropList, CodeLoc)); + Count * sizeof(T), Dev, Ctxt, + PropList, CodeLoc)); } template @@ -204,7 +203,7 @@ T *aligned_alloc_device( return nullptr; } return aligned_alloc_device(Alignment, Count, Q.get_device(), - Q.get_context(), PropList, CodeLoc); + Q.get_context(), PropList, CodeLoc); } template @@ -248,8 +247,8 @@ T *aligned_alloc_host( return nullptr; } return static_cast(aligned_alloc_host(std ::max(Alignment, alignof(T)), - Count * sizeof(T), Ctxt, - PropList, CodeLoc)); + Count * sizeof(T), Ctxt, PropList, + CodeLoc)); } template @@ -261,7 +260,7 @@ T *aligned_alloc_host( return nullptr; } return aligned_alloc_host(Alignment, Count, Q.get_context(), PropList, - CodeLoc); + CodeLoc); } template @@ -273,8 +272,8 @@ T *aligned_alloc_shared( return nullptr; } return static_cast(aligned_alloc_shared(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, - PropList, CodeLoc)); + Count * sizeof(T), Dev, Ctxt, + PropList, CodeLoc)); } template @@ -286,7 +285,7 @@ T *aligned_alloc_shared( return nullptr; } return aligned_alloc_shared(Alignment, Count, Q.get_device(), - Q.get_context(), PropList, CodeLoc); + Q.get_context(), PropList, CodeLoc); } template @@ -316,8 +315,8 @@ T *aligned_alloc( return nullptr; } return static_cast(aligned_alloc(max(Alignment, alignof(T)), - Count * sizeof(T), Dev, Ctxt, Kind, - PropList, CodeLoc)); + Count * sizeof(T), Dev, Ctxt, Kind, + PropList, CodeLoc)); } template @@ -329,7 +328,7 @@ T *aligned_alloc( return nullptr; } return aligned_alloc(Alignment, Count, Q.get_device(), Q.get_context(), - Kind, PropList, CodeLoc); + Kind, PropList, CodeLoc); } // Device copy enhancement APIs, prepare_for and release_from USM. diff --git a/sycl/test-e2e/USM/align.cpp b/sycl/test-e2e/USM/align.cpp index 644bc114f4275..c1da3173f0edc 100755 --- a/sycl/test-e2e/USM/align.cpp +++ b/sycl/test-e2e/USM/align.cpp @@ -1,12 +1,8 @@ // RUN: %{build} -o %t.out // RUN: %{run} %t.out -// UNSUPPORTED: gpu - // E2E tests for annotated USM allocation functions with alignment arguments -// that are not powers of 2. Note this test does not work on gpu because some -// tests expect to return nullptr, e.g. when the alignment argument is not a -// power of 2, while the gpu runtime has different behavior +// that are not powers of 2. #include @@ -52,8 +48,7 @@ template void testAlign(sycl::queue &q, unsigned align) { }; // Test cases that are expected to return null - auto check_null = [&q](auto AllocFn, int Line, - int Case) { + auto check_null = [&q](auto AllocFn, int Line, int Case) { decltype(AllocFn()) Ptr = AllocFn(); auto v = reinterpret_cast(Ptr); if (v != 0) { From e3e899325db2db0e50ff55e74d971376f98fdb96 Mon Sep 17 00:00:00 2001 From: "Hu, Peisen" Date: Thu, 1 Feb 2024 15:40:10 -0800 Subject: [PATCH 5/6] [SYCL] Add back comments indicating GPU is unsupported for test align.cpp Signed-off-by: Hu, Peisen --- sycl/test-e2e/USM/align.cpp | 6 +++++- 1 file changed, 5 insertions(+), 1 deletion(-) diff --git a/sycl/test-e2e/USM/align.cpp b/sycl/test-e2e/USM/align.cpp index c1da3173f0edc..220f3e97f91e5 100755 --- a/sycl/test-e2e/USM/align.cpp +++ b/sycl/test-e2e/USM/align.cpp @@ -1,8 +1,12 @@ // RUN: %{build} -o %t.out // RUN: %{run} %t.out +// UNSUPPORTED: gpu + // E2E tests for annotated USM allocation functions with alignment arguments -// that are not powers of 2. +// that are not powers of 2. Note this test does not work on gpu because some +// tests expect to return nullptr, e.g. when the alignment argument is not a +// power of 2, while the gpu runtime has different behavior #include From 532c706ba8794d8ed427270e9785c60a3d3ad0ed Mon Sep 17 00:00:00 2001 From: "Hu, Peisen" Date: Tue, 6 Feb 2024 16:02:14 -0800 Subject: [PATCH 6/6] [SYCL] Refining align.cpp Signed-off-by: Hu, Peisen --- sycl/test-e2e/USM/align.cpp | 12 ++++++------ 1 file changed, 6 insertions(+), 6 deletions(-) diff --git a/sycl/test-e2e/USM/align.cpp b/sycl/test-e2e/USM/align.cpp index 220f3e97f91e5..9659ec78b9a88 100755 --- a/sycl/test-e2e/USM/align.cpp +++ b/sycl/test-e2e/USM/align.cpp @@ -5,8 +5,10 @@ // E2E tests for annotated USM allocation functions with alignment arguments // that are not powers of 2. Note this test does not work on gpu because some -// tests expect to return nullptr, e.g. when the alignment argument is not a -// power of 2, while the gpu runtime has different behavior +// tests expect non-templated aligned_alloc_xxx functions to return nullptr, +// e.g. when the alignment argument is not a power of 2, while they fail to do +// so when run on gpu. This maybe because the gpu runtime has different +// behavior. Therefore, GPU is unsupported until issue #12638 gets resolved. #include @@ -54,8 +56,7 @@ template void testAlign(sycl::queue &q, unsigned align) { // Test cases that are expected to return null auto check_null = [&q](auto AllocFn, int Line, int Case) { decltype(AllocFn()) Ptr = AllocFn(); - auto v = reinterpret_cast(Ptr); - if (v != 0) { + if (Ptr != nullptr) { free(Ptr, q); std::cout << "Failed at line " << Line << ", case " << Case << std::endl; assert(false && "The return is not null!"); @@ -82,8 +83,7 @@ template void testAlign(sycl::queue &q, unsigned align) { [&]() { return AAnnotated(15, q, alloc::device); }, [&]() { return AAnnotated(17, dev, Ctx, alloc::host); } // Case: aligned_alloc_xxx with no alignment property, and the - // alignment - // argument is not a power of 2, the result is nullptr + // alignment argument is not a power of 2, the result is nullptr , [&]() { return ATDevice(3, q); }, [&]() { return ATDevice(5, dev, Ctx); }, [&]() { return ATHost(7, q); }, [&]() { return ATHost(9, Ctx); },