[llvm] [libsycl] USM Aligned allocation functions (PR #213468)
via llvm-commits
llvm-commits at lists.llvm.org
Sat Aug 1 09:36:40 PDT 2026
https://github.com/Robertkq created https://github.com/llvm/llvm-project/pull/213468
Adds the `aligned_*` versions of the USM functions, specifically for the `device` `host` `shared` and free standing `aligned_alloc`.
Most of the overloads slowly direct calls to each other until they reach `aligned_alloc`, also rewrote templated `malloc_device<T>` to call `aligned_alloc_device<T>`. Although I couldn't find a specification for this, it was a TODO left in the code, maybe the same is expected for the `host` and `shared` versions, to allocate aligned memory if possible? I see this as an implementation detail optimisation, but can be troublesome if the requested type does not have an alignment as power of 2.
Adds more LIT tests in `alloc_functions.cpp` to cover the aligned version.
>From ab886ffeaabce2c5a63a1e4ca1cc68b55e8ffe40 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Sun, 26 Jul 2026 15:20:17 +0300
Subject: [PATCH 1/6] 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 c30229f06b97b1c5a867c6396c621fdc721f5ba5 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Mon, 27 Jul 2026 19:47:57 +0300
Subject: [PATCH 2/6] 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 5b52ca66b50ffa80d15217a08161d28c2046e8dd Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Mon, 27 Jul 2026 20:26:58 +0300
Subject: [PATCH 3/6] 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 41e49b50f0a0efa80ff055e3a2437fce53338192 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Mon, 27 Jul 2026 22:13:14 +0300
Subject: [PATCH 4/6] 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 e68d89cc3579841c7eb646cba14221136bc13fbc Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Sat, 1 Aug 2026 19:25:33 +0300
Subject: [PATCH 5/6] 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 c00502f6e7e8f8b051be8126535c386fea0c4a6b Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Sat, 1 Aug 2026 19:31:05 +0300
Subject: [PATCH 6/6] 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`:
More information about the llvm-commits
mailing list