[llvm] [libsycl] USM Aligned allocation functions (PR #213468)
via llvm-commits
llvm-commits at lists.llvm.org
Sun Aug 16 04:01:51 PDT 2026
https://github.com/Robertkq updated https://github.com/llvm/llvm-project/pull/213468
>From b73c8b586637d66cb4793c76e69ffc56b6a7c7d7 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Sun, 26 Jul 2026 15:20:17 +0300
Subject: [PATCH 01/13] Add API of aligned version of USM alloc free functions
---
libsycl/include/sycl/__impl/usm_functions.hpp | 185 ++++++++++++++++--
1 file changed, 173 insertions(+), 12 deletions(-)
diff --git a/libsycl/include/sycl/__impl/usm_functions.hpp b/libsycl/include/sycl/__impl/usm_functions.hpp
index 10bd4a6dddc8e..793fe2f8d5061 100644
--- a/libsycl/include/sycl/__impl/usm_functions.hpp
+++ b/libsycl/include/sycl/__impl/usm_functions.hpp
@@ -86,22 +86,80 @@ T *malloc_device(std::size_t count, const queue &syclQueue,
return malloc_device<T>(count, syclQueue.get_device(),
syclQueue.get_context(), propList);
}
-/// @}
-/// \name SYCL 2020 4.8.3.3. Host allocation functions.
-/// \brief Allocations in host memory are accessible by a device.
-/// @{
-/// Allocates host USM.
+/// Allocates device USM with specified alignment.
///
+/// \param alignment the alignment of the allocated memory.
/// \param numBytes the number of bytes to allocate.
-/// \param syclContext the context that should have access to the allocated
-/// memory.
+/// \param syclDevice the device to use for the allocation.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak.
-_LIBSYCL_EXPORT void *malloc_host(std::size_t numBytes,
- const context &syclContext,
- const property_list &propList = {});
+/// \return a pointer to the newly allocated memory, which is allocated on
+/// syclDevice and which must eventually be deallocated with sycl::free in order
+/// to avoid a memory leak.
+void *aligned_alloc_device(size_t alignment, size_t numBytes,
+ const device &syclDevice, const context &syclContext,
+ const property_list &propList = {});
+
+/// Allocates device USM with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param count the number of elements of type T to allocate.
+/// \param syclDevice the device to use for the allocation.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which is allocated on
+/// syclDevice and which must eventually be deallocated with sycl::free in order
+/// to avoid a memory leak.
+template <typename T>
+T *aligned_alloc_device(size_t alignment, size_t count,
+ const device &syclDevice, const context &syclContext,
+ const property_list &propList = {});
+
+/// Allocates device USM with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param numBytes the number of bytes to allocate.
+/// \param syclQueue a queue that provides the device and context.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which is allocated on
+/// syclDevice and which must eventually be deallocated with sycl::free in order
+/// to avoid a memory leak.
+void *aligned_alloc_device(size_t alignment, size_t numBytes,
+ const queue &syclQueue,
+ const property_list &propList = {});
+
+/// Allocates device USM with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param count the number of elements of type T to allocate.
+/// \param syclQueue a queue that provides the device and context.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which is allocated on
+/// syclDevice and which must eventually be deallocated with sycl::free in order
+/// to avoid a memory leak.
+template <typename T>
+T *aligned_alloc_device(size_t alignment, size_t count, const queue &syclQueue,
+ const property_list &propList = {})
+
+ /// @}
+
+ /// \name SYCL 2020 4.8.3.3. Host allocation functions.
+ /// \brief Allocations in host memory are accessible by a device.
+ /// @{
+ /// Allocates host USM.
+ ///
+ /// \param numBytes the number of bytes to allocate.
+ /// \param syclContext the context that should have access to the allocated
+ /// memory.
+ /// \param propList the list of properties for the allocation.
+ /// \return a pointer to the newly allocated memory, which must eventually
+ /// be deallocated with sycl::free in order to avoid a memory leak.
+ _LIBSYCL_EXPORT
+ void *malloc_host(std::size_t numBytes, const context &syclContext,
+ const property_list &propList = {});
/// Allocates host USM.
///
@@ -142,6 +200,58 @@ T *malloc_host(std::size_t count, const queue &syclQueue,
const property_list &propList = {}) {
return malloc_host<T>(count, syclQueue.get_context(), propList);
}
+
+/// Allocates host USM with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param numBytes the number of bytes to allocate.
+/// \param syclContext the context that should have access to the allocated
+/// memory.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak.
+void *aligned_alloc_host(size_t alignment, size_t numBytes,
+ const context &syclContext,
+ const property_list &propList = {});
+
+/// Allocates host USM with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param count the number of elements of type T to allocate.
+/// \param syclContext the context that should have access to the allocated
+/// memory.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak.
+template <typename T>
+T *aligned_alloc_host(size_t alignment, size_t count,
+ const context &syclContext,
+ const property_list &propList = {});
+
+/// Allocates host USM with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param numBytes the number of bytes to allocate.
+/// \param syclQueue queue that provides the context.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak.
+void *aligned_alloc_host(size_t alignment, size_t numBytes,
+ const queue &syclQueue,
+ const property_list &propList = {});
+
+/// Allocates host USM with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param count the number of elements of type T to allocate.
+/// \param syclQueue queue that provides the context.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak.
+template <typename T>
+T *aligned_alloc_host(size_t alignment, size_t count, const queue &syclQueue,
+ const property_list &propList = {});
+
/// @}
/// \name SYCL 2020 4.8.3.4. Shared allocation functions.
@@ -204,6 +314,57 @@ T *malloc_shared(std::size_t count, const queue &syclQueue,
return malloc_shared<T>(count, syclQueue.get_device(),
syclQueue.get_context(), propList);
}
+
+/// Allocates shared USM with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param numBytes the number of bytes to allocate.
+/// \param syclDevice the device to use for the allocation.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak.
+void *aligned_alloc_shared(size_t alignment, size_t numBytes,
+ const device &syclDevice, const context &syclContext,
+ const property_list &propList = {});
+
+/// Allocates shared USM with specified alignment.
+/// \param alignment the alignment of the allocated memory.
+/// \param count the number of elements of type T to allocate.
+/// \param syclDevice the device to use for the allocation.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak.
+template <typename T>
+T *aligned_alloc_shared(size_t alignment, size_t count,
+ const device &syclDevice, const context &syclContext,
+ const property_list &propList = {});
+
+/// Allocates shared USM with specified alignment.
+/// \param alignment the alignment of the allocated memory.
+/// \param numBytes the number of bytes to allocate.
+/// \param syclQueue a queue that provides the device and context.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak.
+void *aligned_alloc_shared(size_t alignment, size_t numBytes,
+ const queue &syclQueue,
+ const property_list &propList = {});
+
+/// Allocates shared USM with specified alignment.
+/// \param alignment the alignment of the allocated memory.
+/// \param count the number of elements of type T to allocate.
+/// \param syclQueue a queue that provides the device and context.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak.
+template <typename T>
+T *aligned_alloc_shared(size_t alignment, size_t count, const queue &syclQueue,
+ const property_list &propList = {});
+
/// @}
/// \name SYCL 2020 4.8.3.5. Parameterized allocation functions.
>From d0b015cefd951c26eb80041e6c1ec0b07a55950a Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Mon, 27 Jul 2026 19:47:57 +0300
Subject: [PATCH 02/13] resolve overloads
---
libsycl/include/sycl/__impl/usm_functions.hpp | 67 ++++++++++++-------
libsycl/src/usm_functions.cpp | 39 +++++++++++
2 files changed, 83 insertions(+), 23 deletions(-)
diff --git a/libsycl/include/sycl/__impl/usm_functions.hpp b/libsycl/include/sycl/__impl/usm_functions.hpp
index 793fe2f8d5061..d12c42287bb09 100644
--- a/libsycl/include/sycl/__impl/usm_functions.hpp
+++ b/libsycl/include/sycl/__impl/usm_functions.hpp
@@ -116,7 +116,10 @@ void *aligned_alloc_device(size_t alignment, size_t numBytes,
template <typename T>
T *aligned_alloc_device(size_t alignment, size_t count,
const device &syclDevice, const context &syclContext,
- const property_list &propList = {});
+ const property_list &propList = {}) {
+ return static_cast<T *>(aligned_alloc_device(
+ alignment, count * sizeof(T), syclDevice, syclContext, propList));
+}
/// Allocates device USM with specified alignment.
///
@@ -142,24 +145,27 @@ void *aligned_alloc_device(size_t alignment, size_t numBytes,
/// to avoid a memory leak.
template <typename T>
T *aligned_alloc_device(size_t alignment, size_t count, const queue &syclQueue,
- const property_list &propList = {})
-
- /// @}
-
- /// \name SYCL 2020 4.8.3.3. Host allocation functions.
- /// \brief Allocations in host memory are accessible by a device.
- /// @{
- /// Allocates host USM.
- ///
- /// \param numBytes the number of bytes to allocate.
- /// \param syclContext the context that should have access to the allocated
- /// memory.
- /// \param propList the list of properties for the allocation.
- /// \return a pointer to the newly allocated memory, which must eventually
- /// be deallocated with sycl::free in order to avoid a memory leak.
- _LIBSYCL_EXPORT
- void *malloc_host(std::size_t numBytes, const context &syclContext,
- const property_list &propList = {});
+ const property_list &propList = {}) {
+ return alligned_alloc_device<T>(alignment, count, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
+}
+
+/// @}
+
+/// \name SYCL 2020 4.8.3.3. Host allocation functions.
+/// \brief Allocations in host memory are accessible by a device.
+/// @{
+/// Allocates host USM.
+///
+/// \param numBytes the number of bytes to allocate.
+/// \param syclContext the context that should have access to the allocated
+/// memory.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually
+/// be deallocated with sycl::free in order to avoid a memory leak.
+_LIBSYCL_EXPORT
+void *malloc_host(std::size_t numBytes, const context &syclContext,
+ const property_list &propList = {});
/// Allocates host USM.
///
@@ -226,7 +232,10 @@ void *aligned_alloc_host(size_t alignment, size_t numBytes,
template <typename T>
T *aligned_alloc_host(size_t alignment, size_t count,
const context &syclContext,
- const property_list &propList = {});
+ const property_list &propList = {}) {
+ return static_cast<T *>(
+ aligned_alloc_host(alignment, count * sizeof(T), syclContext, propList));
+}
/// Allocates host USM with specified alignment.
///
@@ -250,7 +259,10 @@ void *aligned_alloc_host(size_t alignment, size_t numBytes,
/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
T *aligned_alloc_host(size_t alignment, size_t count, const queue &syclQueue,
- const property_list &propList = {});
+ const property_list &propList = {}) {
+ return aligned_alloc_host<T>(alignment, count, syclQueue.get_context(),
+ propList);
+}
/// @}
@@ -330,6 +342,7 @@ void *aligned_alloc_shared(size_t alignment, size_t numBytes,
const property_list &propList = {});
/// Allocates shared USM with specified alignment.
+///
/// \param alignment the alignment of the allocated memory.
/// \param count the number of elements of type T to allocate.
/// \param syclDevice the device to use for the allocation.
@@ -341,9 +354,13 @@ void *aligned_alloc_shared(size_t alignment, size_t numBytes,
template <typename T>
T *aligned_alloc_shared(size_t alignment, size_t count,
const device &syclDevice, const context &syclContext,
- const property_list &propList = {});
+ const property_list &propList = {}) {
+ return static_cast<T *>(aligned_alloc_shared(
+ alignment, count * sizeof(T), syclDevice, syclContext, propList));
+}
/// Allocates shared USM with specified alignment.
+///
/// \param alignment the alignment of the allocated memory.
/// \param numBytes the number of bytes to allocate.
/// \param syclQueue a queue that provides the device and context.
@@ -355,6 +372,7 @@ void *aligned_alloc_shared(size_t alignment, size_t numBytes,
const property_list &propList = {});
/// Allocates shared USM with specified alignment.
+///
/// \param alignment the alignment of the allocated memory.
/// \param count the number of elements of type T to allocate.
/// \param syclQueue a queue that provides the device and context.
@@ -363,7 +381,10 @@ void *aligned_alloc_shared(size_t alignment, size_t numBytes,
/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
T *aligned_alloc_shared(size_t alignment, size_t count, const queue &syclQueue,
- const property_list &propList = {});
+ const property_list &propList = {}) {
+ return aligned_alloc_shared<T>(alignment, count, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
+}
/// @}
diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp
index 0f921e055f9a2..4ff3d0d916061 100644
--- a/libsycl/src/usm_functions.cpp
+++ b/libsycl/src/usm_functions.cpp
@@ -31,6 +31,19 @@ void *malloc_device(std::size_t numBytes, const queue &syclQueue,
syclQueue.get_context(), propList);
}
+void *aligned_alloc_device(size_t alignment, size_t numBytes,
+ const device &syclDevice, const context &syclContext,
+ const property_list &propList = {}) {
+ // TODO: implementation here with malloc
+}
+
+void *aligned_alloc_device(size_t alignment, size_t numBytes,
+ const queue &syclQueue,
+ const property_list &propList = {}) {
+ return aligned_alloc_device(alignment, numBytes, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
+}
+
// SYCL 2020 4.8.3.3. Host allocation functions.
void *malloc_host(std::size_t numBytes, const context &syclContext,
@@ -52,6 +65,19 @@ void *malloc_host(std::size_t numBytes, const queue &syclQueue,
return malloc_host(numBytes, syclQueue.get_context(), propList);
}
+void *aligned_alloc_host(size_t alignment, size_t numBytes,
+ const context &syclContext,
+ const property_list &propList = {}) {
+ // TODO: implementation here with malloc
+}
+
+void *aligned_alloc_host(size_t alignment, size_t numBytes,
+ const queue &syclQueue,
+ const property_list &propList = {}) {
+ return aligned_alloc_host(alignment, numBytes, syclQueue.get_context(),
+ propList);
+}
+
// SYCL 2020 4.8.3.4. Shared allocation functions.
void *malloc_shared(std::size_t numBytes, const device &syclDevice,
@@ -66,6 +92,19 @@ void *malloc_shared(std::size_t numBytes, const queue &syclQueue,
syclQueue.get_context(), propList);
}
+void *aligned_alloc_shared(size_t alignment, size_t numBytes,
+ const device &syclDevice, const context &syclContext,
+ const property_list &propList = {}) {
+ // TODO: implementation here with malloc
+}
+
+void *aligned_alloc_shared(size_t alignment, size_t numBytes,
+ const queue &syclQueue,
+ const property_list &propList = {}) {
+ return aligned_alloc_shared(alignment, numBytes, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
+}
+
// SYCL 2020 4.8.3.5. Parameterized allocation functions.
static aspect getAspectByAllocationKind(usm::alloc kind) {
>From 46ec842e3aee90bbb0e1047f2fe15d820572b920 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Mon, 27 Jul 2026 20:26:58 +0300
Subject: [PATCH 03/13] Add aligned_alloc functions in API & cleanup
---
libsycl/include/sycl/__impl/usm_functions.hpp | 109 ++++++++++++++++--
libsycl/src/usm_functions.cpp | 13 +++
2 files changed, 110 insertions(+), 12 deletions(-)
diff --git a/libsycl/include/sycl/__impl/usm_functions.hpp b/libsycl/include/sycl/__impl/usm_functions.hpp
index d12c42287bb09..78847a5828f70 100644
--- a/libsycl/include/sycl/__impl/usm_functions.hpp
+++ b/libsycl/include/sycl/__impl/usm_functions.hpp
@@ -98,7 +98,8 @@ T *malloc_device(std::size_t count, const queue &syclQueue,
/// \return a pointer to the newly allocated memory, which is allocated on
/// syclDevice and which must eventually be deallocated with sycl::free in order
/// to avoid a memory leak.
-void *aligned_alloc_device(size_t alignment, size_t numBytes,
+_LIBSYCL_EXPORT
+void *aligned_alloc_device(std::size_t alignment, std::size_t numBytes,
const device &syclDevice, const context &syclContext,
const property_list &propList = {});
@@ -114,7 +115,7 @@ void *aligned_alloc_device(size_t alignment, size_t numBytes,
/// syclDevice and which must eventually be deallocated with sycl::free in order
/// to avoid a memory leak.
template <typename T>
-T *aligned_alloc_device(size_t alignment, size_t count,
+T *aligned_alloc_device(std::size_t alignment, std::size_t count,
const device &syclDevice, const context &syclContext,
const property_list &propList = {}) {
return static_cast<T *>(aligned_alloc_device(
@@ -130,7 +131,8 @@ T *aligned_alloc_device(size_t alignment, size_t count,
/// \return a pointer to the newly allocated memory, which is allocated on
/// syclDevice and which must eventually be deallocated with sycl::free in order
/// to avoid a memory leak.
-void *aligned_alloc_device(size_t alignment, size_t numBytes,
+_LIBSYCL_EXPORT
+void *aligned_alloc_device(std::size_t alignment, std::size_t numBytes,
const queue &syclQueue,
const property_list &propList = {});
@@ -144,7 +146,8 @@ void *aligned_alloc_device(size_t alignment, size_t numBytes,
/// syclDevice and which must eventually be deallocated with sycl::free in order
/// to avoid a memory leak.
template <typename T>
-T *aligned_alloc_device(size_t alignment, size_t count, const queue &syclQueue,
+T *aligned_alloc_device(std::size_t alignment, std::size_t count,
+ const queue &syclQueue,
const property_list &propList = {}) {
return alligned_alloc_device<T>(alignment, count, syclQueue.get_device(),
syclQueue.get_context(), propList);
@@ -216,7 +219,8 @@ T *malloc_host(std::size_t count, const queue &syclQueue,
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
-void *aligned_alloc_host(size_t alignment, size_t numBytes,
+_LIBSYCL_EXPORT
+void *aligned_alloc_host(std::size_t alignment, std::size_t numBytes,
const context &syclContext,
const property_list &propList = {});
@@ -230,7 +234,7 @@ void *aligned_alloc_host(size_t alignment, size_t numBytes,
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
-T *aligned_alloc_host(size_t alignment, size_t count,
+T *aligned_alloc_host(std::size_t alignment, std::size_t count,
const context &syclContext,
const property_list &propList = {}) {
return static_cast<T *>(
@@ -245,7 +249,8 @@ T *aligned_alloc_host(size_t alignment, size_t count,
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
-void *aligned_alloc_host(size_t alignment, size_t numBytes,
+_LIBSYCL_EXPORT
+void *aligned_alloc_host(std::size_t alignment, std::size_t numBytes,
const queue &syclQueue,
const property_list &propList = {});
@@ -258,7 +263,8 @@ void *aligned_alloc_host(size_t alignment, size_t numBytes,
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
-T *aligned_alloc_host(size_t alignment, size_t count, const queue &syclQueue,
+T *aligned_alloc_host(std::size_t alignment, std::size_t count,
+ const queue &syclQueue,
const property_list &propList = {}) {
return aligned_alloc_host<T>(alignment, count, syclQueue.get_context(),
propList);
@@ -337,7 +343,8 @@ T *malloc_shared(std::size_t count, const queue &syclQueue,
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
-void *aligned_alloc_shared(size_t alignment, size_t numBytes,
+_LIBSYCL_EXPORT
+void *aligned_alloc_shared(std::size_t alignment, std::size_t numBytes,
const device &syclDevice, const context &syclContext,
const property_list &propList = {});
@@ -352,7 +359,7 @@ void *aligned_alloc_shared(size_t alignment, size_t numBytes,
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
-T *aligned_alloc_shared(size_t alignment, size_t count,
+T *aligned_alloc_shared(std::size_t alignment, std::size_t count,
const device &syclDevice, const context &syclContext,
const property_list &propList = {}) {
return static_cast<T *>(aligned_alloc_shared(
@@ -367,7 +374,8 @@ T *aligned_alloc_shared(size_t alignment, size_t count,
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
-void *aligned_alloc_shared(size_t alignment, size_t numBytes,
+_LIBSYCL_EXPORT
+void *aligned_alloc_shared(std::size_t alignment, std::size_t numBytes,
const queue &syclQueue,
const property_list &propList = {});
@@ -380,7 +388,8 @@ void *aligned_alloc_shared(size_t alignment, size_t numBytes,
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
-T *aligned_alloc_shared(size_t alignment, size_t count, const queue &syclQueue,
+T *aligned_alloc_shared(std::size_t alignment, std::size_t count,
+ const queue &syclQueue,
const property_list &propList = {}) {
return aligned_alloc_shared<T>(alignment, count, syclQueue.get_device(),
syclQueue.get_context(), propList);
@@ -459,6 +468,82 @@ T *malloc(std::size_t count, const queue &syclQueue, usm::alloc kind,
return malloc<T>(count, syclQueue.get_device(), syclQueue.get_context(), kind,
propList);
}
+
+/// Allocates USM of type `kind` with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param numBytes the number of bytes to allocate.
+/// \param syclDevice the device to use for the allocation. The syclDevice
+/// parameter is ignored if kind is usm::alloc::host.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
+/// \param kind the type of memory to allocate.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak. If there are
+/// not enough resources to allocate the requested memory, these functions
+/// return nullptr.
+_LIBSYCL_EXPORT
+void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
+ const device &syclDevice, const context &syclContext,
+ usm::alloc kind, const property_list &propList = {});
+
+/// Allocates USM of type `kind` with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param count the number of elements of type T to allocate.
+/// \param syclDevice the device to use for the allocation. The syclDevice
+/// parameter is ignored if kind is usm::alloc::host.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
+/// \param kind the type of memory to allocate.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak. If there are
+/// not enough resources to allocate the requested memory, these functions
+/// return nullptr.
+template <typename T>
+T *aligned_alloc(std::size_t alignment, std::size_t count,
+ const device &syclDevice, const context &syclContext,
+ usm::alloc kind, const property_list &propList = {}) {
+ return static_cast<T *>(aligned_alloc(
+ alignment, count * sizeof(T), syclDevice, syclContext, kind, propList));
+}
+
+/// Allocates USM of type `kind` with specified alignment.
+/// \param alignment the alignment of the allocated memory.
+/// \param numBytes the number of bytes to allocate.
+/// \param syclQueue a queue that provides the device and context.
+/// \param kind the type of memory to allocate.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak. If there are
+/// not enough resources to allocate the requested memory, these functions
+/// return nullptr.
+_LIBSYCL_EXPORT
+void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
+ const queue &syclQueue, usm::alloc kind,
+ const property_list &propList = {});
+
+/// Allocates USM of type `kind` with specified alignment.
+///
+/// \param alignment the alignment of the allocated memory.
+/// \param count the number of elements of type T to allocate.
+/// \param syclQueue a queue that provides the device and context.
+/// \param kind the type of memory to allocate.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak. If there are
+/// not enough resources to allocate the requested memory, these functions
+/// return nullptr.
+template <typename T>
+T *aligned_alloc(std::size_t alignment, std::size_t count,
+ const queue &syclQueue, usm::alloc kind,
+ const property_list &propList = {}) {
+ return aligned_alloc<T>(alignment, count, syclQueue.get_device(),
+ syclQueue.get_context(), kind, propList);
+}
+
/// @}
/// \name SYCL 2020 4.8.3.6. Memory deallocation functions.
diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp
index 4ff3d0d916061..becb6e3466864 100644
--- a/libsycl/src/usm_functions.cpp
+++ b/libsycl/src/usm_functions.cpp
@@ -156,6 +156,19 @@ void *malloc(std::size_t numBytes, const queue &syclQueue, usm::alloc kind,
propList);
}
+void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
+ const device &syclDevice, const context &syclContext,
+ usm::alloc kind, const property_list &propList = {}) {
+ // TODO: this is the important function to implement
+}
+
+void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
+ const queue &syclQueue, usm::alloc kind,
+ const property_list &propList = {}) {
+ return aligned_alloc(alignment, numBytes, syclQueue.get_device(),
+ syclQueue.get_context(), kind, propList);
+}
+
// SYCL 2020 4.8.3.6. Memory deallocation functions.
void free(void *ptr, const context &ctxt) {
>From 33af44e0fa7b212200024762bb83d22a6af39f18 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Mon, 27 Jul 2026 22:13:14 +0300
Subject: [PATCH 04/13] Add implementation of aligned_alloc free function
---
libsycl/include/sycl/__impl/usm_functions.hpp | 2 +
libsycl/src/usm_functions.cpp | 66 +++++++++++++++----
2 files changed, 55 insertions(+), 13 deletions(-)
diff --git a/libsycl/include/sycl/__impl/usm_functions.hpp b/libsycl/include/sycl/__impl/usm_functions.hpp
index 78847a5828f70..acf3bf56a7e6f 100644
--- a/libsycl/include/sycl/__impl/usm_functions.hpp
+++ b/libsycl/include/sycl/__impl/usm_functions.hpp
@@ -56,6 +56,8 @@ T *malloc_device(std::size_t count, const device &syclDevice,
const property_list &propList = {}) {
// TODO: to rewrite with aligned_malloc_device once it's supported in
// liboffload.
+ // Why does this need to be rewrited to use aligned version when there is
+ // explicit aligned_malloc_device ?
return static_cast<T *>(
malloc_device(count * sizeof(T), syclDevice, syclContext, propList));
}
diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp
index becb6e3466864..eede544cff215 100644
--- a/libsycl/src/usm_functions.cpp
+++ b/libsycl/src/usm_functions.cpp
@@ -33,13 +33,14 @@ void *malloc_device(std::size_t numBytes, const queue &syclQueue,
void *aligned_alloc_device(size_t alignment, size_t numBytes,
const device &syclDevice, const context &syclContext,
- const property_list &propList = {}) {
- // TODO: implementation here with malloc
+ const property_list &propList) {
+ return aligned_alloc(alignment, numBytes, syclDevice, syclContext,
+ usm::alloc::device, propList);
}
void *aligned_alloc_device(size_t alignment, size_t numBytes,
const queue &syclQueue,
- const property_list &propList = {}) {
+ const property_list &propList) {
return aligned_alloc_device(alignment, numBytes, syclQueue.get_device(),
syclQueue.get_context(), propList);
}
@@ -50,12 +51,14 @@ void *malloc_host(std::size_t numBytes, const context &syclContext,
const property_list &propList) {
auto ContextDevices = syclContext.get_devices();
assert(!ContextDevices.empty() && "Context can't be created without device");
- if (std::none_of(
- ContextDevices.begin(), ContextDevices.end(),
- [](device Dev) { return Dev.has(aspect::usm_host_allocations); }))
+ if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
+ [](const device &Dev) {
+ return Dev.has(aspect::usm_host_allocations);
+ })) {
throw sycl::exception(
sycl::errc::feature_not_supported,
- "All devices of context do not support host USM allocations.");
+ "None of the context's devices support host USM allocations.");
+ }
return malloc(numBytes, ContextDevices[0], syclContext, usm::alloc::host,
propList);
}
@@ -67,13 +70,25 @@ void *malloc_host(std::size_t numBytes, const queue &syclQueue,
void *aligned_alloc_host(size_t alignment, size_t numBytes,
const context &syclContext,
- const property_list &propList = {}) {
+ const property_list &propList) {
// TODO: implementation here with malloc
+ auto ContextDevices = syclContext.get_devices();
+ assert(!ContextDevices.empty() && "Context can't be created without device");
+ if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
+ [](const device &Dev) {
+ return Dev.has(aspect::usm_host_allocations);
+ })) {
+ throw sycl::exception(
+ sycl::errc::feature_not_supported,
+ "None of the context's devices support host USM allocations.");
+ }
+ return aligned_alloc(alignment, numBytes, ContextDevices[0], syclContext,
+ usm::alloc::host, propList);
}
void *aligned_alloc_host(size_t alignment, size_t numBytes,
const queue &syclQueue,
- const property_list &propList = {}) {
+ const property_list &propList) {
return aligned_alloc_host(alignment, numBytes, syclQueue.get_context(),
propList);
}
@@ -94,13 +109,15 @@ void *malloc_shared(std::size_t numBytes, const queue &syclQueue,
void *aligned_alloc_shared(size_t alignment, size_t numBytes,
const device &syclDevice, const context &syclContext,
- const property_list &propList = {}) {
+ const property_list &propList) {
// TODO: implementation here with malloc
+ return aligned_alloc(alignment, numBytes, syclDevice, syclContext,
+ usm::alloc::shared, propList);
}
void *aligned_alloc_shared(size_t alignment, size_t numBytes,
const queue &syclQueue,
- const property_list &propList = {}) {
+ const property_list &propList) {
return aligned_alloc_shared(alignment, numBytes, syclQueue.get_device(),
syclQueue.get_context(), propList);
}
@@ -158,13 +175,36 @@ void *malloc(std::size_t numBytes, const queue &syclQueue, usm::alloc kind,
void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
const device &syclDevice, const context &syclContext,
- usm::alloc kind, const property_list &propList = {}) {
+ usm::alloc kind, const property_list &propList) {
// TODO: this is the important function to implement
+ auto ContextDevices = syclContext.get_devices();
+ assert(!ContextDevices.empty() && "Context can't be created without device");
+ if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
+ [&syclDevice](device Dev) { return Dev == syclDevice; }))
+ throw exception(make_error_code(errc::invalid),
+ "Specified device is not contained by specified context.");
+ if (!syclDevice.has(getAspectByAllocationKind(kind)))
+ throw sycl::exception(
+ sycl::errc::feature_not_supported,
+ "Device doesn't support requested kind of USM allocation");
+
+ if (!numBytes)
+ return nullptr;
+
+ void *Ptr{};
+ auto OLDevice = detail::getSyclObjImpl(syclDevice)->getOLHandle();
+ auto Result = kind == usm::alloc::host
+ ? detail::callNoCheck(olMemAllocAlignedHost, OLDevice,
+ numBytes, alignment, &Ptr)
+ : detail::callNoCheck(olMemAllocAligned, OLDevice,
+ detail::getOlAllocType(kind),
+ numBytes, alignment, &Ptr);
+ return detail::isFailed(Result) ? nullptr : Ptr;
}
void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
const queue &syclQueue, usm::alloc kind,
- const property_list &propList = {}) {
+ const property_list &propList) {
return aligned_alloc(alignment, numBytes, syclQueue.get_device(),
syclQueue.get_context(), kind, propList);
}
>From 633d81063f85b55a4603ccda7f133b4017e61580 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Sat, 1 Aug 2026 19:25:33 +0300
Subject: [PATCH 05/13] Verify alignment is power of 2 & add tests
---
libsycl/include/sycl/__impl/detail/common.hpp | 30 ++++++++
libsycl/include/sycl/__impl/usm_functions.hpp | 28 +++----
libsycl/src/usm_functions.cpp | 6 +-
libsycl/test/usm/alloc_functions.cpp | 74 +++++++++++++++++++
4 files changed, 124 insertions(+), 14 deletions(-)
create mode 100644 libsycl/include/sycl/__impl/detail/common.hpp
diff --git a/libsycl/include/sycl/__impl/detail/common.hpp b/libsycl/include/sycl/__impl/detail/common.hpp
new file mode 100644
index 0000000000000..fe47f84bbad7b
--- /dev/null
+++ b/libsycl/include/sycl/__impl/detail/common.hpp
@@ -0,0 +1,30 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+///
+/// \file
+/// This file contains the declaration of functions commonly used in SYCL
+/// runtime implementation.
+///
+//===----------------------------------------------------------------------===//
+
+#ifndef _LIBSYCL___IMPL_DETAIL_COMMON_HPP
+#define _LIBSYCL___IMPL_DETAIL_COMMON_HPP
+
+#include <sycl/__impl/detail/config.hpp>
+
+_LIBSYCL_BEGIN_NAMESPACE_SYCL
+
+namespace detail {
+
+constexpr bool isPowerOf2(std::size_t n) { return (n & (n - 1)) == 0; }
+
+} // namespace detail
+
+_LIBSYCL_END_NAMESPACE_SYCL
+
+#endif // _LIBSYCL___IMPL_DETAIL_COMMON_HPP
\ No newline at end of file
diff --git a/libsycl/include/sycl/__impl/usm_functions.hpp b/libsycl/include/sycl/__impl/usm_functions.hpp
index acf3bf56a7e6f..a0f32e2a9f538 100644
--- a/libsycl/include/sycl/__impl/usm_functions.hpp
+++ b/libsycl/include/sycl/__impl/usm_functions.hpp
@@ -40,26 +40,28 @@ _LIBSYCL_EXPORT void *malloc_device(std::size_t numBytes,
const context &syclContext,
const property_list &propList = {});
+/// Forward declaration of aligned_alloc_device for use in malloc_device.
+template <typename T>
+T *aligned_alloc_device(std::size_t alignment, std::size_t count,
+ const device &syclDevice, const context &syclContext,
+ const property_list &propList = {});
+
/// Allocates device USM.
///
/// \param count the number of elements of type T to allocate.
/// \param syclDevice the device to use for the allocation.
-/// \param syclContext a context containing syclDevice or its parent device if
-/// syclDevice is a subdevice.
+/// \param syclContext a context containing syclDevice or its parent device
+/// if syclDevice is a subdevice.
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which is allocated on
-/// syclDevice and which must eventually be deallocated with sycl::free in order
-/// to avoid a memory leak.
+/// syclDevice and which must eventually be deallocated with sycl::free in
+/// order to avoid a memory leak.
template <typename T>
T *malloc_device(std::size_t count, const device &syclDevice,
const context &syclContext,
const property_list &propList = {}) {
- // TODO: to rewrite with aligned_malloc_device once it's supported in
- // liboffload.
- // Why does this need to be rewrited to use aligned version when there is
- // explicit aligned_malloc_device ?
- return static_cast<T *>(
- malloc_device(count * sizeof(T), syclDevice, syclContext, propList));
+ return aligned_alloc_device<T>(alignof(T), count, syclDevice, syclContext,
+ propList);
}
/// Allocates device USM.
@@ -119,7 +121,7 @@ void *aligned_alloc_device(std::size_t alignment, std::size_t numBytes,
template <typename T>
T *aligned_alloc_device(std::size_t alignment, std::size_t count,
const device &syclDevice, const context &syclContext,
- const property_list &propList = {}) {
+ const property_list &propList) {
return static_cast<T *>(aligned_alloc_device(
alignment, count * sizeof(T), syclDevice, syclContext, propList));
}
@@ -151,8 +153,8 @@ template <typename T>
T *aligned_alloc_device(std::size_t alignment, std::size_t count,
const queue &syclQueue,
const property_list &propList = {}) {
- return alligned_alloc_device<T>(alignment, count, syclQueue.get_device(),
- syclQueue.get_context(), propList);
+ return aligned_alloc_device<T>(alignment, count, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
}
/// @}
diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp
index eede544cff215..1f6aa55559f78 100644
--- a/libsycl/src/usm_functions.cpp
+++ b/libsycl/src/usm_functions.cpp
@@ -10,6 +10,7 @@
#include <detail/device_impl.hpp>
#include <detail/offload/offload_utils.hpp>
+#include <sycl/__impl/detail/common.hpp>
#include <OffloadAPI.h>
@@ -176,7 +177,10 @@ void *malloc(std::size_t numBytes, const queue &syclQueue, usm::alloc kind,
void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
const device &syclDevice, const context &syclContext,
usm::alloc kind, const property_list &propList) {
- // TODO: this is the important function to implement
+ if (alignment == 0 || !detail::isPowerOf2(alignment))
+ throw exception(make_error_code(errc::invalid),
+ "Alignment must be a non-zero power of two");
+
auto ContextDevices = syclContext.get_devices();
assert(!ContextDevices.empty() && "Context can't be created without device");
if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
diff --git a/libsycl/test/usm/alloc_functions.cpp b/libsycl/test/usm/alloc_functions.cpp
index 7df7206ef532e..63d663f05aff9 100644
--- a/libsycl/test/usm/alloc_functions.cpp
+++ b/libsycl/test/usm/alloc_functions.cpp
@@ -4,6 +4,7 @@
#include <sycl/sycl.hpp>
+#include <cassert>
#include <cstddef>
#include <iostream>
#include <tuple>
@@ -66,14 +67,37 @@ int main() {
[&]() { return MDevice(q, property_list{}); },
[&]() { return MDevice(d, ctx, property_list{}); }});
+ auto ADevice = [&](auto... args) {
+ return aligned_alloc_device(Align, 1024, args...);
+ };
+
+ CheckAll(Align, std::tuple{
+ [&]() { return ADevice(q); },
+ [&]() { return ADevice(d, ctx); },
+ [&]() { return ADevice(q, property_list{}); },
+ [&]() { return ADevice(d, ctx, property_list{}); },
+ });
+
auto MHost = [&](auto... args) {
return malloc_host(sizeof(std::max_align_t), args...);
};
+
CheckAll(FAlign,
std::tuple{[&]() { return MHost(q); }, [&]() { return MHost(ctx); },
[&]() { return MHost(q, property_list{}); },
[&]() { return MHost(ctx, property_list{}); }});
+ auto AHost = [&](auto... args) {
+ return aligned_alloc_host(Align, 1024, args...);
+ };
+
+ CheckAll(Align, std::tuple{
+ [&]() { return AHost(q); },
+ [&]() { return AHost(ctx); },
+ [&]() { return AHost(q, property_list{}); },
+ [&]() { return AHost(ctx, property_list{}); },
+ });
+
if (d.has(aspect::usm_shared_allocations)) {
auto MShared = [&](auto... args) {
return malloc_shared(sizeof(std::max_align_t), args...);
@@ -84,6 +108,16 @@ int main() {
[&]() { return MShared(d, ctx); },
[&]() { return MShared(q, property_list{}); },
[&]() { return MShared(d, ctx, property_list{}); }});
+
+ auto AShared = [&](auto... args) {
+ return aligned_alloc_shared(Align, 1024, args...);
+ };
+ CheckAll(Align, std::tuple{
+ [&]() { return AShared(q); },
+ [&]() { return AShared(d, ctx); },
+ [&]() { return AShared(q, property_list{}); },
+ [&]() { return AShared(d, ctx, property_list{}); },
+ });
}
auto TDevice = [&](auto... args) {
@@ -92,21 +126,40 @@ int main() {
CheckAll(Align, std::tuple{[&]() { return TDevice(q); },
[&]() { return TDevice(d, ctx); }});
+ auto TADevice = [&](auto... args) {
+ return aligned_alloc_device<Aligned>(Align, 1, args...);
+ };
+
+ CheckAll(Align, std::tuple{[&]() { return TADevice(q); },
+ [&]() { return TADevice(d, ctx); }});
+
auto THost = [&](auto... args) { return malloc_host<Aligned>(1, args...); };
CheckAll(Align, std::tuple{[&]() { return THost(q); },
[&]() { return THost(ctx); }});
+ auto TAHost = [&](auto... args) {
+ return aligned_alloc_host<Aligned>(Align, 1, args...);
+ };
+ CheckAll(Align, std::tuple{[&]() { return TAHost(q); },
+ [&]() { return TAHost(ctx); }});
+
if (d.has(aspect::usm_shared_allocations)) {
auto TShared = [&](auto... args) {
return malloc_shared<Aligned>(1, args...);
};
CheckAll(Align, std::tuple{[&]() { return TShared(q); },
[&]() { return TShared(d, ctx); }});
+ auto TAShared = [&](auto... args) {
+ return aligned_alloc_shared<Aligned>(Align, 1, args...);
+ };
+ CheckAll(Align, std::tuple{[&]() { return TAShared(q); },
+ [&]() { return TAShared(d, ctx); }});
}
auto Malloc = [&](auto... args) {
return malloc(sizeof(std::max_align_t), args...);
};
+
CheckAll(
FAlign,
std::tuple{
@@ -115,10 +168,31 @@ int main() {
[&]() { return Malloc(q, usm::alloc::host, property_list{}); },
[&]() { return Malloc(d, ctx, usm::alloc::host, property_list{}); }});
+ auto AMalloc = [&](auto... args) {
+ return aligned_alloc(Align, 1024, args...);
+ };
+
+ CheckAll(
+ Align,
+ std::tuple{
+ [&]() { return AMalloc(q, usm::alloc::host); },
+ [&]() { return AMalloc(d, ctx, usm::alloc::host); },
+ [&]() { return AMalloc(q, usm::alloc::host, property_list{}); },
+ [&]() { return AMalloc(d, ctx, usm::alloc::host, property_list{}); },
+ });
+
auto TMalloc = [&](auto... args) { return malloc<Aligned>(1, args...); };
CheckAll(Align,
std::tuple{[&]() { return TMalloc(q, usm::alloc::host); },
[&]() { return TMalloc(d, ctx, usm::alloc::host); }});
+ auto TAMalloc = [&](auto... args) {
+ return aligned_alloc<Aligned>(Align, 1, args...);
+ };
+
+ CheckAll(Align,
+ std::tuple{[&]() { return TAMalloc(q, usm::alloc::host); },
+ [&]() { return TAMalloc(d, ctx, usm::alloc::host); }});
+
return 0;
}
>From 6cf8583705c37c3dfe07ff77487025d4af6badfd Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Sat, 1 Aug 2026 19:31:05 +0300
Subject: [PATCH 06/13] removed solved issues from index.md
---
libsycl/docs/index.md | 3 ---
1 file changed, 3 deletions(-)
diff --git a/libsycl/docs/index.md b/libsycl/docs/index.md
index de14261322039..43ecadac29fdb 100644
--- a/libsycl/docs/index.md
+++ b/libsycl/docs/index.md
@@ -129,9 +129,6 @@ which doesn't currently support Windows.
- `property_list`: to fully implement and integrate with existing SYCL runtime classes supporting it
- usm allocations:
-
- - add aligned functions (blocked by liboffload support)
- - forward templated funcs to alignment methods (rewrite current impl)
- handle sub devices once they are implemented (blocked by liboffload support)
- `event`:
>From d579531f3613f6b0c23819f6952c2ba50e8adb37 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Wed, 12 Aug 2026 18:29:26 +0300
Subject: [PATCH 07/13] reorder functions to ensure aligned version is declared
---
libsycl/include/sycl/__impl/usm_functions.hpp | 358 +++++++++---------
libsycl/src/usm_functions.cpp | 138 +++----
2 files changed, 247 insertions(+), 249 deletions(-)
diff --git a/libsycl/include/sycl/__impl/usm_functions.hpp b/libsycl/include/sycl/__impl/usm_functions.hpp
index a0f32e2a9f538..eb4473729f950 100644
--- a/libsycl/include/sycl/__impl/usm_functions.hpp
+++ b/libsycl/include/sycl/__impl/usm_functions.hpp
@@ -25,71 +25,6 @@ _LIBSYCL_BEGIN_NAMESPACE_SYCL
/// \name SYCL 2020 4.8.3.2. Device allocation functions.
/// \brief Allocations in device memory are not accessible by the host.
/// @{
-/// Allocates device USM.
-///
-/// \param numBytes the number of bytes to allocate.
-/// \param syclDevice the device to use for the allocation.
-/// \param syclContext a context containing syclDevice or its parent device if
-/// syclDevice is a subdevice.
-/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which is allocated on
-/// syclDevice and which must eventually be deallocated with sycl::free in order
-/// to avoid a memory leak.
-_LIBSYCL_EXPORT void *malloc_device(std::size_t numBytes,
- const device &syclDevice,
- const context &syclContext,
- const property_list &propList = {});
-
-/// Forward declaration of aligned_alloc_device for use in malloc_device.
-template <typename T>
-T *aligned_alloc_device(std::size_t alignment, std::size_t count,
- const device &syclDevice, const context &syclContext,
- const property_list &propList = {});
-
-/// Allocates device USM.
-///
-/// \param count the number of elements of type T to allocate.
-/// \param syclDevice the device to use for the allocation.
-/// \param syclContext a context containing syclDevice or its parent device
-/// if syclDevice is a subdevice.
-/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which is allocated on
-/// syclDevice and which must eventually be deallocated with sycl::free in
-/// order to avoid a memory leak.
-template <typename T>
-T *malloc_device(std::size_t count, const device &syclDevice,
- const context &syclContext,
- const property_list &propList = {}) {
- return aligned_alloc_device<T>(alignof(T), count, syclDevice, syclContext,
- propList);
-}
-
-/// Allocates device USM.
-///
-/// \param numBytes the number of bytes to allocate.
-/// \param syclQueue a queue that provides the device and context.
-/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which is allocated on
-/// syclDevice and which must eventually be deallocated with sycl::free in order
-/// to avoid a memory leak.
-_LIBSYCL_EXPORT void *malloc_device(std::size_t numBytes,
- const queue &syclQueue,
- const property_list &propList = {});
-
-/// Allocates device USM.
-///
-/// \param count the number of elements of type T to allocate.
-/// \param syclQueue a queue that provides the device and context.
-/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which is allocated on
-/// syclDevice and which must eventually be deallocated with sycl::free in order
-/// to avoid a memory leak.
-template <typename T>
-T *malloc_device(std::size_t count, const queue &syclQueue,
- const property_list &propList = {}) {
- return malloc_device<T>(count, syclQueue.get_device(),
- syclQueue.get_context(), propList);
-}
/// Allocates device USM with specified alignment.
///
@@ -157,63 +92,72 @@ T *aligned_alloc_device(std::size_t alignment, std::size_t count,
syclQueue.get_context(), propList);
}
-/// @}
-
-/// \name SYCL 2020 4.8.3.3. Host allocation functions.
-/// \brief Allocations in host memory are accessible by a device.
-/// @{
-/// Allocates host USM.
+/// Allocates device USM.
///
/// \param numBytes the number of bytes to allocate.
-/// \param syclContext the context that should have access to the allocated
-/// memory.
+/// \param syclDevice the device to use for the allocation.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which must eventually
-/// be deallocated with sycl::free in order to avoid a memory leak.
-_LIBSYCL_EXPORT
-void *malloc_host(std::size_t numBytes, const context &syclContext,
- const property_list &propList = {});
+/// \return a pointer to the newly allocated memory, which is allocated on
+/// syclDevice and which must eventually be deallocated with sycl::free in order
+/// to avoid a memory leak.
+_LIBSYCL_EXPORT void *malloc_device(std::size_t numBytes,
+ const device &syclDevice,
+ const context &syclContext,
+ const property_list &propList = {});
-/// Allocates host USM.
+/// Allocates device USM.
///
/// \param count the number of elements of type T to allocate.
-/// \param syclContext the context that should have access to the allocated
-/// memory.
+/// \param syclDevice the device to use for the allocation.
+/// \param syclContext a context containing syclDevice or its parent device
+/// if syclDevice is a subdevice.
/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak.
+/// \return a pointer to the newly allocated memory, which is allocated on
+/// syclDevice and which must eventually be deallocated with sycl::free in
+/// order to avoid a memory leak.
template <typename T>
-T *malloc_host(std::size_t count, const context &syclContext,
- const property_list &propList = {}) {
- // TODO: to rewrite with aligned_malloc_host once it's supported in
- // liboffload.
- return static_cast<T *>(
- malloc_host(count * sizeof(T), syclContext, propList));
+T *malloc_device(std::size_t count, const device &syclDevice,
+ const context &syclContext,
+ const property_list &propList = {}) {
+ return aligned_alloc_device<T>(alignof(T), count, syclDevice, syclContext,
+ propList);
}
-/// Allocates host USM.
+/// Allocates device USM.
///
/// \param numBytes the number of bytes to allocate.
-/// \param syclQueue queue that provides the context.
+/// \param syclQueue a queue that provides the device and context.
/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak.
-_LIBSYCL_EXPORT void *malloc_host(std::size_t numBytes, const queue &syclQueue,
- const property_list &propList = {});
+/// \return a pointer to the newly allocated memory, which is allocated on
+/// syclDevice and which must eventually be deallocated with sycl::free in order
+/// to avoid a memory leak.
+_LIBSYCL_EXPORT void *malloc_device(std::size_t numBytes,
+ const queue &syclQueue,
+ const property_list &propList = {});
-/// Allocates host USM.
+/// Allocates device USM.
///
/// \param count the number of elements of type T to allocate.
-/// \param syclQueue queue that provides the context.
+/// \param syclQueue a queue that provides the device and context.
/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak.
+/// \return a pointer to the newly allocated memory, which is allocated on
+/// syclDevice and which must eventually be deallocated with sycl::free in order
+/// to avoid a memory leak.
template <typename T>
-T *malloc_host(std::size_t count, const queue &syclQueue,
- const property_list &propList = {}) {
- return malloc_host<T>(count, syclQueue.get_context(), propList);
+T *malloc_device(std::size_t count, const queue &syclQueue,
+ const property_list &propList = {}) {
+ return malloc_device<T>(count, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
}
+/// @}
+
+/// \name SYCL 2020 4.8.3.3. Host allocation functions.
+/// \brief Allocations in host memory are accessible by a device.
+/// @{
+
/// Allocates host USM with specified alignment.
///
/// \param alignment the alignment of the allocated memory.
@@ -274,69 +218,64 @@ T *aligned_alloc_host(std::size_t alignment, std::size_t count,
propList);
}
-/// @}
-
-/// \name SYCL 2020 4.8.3.4. Shared allocation functions.
-/// \brief Allocations in shared memory are accessible by both host and device.
-/// @{
-/// Allocates shared USM.
+/// Allocates host USM.
///
/// \param numBytes the number of bytes to allocate.
-/// \param syclDevice the device to use for the allocation.
-/// \param syclContext a context containing syclDevice or its parent device if
-/// syclDevice is a subdevice.
+/// \param syclContext the context that should have access to the allocated
+/// memory.
/// \param propList the list of properties for the allocation.
-/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak.
-_LIBSYCL_EXPORT void *malloc_shared(std::size_t numBytes,
- const device &syclDevice,
- const context &syclContext,
- const property_list &propList = {});
+/// \return a pointer to the newly allocated memory, which must eventually
+/// be deallocated with sycl::free in order to avoid a memory leak.
+_LIBSYCL_EXPORT
+void *malloc_host(std::size_t numBytes, const context &syclContext,
+ const property_list &propList = {});
-/// Allocates shared USM.
+/// Allocates host USM.
///
/// \param count the number of elements of type T to allocate.
-/// \param syclDevice the device to use for the allocation.
-/// \param syclContext a context containing syclDevice or its parent device if
-/// syclDevice is a subdevice.
+/// \param syclContext the context that should have access to the allocated
+/// memory.
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
-T *malloc_shared(std::size_t count, const device &syclDevice,
- const context &syclContext,
- const property_list &propList = {}) {
- // TODO: to rewrite with aligned_malloc_shared once it's supported in
+T *malloc_host(std::size_t count, const context &syclContext,
+ const property_list &propList = {}) {
+ // TODO: to rewrite with aligned_malloc_host once it's supported in
// liboffload.
return static_cast<T *>(
- malloc_shared(count * sizeof(T), syclDevice, syclContext, propList));
+ malloc_host(count * sizeof(T), syclContext, propList));
}
-/// Allocates shared USM.
+/// Allocates host USM.
///
/// \param numBytes the number of bytes to allocate.
-/// \param syclQueue a queue that provides the device and context.
+/// \param syclQueue queue that provides the context.
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
-_LIBSYCL_EXPORT void *malloc_shared(std::size_t numBytes,
- const queue &syclQueue,
- const property_list &propList = {});
+_LIBSYCL_EXPORT void *malloc_host(std::size_t numBytes, const queue &syclQueue,
+ const property_list &propList = {});
-/// Allocates shared USM.
+/// Allocates host USM.
///
/// \param count the number of elements of type T to allocate.
-/// \param syclQueue a queue that provides the device and context.
+/// \param syclQueue queue that provides the context.
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
-T *malloc_shared(std::size_t count, const queue &syclQueue,
- const property_list &propList = {}) {
- return malloc_shared<T>(count, syclQueue.get_device(),
- syclQueue.get_context(), propList);
+T *malloc_host(std::size_t count, const queue &syclQueue,
+ const property_list &propList = {}) {
+ return malloc_host<T>(count, syclQueue.get_context(), propList);
}
+/// @}
+
+/// \name SYCL 2020 4.8.3.4. Shared allocation functions.
+/// \brief Allocations in shared memory are accessible by both host and device.
+/// @{
+
/// Allocates shared USM with specified alignment.
///
/// \param alignment the alignment of the allocated memory.
@@ -399,80 +338,69 @@ T *aligned_alloc_shared(std::size_t alignment, std::size_t count,
syclQueue.get_context(), propList);
}
-/// @}
-
-/// \name SYCL 2020 4.8.3.5. Parameterized allocation functions.
-/// @{
-/// Allocates USM of type `kind`.
+/// Allocates shared USM.
///
/// \param numBytes the number of bytes to allocate.
-/// \param syclDevice the device to use for the allocation. The syclDevice
-/// parameter is ignored if kind is usm::alloc::host.
+/// \param syclDevice the device to use for the allocation.
/// \param syclContext a context containing syclDevice or its parent device if
/// syclDevice is a subdevice.
-/// \param kind the type of memory to allocate.
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak. If there are
-/// not enough resources to allocate the requested memory, these functions
-/// return nullptr.
-_LIBSYCL_EXPORT void *malloc(std::size_t numBytes, const device &syclDevice,
- const context &syclContext, usm::alloc kind,
- const property_list &propList = {});
+/// deallocated with sycl::free in order to avoid a memory leak.
+_LIBSYCL_EXPORT void *malloc_shared(std::size_t numBytes,
+ const device &syclDevice,
+ const context &syclContext,
+ const property_list &propList = {});
-/// Allocates USM of type `kind`.
+/// Allocates shared USM.
///
/// \param count the number of elements of type T to allocate.
-/// \param syclDevice the device to use for the allocation. The syclDevice
-/// parameter is ignored if kind is usm::alloc::host.
+/// \param syclDevice the device to use for the allocation.
/// \param syclContext a context containing syclDevice or its parent device if
/// syclDevice is a subdevice.
-/// \param kind the type of memory to allocate.
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak. If there are
-/// not enough resources to allocate the requested memory, these functions
-/// return nullptr.
+/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
-T *malloc(std::size_t count, const device &syclDevice,
- const context &syclContext, usm::alloc kind,
- const property_list &propList = {}) {
- // TODO: to rewrite with aligned_malloc once it's supported in liboffload.
+T *malloc_shared(std::size_t count, const device &syclDevice,
+ const context &syclContext,
+ const property_list &propList = {}) {
+ // TODO: to rewrite with aligned_malloc_shared once it's supported in
+ // liboffload.
return static_cast<T *>(
- malloc(count * sizeof(T), syclDevice, syclContext, kind, propList));
+ malloc_shared(count * sizeof(T), syclDevice, syclContext, propList));
}
-/// Allocates USM of type `kind`.
+/// Allocates shared USM.
///
/// \param numBytes the number of bytes to allocate.
/// \param syclQueue a queue that provides the device and context.
-/// \param kind the type of memory to allocate.
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak. If there are
-/// not enough resources to allocate the requested memory, these functions
-/// return nullptr.
-_LIBSYCL_EXPORT void *malloc(std::size_t numBytes, const queue &syclQueue,
- usm::alloc kind,
- const property_list &propList = {});
+/// deallocated with sycl::free in order to avoid a memory leak.
+_LIBSYCL_EXPORT void *malloc_shared(std::size_t numBytes,
+ const queue &syclQueue,
+ const property_list &propList = {});
-/// Allocates USM of type `kind`.
+/// Allocates shared USM.
///
/// \param count the number of elements of type T to allocate.
/// \param syclQueue a queue that provides the device and context.
-/// \param kind the type of memory to allocate.
/// \param propList the list of properties for the allocation.
/// \return a pointer to the newly allocated memory, which must eventually be
-/// deallocated with sycl::free in order to avoid a memory leak. If there are
-/// not enough resources to allocate the requested memory, these functions
-/// return nullptr.
+/// deallocated with sycl::free in order to avoid a memory leak.
template <typename T>
-T *malloc(std::size_t count, const queue &syclQueue, usm::alloc kind,
- const property_list &propList = {}) {
- return malloc<T>(count, syclQueue.get_device(), syclQueue.get_context(), kind,
- propList);
+T *malloc_shared(std::size_t count, const queue &syclQueue,
+ const property_list &propList = {}) {
+ return malloc_shared<T>(count, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
}
+/// @}
+
+/// \name SYCL 2020 4.8.3.5. Parameterized allocation functions.
+/// @{
+
/// Allocates USM of type `kind` with specified alignment.
///
/// \param alignment the alignment of the allocated memory.
@@ -548,6 +476,76 @@ T *aligned_alloc(std::size_t alignment, std::size_t count,
syclQueue.get_context(), kind, propList);
}
+/// Allocates USM of type `kind`.
+///
+/// \param numBytes the number of bytes to allocate.
+/// \param syclDevice the device to use for the allocation. The syclDevice
+/// parameter is ignored if kind is usm::alloc::host.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
+/// \param kind the type of memory to allocate.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak. If there are
+/// not enough resources to allocate the requested memory, these functions
+/// return nullptr.
+_LIBSYCL_EXPORT void *malloc(std::size_t numBytes, const device &syclDevice,
+ const context &syclContext, usm::alloc kind,
+ const property_list &propList = {});
+
+/// Allocates USM of type `kind`.
+///
+/// \param count the number of elements of type T to allocate.
+/// \param syclDevice the device to use for the allocation. The syclDevice
+/// parameter is ignored if kind is usm::alloc::host.
+/// \param syclContext a context containing syclDevice or its parent device if
+/// syclDevice is a subdevice.
+/// \param kind the type of memory to allocate.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak. If there are
+/// not enough resources to allocate the requested memory, these functions
+/// return nullptr.
+template <typename T>
+T *malloc(std::size_t count, const device &syclDevice,
+ const context &syclContext, usm::alloc kind,
+ const property_list &propList = {}) {
+ // TODO: to rewrite with aligned_malloc once it's supported in liboffload.
+ return static_cast<T *>(
+ malloc(count * sizeof(T), syclDevice, syclContext, kind, propList));
+}
+
+/// Allocates USM of type `kind`.
+///
+/// \param numBytes the number of bytes to allocate.
+/// \param syclQueue a queue that provides the device and context.
+/// \param kind the type of memory to allocate.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak. If there are
+/// not enough resources to allocate the requested memory, these functions
+/// return nullptr.
+_LIBSYCL_EXPORT void *malloc(std::size_t numBytes, const queue &syclQueue,
+ usm::alloc kind,
+ const property_list &propList = {});
+
+/// Allocates USM of type `kind`.
+///
+/// \param count the number of elements of type T to allocate.
+/// \param syclQueue a queue that provides the device and context.
+/// \param kind the type of memory to allocate.
+/// \param propList the list of properties for the allocation.
+/// \return a pointer to the newly allocated memory, which must eventually be
+/// deallocated with sycl::free in order to avoid a memory leak. If there are
+/// not enough resources to allocate the requested memory, these functions
+/// return nullptr.
+template <typename T>
+T *malloc(std::size_t count, const queue &syclQueue, usm::alloc kind,
+ const property_list &propList = {}) {
+ return malloc<T>(count, syclQueue.get_device(), syclQueue.get_context(), kind,
+ propList);
+}
+
/// @}
/// \name SYCL 2020 4.8.3.6. Memory deallocation functions.
diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp
index 1f6aa55559f78..90c6699533861 100644
--- a/libsycl/src/usm_functions.cpp
+++ b/libsycl/src/usm_functions.cpp
@@ -20,18 +20,6 @@ _LIBSYCL_BEGIN_NAMESPACE_SYCL
// SYCL 2020 4.8.3.2. Device allocation functions.
-void *malloc_device(std::size_t numBytes, const device &syclDevice,
- const context &syclContext, const property_list &propList) {
- return malloc(numBytes, syclDevice, syclContext, usm::alloc::device,
- propList);
-}
-
-void *malloc_device(std::size_t numBytes, const queue &syclQueue,
- const property_list &propList) {
- return malloc_device(numBytes, syclQueue.get_device(),
- syclQueue.get_context(), propList);
-}
-
void *aligned_alloc_device(size_t alignment, size_t numBytes,
const device &syclDevice, const context &syclContext,
const property_list &propList) {
@@ -46,29 +34,20 @@ void *aligned_alloc_device(size_t alignment, size_t numBytes,
syclQueue.get_context(), propList);
}
-// SYCL 2020 4.8.3.3. Host allocation functions.
-
-void *malloc_host(std::size_t numBytes, const context &syclContext,
- const property_list &propList) {
- auto ContextDevices = syclContext.get_devices();
- assert(!ContextDevices.empty() && "Context can't be created without device");
- if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
- [](const device &Dev) {
- return Dev.has(aspect::usm_host_allocations);
- })) {
- throw sycl::exception(
- sycl::errc::feature_not_supported,
- "None of the context's devices support host USM allocations.");
- }
- return malloc(numBytes, ContextDevices[0], syclContext, usm::alloc::host,
+void *malloc_device(std::size_t numBytes, const device &syclDevice,
+ const context &syclContext, const property_list &propList) {
+ return malloc(numBytes, syclDevice, syclContext, usm::alloc::device,
propList);
}
-void *malloc_host(std::size_t numBytes, const queue &syclQueue,
- const property_list &propList) {
- return malloc_host(numBytes, syclQueue.get_context(), propList);
+void *malloc_device(std::size_t numBytes, const queue &syclQueue,
+ const property_list &propList) {
+ return malloc_device(numBytes, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
}
+// SYCL 2020 4.8.3.3. Host allocation functions.
+
void *aligned_alloc_host(size_t alignment, size_t numBytes,
const context &syclContext,
const property_list &propList) {
@@ -94,20 +73,29 @@ void *aligned_alloc_host(size_t alignment, size_t numBytes,
propList);
}
-// SYCL 2020 4.8.3.4. Shared allocation functions.
-
-void *malloc_shared(std::size_t numBytes, const device &syclDevice,
- const context &syclContext, const property_list &propList) {
- return malloc(numBytes, syclDevice, syclContext, usm::alloc::shared,
+void *malloc_host(std::size_t numBytes, const context &syclContext,
+ const property_list &propList) {
+ auto ContextDevices = syclContext.get_devices();
+ assert(!ContextDevices.empty() && "Context can't be created without device");
+ if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
+ [](const device &Dev) {
+ return Dev.has(aspect::usm_host_allocations);
+ })) {
+ throw sycl::exception(
+ sycl::errc::feature_not_supported,
+ "None of the context's devices support host USM allocations.");
+ }
+ return malloc(numBytes, ContextDevices[0], syclContext, usm::alloc::host,
propList);
}
-void *malloc_shared(std::size_t numBytes, const queue &syclQueue,
- const property_list &propList) {
- return malloc_shared(numBytes, syclQueue.get_device(),
- syclQueue.get_context(), propList);
+void *malloc_host(std::size_t numBytes, const queue &syclQueue,
+ const property_list &propList) {
+ return malloc_host(numBytes, syclQueue.get_context(), propList);
}
+// SYCL 2020 4.8.3.4. Shared allocation functions.
+
void *aligned_alloc_shared(size_t alignment, size_t numBytes,
const device &syclDevice, const context &syclContext,
const property_list &propList) {
@@ -123,6 +111,18 @@ void *aligned_alloc_shared(size_t alignment, size_t numBytes,
syclQueue.get_context(), propList);
}
+void *malloc_shared(std::size_t numBytes, const device &syclDevice,
+ const context &syclContext, const property_list &propList) {
+ return malloc(numBytes, syclDevice, syclContext, usm::alloc::shared,
+ propList);
+}
+
+void *malloc_shared(std::size_t numBytes, const queue &syclQueue,
+ const property_list &propList) {
+ return malloc_shared(numBytes, syclQueue.get_device(),
+ syclQueue.get_context(), propList);
+}
+
// SYCL 2020 4.8.3.5. Parameterized allocation functions.
static aspect getAspectByAllocationKind(usm::alloc kind) {
@@ -141,9 +141,13 @@ static aspect getAspectByAllocationKind(usm::alloc kind) {
}
}
-void *malloc(std::size_t numBytes, const device &syclDevice,
- const context &syclContext, usm::alloc kind,
- const property_list &propList) {
+void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
+ const device &syclDevice, const context &syclContext,
+ usm::alloc kind, const property_list &propList) {
+ if (alignment == 0 || !detail::isPowerOf2(alignment))
+ throw exception(make_error_code(errc::invalid),
+ "Alignment must be a non-zero power of two");
+
auto ContextDevices = syclContext.get_devices();
assert(!ContextDevices.empty() && "Context can't be created without device");
if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
@@ -160,27 +164,25 @@ void *malloc(std::size_t numBytes, const device &syclDevice,
void *Ptr{};
auto OLDevice = detail::getSyclObjImpl(syclDevice)->getOLHandle();
- auto Result =
- kind == usm::alloc::host
- ? detail::callNoCheck(olMemAllocHost, OLDevice, numBytes, &Ptr)
- : detail::callNoCheck(olMemAlloc, OLDevice,
- detail::getOlAllocType(kind), numBytes, &Ptr);
+ auto Result = kind == usm::alloc::host
+ ? detail::callNoCheck(olMemAllocAlignedHost, OLDevice,
+ numBytes, alignment, &Ptr)
+ : detail::callNoCheck(olMemAllocAligned, OLDevice,
+ detail::getOlAllocType(kind),
+ numBytes, alignment, &Ptr);
return detail::isFailed(Result) ? nullptr : Ptr;
}
-void *malloc(std::size_t numBytes, const queue &syclQueue, usm::alloc kind,
- const property_list &propList) {
- return malloc(numBytes, syclQueue.get_device(), syclQueue.get_context(), kind,
- propList);
-}
-
void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
- const device &syclDevice, const context &syclContext,
- usm::alloc kind, const property_list &propList) {
- if (alignment == 0 || !detail::isPowerOf2(alignment))
- throw exception(make_error_code(errc::invalid),
- "Alignment must be a non-zero power of two");
+ const queue &syclQueue, usm::alloc kind,
+ const property_list &propList) {
+ return aligned_alloc(alignment, numBytes, syclQueue.get_device(),
+ syclQueue.get_context(), kind, propList);
+}
+void *malloc(std::size_t numBytes, const device &syclDevice,
+ const context &syclContext, usm::alloc kind,
+ const property_list &propList) {
auto ContextDevices = syclContext.get_devices();
assert(!ContextDevices.empty() && "Context can't be created without device");
if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
@@ -197,20 +199,18 @@ void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
void *Ptr{};
auto OLDevice = detail::getSyclObjImpl(syclDevice)->getOLHandle();
- auto Result = kind == usm::alloc::host
- ? detail::callNoCheck(olMemAllocAlignedHost, OLDevice,
- numBytes, alignment, &Ptr)
- : detail::callNoCheck(olMemAllocAligned, OLDevice,
- detail::getOlAllocType(kind),
- numBytes, alignment, &Ptr);
+ auto Result =
+ kind == usm::alloc::host
+ ? detail::callNoCheck(olMemAllocHost, OLDevice, numBytes, &Ptr)
+ : detail::callNoCheck(olMemAlloc, OLDevice,
+ detail::getOlAllocType(kind), numBytes, &Ptr);
return detail::isFailed(Result) ? nullptr : Ptr;
}
-void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
- const queue &syclQueue, usm::alloc kind,
- const property_list &propList) {
- return aligned_alloc(alignment, numBytes, syclQueue.get_device(),
- syclQueue.get_context(), kind, propList);
+void *malloc(std::size_t numBytes, const queue &syclQueue, usm::alloc kind,
+ const property_list &propList) {
+ return malloc(numBytes, syclQueue.get_device(), syclQueue.get_context(), kind,
+ propList);
}
// SYCL 2020 4.8.3.6. Memory deallocation functions.
>From 21aaceca183c7241ab5f32770ed103189f6b7231 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Wed, 12 Aug 2026 19:22:55 +0300
Subject: [PATCH 08/13] Satisfy alignment minimums on templated types, cleanup
---
libsycl/include/sycl/__impl/usm_functions.hpp | 40 ++++++------
libsycl/src/usm_functions.cpp | 65 +++++--------------
2 files changed, 37 insertions(+), 68 deletions(-)
diff --git a/libsycl/include/sycl/__impl/usm_functions.hpp b/libsycl/include/sycl/__impl/usm_functions.hpp
index eb4473729f950..2514dc51482db 100644
--- a/libsycl/include/sycl/__impl/usm_functions.hpp
+++ b/libsycl/include/sycl/__impl/usm_functions.hpp
@@ -20,6 +20,8 @@
#include <sycl/__impl/detail/config.hpp>
+#include <algorithm>
+
_LIBSYCL_BEGIN_NAMESPACE_SYCL
/// \name SYCL 2020 4.8.3.2. Device allocation functions.
@@ -56,9 +58,10 @@ void *aligned_alloc_device(std::size_t alignment, std::size_t numBytes,
template <typename T>
T *aligned_alloc_device(std::size_t alignment, std::size_t count,
const device &syclDevice, const context &syclContext,
- const property_list &propList) {
- return static_cast<T *>(aligned_alloc_device(
- alignment, count * sizeof(T), syclDevice, syclContext, propList));
+ const property_list &propList = {}) {
+ return static_cast<T *>(aligned_alloc_device(std::max(alignment, alignof(T)),
+ count * sizeof(T), syclDevice,
+ syclContext, propList));
}
/// Allocates device USM with specified alignment.
@@ -185,8 +188,9 @@ template <typename T>
T *aligned_alloc_host(std::size_t alignment, std::size_t count,
const context &syclContext,
const property_list &propList = {}) {
- return static_cast<T *>(
- aligned_alloc_host(alignment, count * sizeof(T), syclContext, propList));
+ return static_cast<T *>(aligned_alloc_host(std::max(alignment, alignof(T)),
+ count * sizeof(T), syclContext,
+ propList));
}
/// Allocates host USM with specified alignment.
@@ -241,10 +245,7 @@ void *malloc_host(std::size_t numBytes, const context &syclContext,
template <typename T>
T *malloc_host(std::size_t count, const context &syclContext,
const property_list &propList = {}) {
- // TODO: to rewrite with aligned_malloc_host once it's supported in
- // liboffload.
- return static_cast<T *>(
- malloc_host(count * sizeof(T), syclContext, propList));
+ return aligned_alloc_host<T>(alignof(T), count, syclContext, propList);
}
/// Allocates host USM.
@@ -305,8 +306,9 @@ template <typename T>
T *aligned_alloc_shared(std::size_t alignment, std::size_t count,
const device &syclDevice, const context &syclContext,
const property_list &propList = {}) {
- return static_cast<T *>(aligned_alloc_shared(
- alignment, count * sizeof(T), syclDevice, syclContext, propList));
+ return static_cast<T *>(aligned_alloc_shared(std::max(alignment, alignof(T)),
+ count * sizeof(T), syclDevice,
+ syclContext, propList));
}
/// Allocates shared USM with specified alignment.
@@ -365,10 +367,8 @@ template <typename T>
T *malloc_shared(std::size_t count, const device &syclDevice,
const context &syclContext,
const property_list &propList = {}) {
- // TODO: to rewrite with aligned_malloc_shared once it's supported in
- // liboffload.
- return static_cast<T *>(
- malloc_shared(count * sizeof(T), syclDevice, syclContext, propList));
+ return aligned_alloc_shared<T>(alignof(T), count, syclDevice, syclContext,
+ propList);
}
/// Allocates shared USM.
@@ -438,8 +438,9 @@ template <typename T>
T *aligned_alloc(std::size_t alignment, std::size_t count,
const device &syclDevice, const context &syclContext,
usm::alloc kind, const property_list &propList = {}) {
- return static_cast<T *>(aligned_alloc(
- alignment, count * sizeof(T), syclDevice, syclContext, kind, propList));
+ return static_cast<T *>(aligned_alloc(std::max(alignment, alignof(T)),
+ count * sizeof(T), syclDevice,
+ syclContext, kind, propList));
}
/// Allocates USM of type `kind` with specified alignment.
@@ -510,9 +511,8 @@ template <typename T>
T *malloc(std::size_t count, const device &syclDevice,
const context &syclContext, usm::alloc kind,
const property_list &propList = {}) {
- // TODO: to rewrite with aligned_malloc once it's supported in liboffload.
- return static_cast<T *>(
- malloc(count * sizeof(T), syclDevice, syclContext, kind, propList));
+ return aligned_alloc<T>(alignof(T), count, syclDevice, syclContext, kind,
+ propList);
}
/// Allocates USM of type `kind`.
diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp
index 90c6699533861..0f675ad0de9fb 100644
--- a/libsycl/src/usm_functions.cpp
+++ b/libsycl/src/usm_functions.cpp
@@ -48,21 +48,26 @@ void *malloc_device(std::size_t numBytes, const queue &syclQueue,
// SYCL 2020 4.8.3.3. Host allocation functions.
-void *aligned_alloc_host(size_t alignment, size_t numBytes,
- const context &syclContext,
- const property_list &propList) {
- // TODO: implementation here with malloc
+static device getHostAllocDevice(const context &syclContext) {
auto ContextDevices = syclContext.get_devices();
- assert(!ContextDevices.empty() && "Context can't be created without device");
- if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
- [](const device &Dev) {
- return Dev.has(aspect::usm_host_allocations);
- })) {
+
+ auto It = std::find_if(
+ ContextDevices.begin(), ContextDevices.end(),
+ [](const device &Dev) { return Dev.has(aspect::usm_host_allocations); });
+
+ if (It == ContextDevices.end()) {
throw sycl::exception(
sycl::errc::feature_not_supported,
"None of the context's devices support host USM allocations.");
}
- return aligned_alloc(alignment, numBytes, ContextDevices[0], syclContext,
+ return *It;
+}
+
+void *aligned_alloc_host(size_t alignment, size_t numBytes,
+ const context &syclContext,
+ const property_list &propList) {
+ auto device = getHostAllocDevice(syclContext);
+ return aligned_alloc(alignment, numBytes, device, syclContext,
usm::alloc::host, propList);
}
@@ -75,18 +80,7 @@ void *aligned_alloc_host(size_t alignment, size_t numBytes,
void *malloc_host(std::size_t numBytes, const context &syclContext,
const property_list &propList) {
- auto ContextDevices = syclContext.get_devices();
- assert(!ContextDevices.empty() && "Context can't be created without device");
- if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
- [](const device &Dev) {
- return Dev.has(aspect::usm_host_allocations);
- })) {
- throw sycl::exception(
- sycl::errc::feature_not_supported,
- "None of the context's devices support host USM allocations.");
- }
- return malloc(numBytes, ContextDevices[0], syclContext, usm::alloc::host,
- propList);
+ return aligned_alloc_host(0, numBytes, syclContext, propList);
}
void *malloc_host(std::size_t numBytes, const queue &syclQueue,
@@ -99,7 +93,6 @@ void *malloc_host(std::size_t numBytes, const queue &syclQueue,
void *aligned_alloc_shared(size_t alignment, size_t numBytes,
const device &syclDevice, const context &syclContext,
const property_list &propList) {
- // TODO: implementation here with malloc
return aligned_alloc(alignment, numBytes, syclDevice, syclContext,
usm::alloc::shared, propList);
}
@@ -144,9 +137,6 @@ static aspect getAspectByAllocationKind(usm::alloc kind) {
void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
const device &syclDevice, const context &syclContext,
usm::alloc kind, const property_list &propList) {
- if (alignment == 0 || !detail::isPowerOf2(alignment))
- throw exception(make_error_code(errc::invalid),
- "Alignment must be a non-zero power of two");
auto ContextDevices = syclContext.get_devices();
assert(!ContextDevices.empty() && "Context can't be created without device");
@@ -183,28 +173,7 @@ void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
void *malloc(std::size_t numBytes, const device &syclDevice,
const context &syclContext, usm::alloc kind,
const property_list &propList) {
- auto ContextDevices = syclContext.get_devices();
- assert(!ContextDevices.empty() && "Context can't be created without device");
- if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
- [&syclDevice](device Dev) { return Dev == syclDevice; }))
- throw exception(make_error_code(errc::invalid),
- "Specified device is not contained by specified context.");
- if (!syclDevice.has(getAspectByAllocationKind(kind)))
- throw sycl::exception(
- sycl::errc::feature_not_supported,
- "Device doesn't support requested kind of USM allocation");
-
- if (!numBytes)
- return nullptr;
-
- void *Ptr{};
- auto OLDevice = detail::getSyclObjImpl(syclDevice)->getOLHandle();
- auto Result =
- kind == usm::alloc::host
- ? detail::callNoCheck(olMemAllocHost, OLDevice, numBytes, &Ptr)
- : detail::callNoCheck(olMemAlloc, OLDevice,
- detail::getOlAllocType(kind), numBytes, &Ptr);
- return detail::isFailed(Result) ? nullptr : Ptr;
+ return aligned_alloc(0, numBytes, syclDevice, syclContext, kind, propList);
}
void *malloc(std::size_t numBytes, const queue &syclQueue, usm::alloc kind,
>From a4d050f12cb502cf9e859f08f9caa0071f3ca5f9 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Wed, 12 Aug 2026 20:28:35 +0300
Subject: [PATCH 09/13] add unittests
---
libsycl/src/usm_functions.cpp | 12 ++-
libsycl/unittests/CMakeLists.txt | 1 +
libsycl/unittests/mock/helpers.cpp | 31 ++++++
libsycl/unittests/mock/helpers.hpp | 6 ++
libsycl/unittests/mock/mock.cpp | 17 ++++
libsycl/unittests/usm/CMakeLists.txt | 3 +
libsycl/unittests/usm/alloc.cpp | 147 +++++++++++++++++++++++++++
7 files changed, 214 insertions(+), 3 deletions(-)
create mode 100644 libsycl/unittests/usm/CMakeLists.txt
create mode 100644 libsycl/unittests/usm/alloc.cpp
diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp
index 0f675ad0de9fb..7fb69b4269252 100644
--- a/libsycl/src/usm_functions.cpp
+++ b/libsycl/src/usm_functions.cpp
@@ -80,7 +80,8 @@ void *aligned_alloc_host(size_t alignment, size_t numBytes,
void *malloc_host(std::size_t numBytes, const context &syclContext,
const property_list &propList) {
- return aligned_alloc_host(0, numBytes, syclContext, propList);
+ return aligned_alloc_host(alignof(std::max_align_t), numBytes, syclContext,
+ propList);
}
void *malloc_host(std::size_t numBytes, const queue &syclQueue,
@@ -138,12 +139,16 @@ void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
const device &syclDevice, const context &syclContext,
usm::alloc kind, const property_list &propList) {
+ if (alignment == 0 || !detail::isPowerOf2(alignment))
+ throw exception(sycl::make_error_code(sycl::errc::invalid),
+ "Alignment must be a non-zero power of 2");
+
auto ContextDevices = syclContext.get_devices();
- assert(!ContextDevices.empty() && "Context can't be created without device");
if (std::none_of(ContextDevices.begin(), ContextDevices.end(),
[&syclDevice](device Dev) { return Dev == syclDevice; }))
throw exception(make_error_code(errc::invalid),
"Specified device is not contained by specified context.");
+
if (!syclDevice.has(getAspectByAllocationKind(kind)))
throw sycl::exception(
sycl::errc::feature_not_supported,
@@ -173,7 +178,8 @@ void *aligned_alloc(std::size_t alignment, std::size_t numBytes,
void *malloc(std::size_t numBytes, const device &syclDevice,
const context &syclContext, usm::alloc kind,
const property_list &propList) {
- return aligned_alloc(0, numBytes, syclDevice, syclContext, kind, propList);
+ return aligned_alloc(alignof(std::max_align_t), numBytes, syclDevice,
+ syclContext, kind, propList);
}
void *malloc(std::size_t numBytes, const queue &syclQueue, usm::alloc kind,
diff --git a/libsycl/unittests/CMakeLists.txt b/libsycl/unittests/CMakeLists.txt
index 597651abd0bfc..5d515c8db25cc 100644
--- a/libsycl/unittests/CMakeLists.txt
+++ b/libsycl/unittests/CMakeLists.txt
@@ -10,6 +10,7 @@ add_subdirectory(event)
add_subdirectory(platform)
add_subdirectory(program_manager)
add_subdirectory(queue)
+add_subdirectory(usm)
# The unit test executables are GoogleTest binaries that are discovered and run
# by lit, mirroring how liboffload runs its unit tests.
diff --git a/libsycl/unittests/mock/helpers.cpp b/libsycl/unittests/mock/helpers.cpp
index f12e8509f1dce..4177dffd94f01 100644
--- a/libsycl/unittests/mock/helpers.cpp
+++ b/libsycl/unittests/mock/helpers.cpp
@@ -354,4 +354,35 @@ void mock::MockLiboffload::initDefault() {
mock::releaseDummyHandle(Address);
return OL_SUCCESS;
});
+
+ ON_CALL(*this, olMemAllocAligned)
+ .WillByDefault([this](ol_device_handle_t Device,
+ ol_alloc_type_t AllocType, size_t Size,
+ size_t Alignment,
+ void **AllocationOut) -> ol_result_t {
+ EXPECT_NE(Device, nullptr);
+ EXPECT_TRUE(AllocType == OL_ALLOC_TYPE_DEVICE ||
+ AllocType == OL_ALLOC_TYPE_MANAGED);
+ EXPECT_GT(Size, 0);
+ EXPECT_GT(Alignment, 0);
+ EXPECT_EQ(Alignment & (Alignment - 1), 0);
+ EXPECT_NE(AllocationOut, nullptr);
+
+ *AllocationOut = mock::createDummyHandle<void *>();
+ return OL_SUCCESS;
+ });
+
+ ON_CALL(*this, olMemAllocAlignedHost)
+ .WillByDefault([this](ol_device_handle_t Device, size_t Size,
+ size_t Alignment,
+ void **AllocationOut) -> ol_result_t {
+ EXPECT_NE(Device, nullptr);
+ EXPECT_GT(Size, 0);
+ EXPECT_GT(Alignment, 0);
+ EXPECT_EQ(Alignment & (Alignment - 1), 0);
+ EXPECT_NE(AllocationOut, nullptr);
+
+ *AllocationOut = mock::createDummyHandle<void *>();
+ return OL_SUCCESS;
+ });
}
diff --git a/libsycl/unittests/mock/helpers.hpp b/libsycl/unittests/mock/helpers.hpp
index 174bd24a139b7..b4b47ebfb21ac 100644
--- a/libsycl/unittests/mock/helpers.hpp
+++ b/libsycl/unittests/mock/helpers.hpp
@@ -135,6 +135,12 @@ class MockLiboffload {
MOCK_METHOD(ol_result_t, olMemAllocHost,
(ol_device_handle_t Device, size_t Size, void **AllocationOut));
MOCK_METHOD(ol_result_t, olMemFree, (void *Address));
+ MOCK_METHOD(ol_result_t, olMemAllocAligned,
+ (ol_device_handle_t Device, ol_alloc_type_t AllocType,
+ size_t Size, size_t Alignment, void **AllocationOut));
+ MOCK_METHOD(ol_result_t, olMemAllocAlignedHost,
+ (ol_device_handle_t Device, size_t Size, size_t Alignment,
+ void **AllocationOut));
ol_result_t makeEmptyStrError(ol_errc_t Code) {
auto [Iterator, Flag] =
diff --git a/libsycl/unittests/mock/mock.cpp b/libsycl/unittests/mock/mock.cpp
index 02619752acf7c..bef01f3904514 100644
--- a/libsycl/unittests/mock/mock.cpp
+++ b/libsycl/unittests/mock/mock.cpp
@@ -146,3 +146,20 @@ ol_result_t olCreateEvent(ol_queue_handle_t Queue, ol_event_flags_t Flags,
ol_result_t olDestroyEvent(ol_event_handle_t Event) {
return mock::getMockLiboffload().olDestroyEvent(Event);
}
+
+ol_result_t olMemAllocAligned(ol_device_handle_t Device,
+ ol_alloc_type_t AllocType, size_t Size,
+ size_t Alignment, void **OutPtr) {
+ return mock::getMockLiboffload().olMemAllocAligned(Device, AllocType, Size,
+ Alignment, OutPtr);
+}
+
+ol_result_t olMemAllocAlignedHost(ol_device_handle_t Device, size_t Size,
+ size_t Alignment, void **OutPtr) {
+ return mock::getMockLiboffload().olMemAllocAlignedHost(Device, Size,
+ Alignment, OutPtr);
+}
+
+ol_result_t olMemFree(void *Address) {
+ return mock::getMockLiboffload().olMemFree(Address);
+}
diff --git a/libsycl/unittests/usm/CMakeLists.txt b/libsycl/unittests/usm/CMakeLists.txt
new file mode 100644
index 0000000000000..6f239e961102f
--- /dev/null
+++ b/libsycl/unittests/usm/CMakeLists.txt
@@ -0,0 +1,3 @@
+add_sycl_unittest(USMTests
+ alloc.cpp
+)
diff --git a/libsycl/unittests/usm/alloc.cpp b/libsycl/unittests/usm/alloc.cpp
new file mode 100644
index 0000000000000..ee4f101dc1c83
--- /dev/null
+++ b/libsycl/unittests/usm/alloc.cpp
@@ -0,0 +1,147 @@
+#include <mock/helpers.hpp>
+
+#include <sycl/__impl/device.hpp>
+#include <sycl/__impl/queue.hpp>
+#include <sycl/__impl/usm_functions.hpp>
+
+#include <detail/device_impl.hpp>
+#include <detail/queue_impl.hpp>
+
+#include <cstddef>
+#include <gmock/gmock.h>
+#include <gtest/gtest.h>
+
+using namespace sycl;
+using namespace ::testing;
+
+TEST(USMFunctions, DeviceAllocation) {
+ constexpr size_t NumBytes = 1024;
+ constexpr size_t Alignment = 256;
+ constexpr size_t DefaultAlign = alignof(std::max_align_t);
+
+ mock::MockWrapper Mock;
+ queue Q;
+ device Dev = Q.get_device();
+ context Ctx = Q.get_context();
+ ol_device_handle_t OLDev = detail::getSyclObjImpl(Dev)->getOLHandle();
+
+ // 1. Test malloc_device
+ void *DummyPtr1 = mock::createDummyHandle<void *>();
+ EXPECT_CALL(Mock.get(), olMemAllocAligned(OLDev, OL_ALLOC_TYPE_DEVICE,
+ NumBytes, DefaultAlign, _))
+ .Times(1)
+ .WillOnce([&](ol_device_handle_t Device, ol_alloc_type_t AllocType,
+ size_t Size, size_t Alignment,
+ void **OutPtr) -> ol_result_t {
+ *OutPtr = DummyPtr1;
+ return OL_SUCCESS;
+ });
+
+ void *Ptr1 = malloc_device(NumBytes, Dev, Ctx);
+ EXPECT_EQ(Ptr1, DummyPtr1);
+
+ EXPECT_CALL(Mock.get(), olMemFree(Ptr1))
+ .Times(1)
+ .WillOnce(Return(ol_result_t(OL_SUCCESS)));
+ free(Ptr1, Ctx);
+
+ // 2. Test aligned_alloc_device
+ void *DummyPtr2 = mock::createDummyHandle<void *>();
+ EXPECT_CALL(Mock.get(), olMemAllocAligned(OLDev, OL_ALLOC_TYPE_DEVICE,
+ NumBytes, Alignment, _))
+ .Times(1)
+ .WillOnce([&](ol_device_handle_t Device, ol_alloc_type_t AllocType,
+ size_t Size, size_t Alignment,
+ void **OutPtr) -> ol_result_t {
+ *OutPtr = DummyPtr2;
+ return OL_SUCCESS;
+ });
+
+ void *Ptr2 = aligned_alloc_device(Alignment, NumBytes, Q);
+ EXPECT_EQ(Ptr2, DummyPtr2);
+
+ EXPECT_CALL(Mock.get(), olMemFree(Ptr2))
+ .Times(1)
+ .WillOnce(Return(ol_result_t(OL_SUCCESS)));
+ free(Ptr2, Q);
+}
+
+TEST(USMFunctions, HostAllocation) {
+ constexpr size_t NumBytes = 512;
+ constexpr size_t Alignment = 64;
+ void *DummyPtr = mock::createDummyHandle<void *>();
+
+ mock::MockWrapper Mock;
+ queue Q;
+ context Ctx = Q.get_context();
+ ol_device_handle_t OLDev =
+ detail::getSyclObjImpl(Q.get_device())->getOLHandle();
+
+ EXPECT_CALL(Mock.get(), olMemAllocAlignedHost(OLDev, NumBytes, Alignment, _))
+ .Times(1)
+ .WillOnce([&](ol_device_handle_t Device, size_t Size, size_t Alignment,
+ void **OutPtr) -> ol_result_t {
+ *OutPtr = DummyPtr;
+ return OL_SUCCESS;
+ });
+
+ void *Ptr = aligned_alloc_host(Alignment, NumBytes, Ctx);
+ EXPECT_EQ(Ptr, DummyPtr);
+
+ EXPECT_CALL(Mock.get(), olMemFree(Ptr))
+ .Times(1)
+ .WillOnce(Return(ol_result_t(OL_SUCCESS)));
+ free(Ptr, Ctx);
+}
+
+TEST(USMFunctions, SharedAllocation) {
+ constexpr size_t NumBytes = 1024;
+ constexpr size_t Alignment = 128;
+ constexpr size_t DefaultAlign = alignof(std::max_align_t);
+
+ mock::MockWrapper Mock;
+ queue Q;
+ device Dev = Q.get_device();
+ context Ctx = Q.get_context();
+ ol_device_handle_t OLDev = detail::getSyclObjImpl(Dev)->getOLHandle();
+
+ // 1. Test malloc_shared
+ void *DummyPtr1 = mock::createDummyHandle<void *>();
+ EXPECT_CALL(Mock.get(), olMemAllocAligned(OLDev, OL_ALLOC_TYPE_MANAGED,
+ NumBytes, DefaultAlign, _))
+ .Times(1)
+ .WillOnce([&](ol_device_handle_t Device, ol_alloc_type_t AllocType,
+ size_t Size, size_t Alignment,
+ void **OutPtr) -> ol_result_t {
+ *OutPtr = DummyPtr1;
+ return OL_SUCCESS;
+ });
+
+ void *Ptr1 = malloc_shared(NumBytes, Dev, Ctx);
+ EXPECT_EQ(Ptr1, DummyPtr1);
+
+ EXPECT_CALL(Mock.get(), olMemFree(Ptr1))
+ .Times(1)
+ .WillOnce(Return(ol_result_t(OL_SUCCESS)));
+ free(Ptr1, Ctx);
+
+ // 2. Test aligned_alloc_shared
+ void *DummyPtr2 = mock::createDummyHandle<void *>();
+ EXPECT_CALL(Mock.get(), olMemAllocAligned(OLDev, OL_ALLOC_TYPE_MANAGED,
+ NumBytes, Alignment, _))
+ .Times(1)
+ .WillOnce([&](ol_device_handle_t Device, ol_alloc_type_t AllocType,
+ size_t Size, size_t Alignment,
+ void **OutPtr) -> ol_result_t {
+ *OutPtr = DummyPtr2;
+ return OL_SUCCESS;
+ });
+
+ void *Ptr2 = aligned_alloc_shared(Alignment, NumBytes, Q);
+ EXPECT_EQ(Ptr2, DummyPtr2);
+
+ EXPECT_CALL(Mock.get(), olMemFree(Ptr2))
+ .Times(1)
+ .WillOnce(Return(ol_result_t(OL_SUCCESS)));
+ free(Ptr2, Q);
+}
\ No newline at end of file
>From 119c396dafd1d7dc71faae8d944d835a7deeaeb2 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Wed, 12 Aug 2026 20:49:34 +0300
Subject: [PATCH 10/13] Move detail/common.hpp to src as its non-public
---
libsycl/{include/sycl/__impl => src}/detail/common.hpp | 2 +-
libsycl/src/usm_functions.cpp | 2 +-
2 files changed, 2 insertions(+), 2 deletions(-)
rename libsycl/{include/sycl/__impl => src}/detail/common.hpp (95%)
diff --git a/libsycl/include/sycl/__impl/detail/common.hpp b/libsycl/src/detail/common.hpp
similarity index 95%
rename from libsycl/include/sycl/__impl/detail/common.hpp
rename to libsycl/src/detail/common.hpp
index fe47f84bbad7b..e39cb4555a6ba 100644
--- a/libsycl/include/sycl/__impl/detail/common.hpp
+++ b/libsycl/src/detail/common.hpp
@@ -27,4 +27,4 @@ constexpr bool isPowerOf2(std::size_t n) { return (n & (n - 1)) == 0; }
_LIBSYCL_END_NAMESPACE_SYCL
-#endif // _LIBSYCL___IMPL_DETAIL_COMMON_HPP
\ No newline at end of file
+#endif // _LIBSYCL___IMPL_DETAIL_COMMON_HPP
diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp
index 7fb69b4269252..0bc62300ad1b1 100644
--- a/libsycl/src/usm_functions.cpp
+++ b/libsycl/src/usm_functions.cpp
@@ -8,9 +8,9 @@
#include <sycl/__impl/usm_functions.hpp>
+#include <detail/common.hpp>
#include <detail/device_impl.hpp>
#include <detail/offload/offload_utils.hpp>
-#include <sycl/__impl/detail/common.hpp>
#include <OffloadAPI.h>
>From 22649fd48e3b2b205bcd7a410971bb742c74acdc Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Wed, 12 Aug 2026 21:05:29 +0300
Subject: [PATCH 11/13] more missing eof
---
libsycl/unittests/usm/alloc.cpp | 2 +-
1 file changed, 1 insertion(+), 1 deletion(-)
diff --git a/libsycl/unittests/usm/alloc.cpp b/libsycl/unittests/usm/alloc.cpp
index ee4f101dc1c83..d1c61460c0ab4 100644
--- a/libsycl/unittests/usm/alloc.cpp
+++ b/libsycl/unittests/usm/alloc.cpp
@@ -144,4 +144,4 @@ TEST(USMFunctions, SharedAllocation) {
.Times(1)
.WillOnce(Return(ol_result_t(OL_SUCCESS)));
free(Ptr2, Q);
-}
\ No newline at end of file
+}
>From 224983a7d5d624e1299f17a8ea2f14be5d1f93fc Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Sun, 16 Aug 2026 14:00:10 +0300
Subject: [PATCH 12/13] Rebase & fix conflicts
---
libsycl/unittests/mock/mock.cpp | 6 +-----
1 file changed, 1 insertion(+), 5 deletions(-)
diff --git a/libsycl/unittests/mock/mock.cpp b/libsycl/unittests/mock/mock.cpp
index bef01f3904514..a40af182bc464 100644
--- a/libsycl/unittests/mock/mock.cpp
+++ b/libsycl/unittests/mock/mock.cpp
@@ -158,8 +158,4 @@ ol_result_t olMemAllocAlignedHost(ol_device_handle_t Device, size_t Size,
size_t Alignment, void **OutPtr) {
return mock::getMockLiboffload().olMemAllocAlignedHost(Device, Size,
Alignment, OutPtr);
-}
-
-ol_result_t olMemFree(void *Address) {
- return mock::getMockLiboffload().olMemFree(Address);
-}
+}
\ No newline at end of file
>From 4fd52685b841cecf3a9f8050d0eb34c931b20d24 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Sun, 16 Aug 2026 14:01:33 +0300
Subject: [PATCH 13/13] add missing eof to mock.cpp
---
libsycl/unittests/mock/mock.cpp | 2 +-
1 file changed, 1 insertion(+), 1 deletion(-)
diff --git a/libsycl/unittests/mock/mock.cpp b/libsycl/unittests/mock/mock.cpp
index a40af182bc464..0eb852aef7918 100644
--- a/libsycl/unittests/mock/mock.cpp
+++ b/libsycl/unittests/mock/mock.cpp
@@ -158,4 +158,4 @@ ol_result_t olMemAllocAlignedHost(ol_device_handle_t Device, size_t Size,
size_t Alignment, void **OutPtr) {
return mock::getMockLiboffload().olMemAllocAlignedHost(Device, Size,
Alignment, OutPtr);
-}
\ No newline at end of file
+}
More information about the llvm-commits
mailing list