https://github.com/KseniyaTikhomirova created https://github.com/llvm/llvm-project/pull/184111
Depends on https://github.com/llvm/llvm-project/pull/184110 This is part of the SYCL support upstreaming effort. The relevant RFCs can be found here: https://discourse.llvm.org/t/rfc-add-full-support-for-the-sycl-programming-model/74080 https://discourse.llvm.org/t/rfc-sycl-runtime-upstreaming/74479 >From fb131918cbbcbf208ed28cc7e6bc06c8d1893d8e Mon Sep 17 00:00:00 2001 From: "Tikhomirova, Kseniya" <[email protected]> Date: Fri, 27 Feb 2026 10:44:21 -0800 Subject: [PATCH] [libsycl] add USM alloc/free functions Signed-off-by: Tikhomirova, Kseniya <[email protected]> --- libsycl/docs/index.rst | 5 + .../include/sycl/__impl/usm_alloc_type.hpp | 25 ++ libsycl/include/sycl/__impl/usm_functions.hpp | 300 ++++++++++++++++++ libsycl/include/sycl/sycl.hpp | 1 + libsycl/src/CMakeLists.txt | 1 + libsycl/src/detail/device_impl.cpp | 7 + libsycl/src/detail/global_objects.cpp | 3 + libsycl/src/detail/global_objects.hpp | 13 + libsycl/src/detail/offload/offload_utils.cpp | 17 + libsycl/src/detail/offload/offload_utils.hpp | 8 + libsycl/src/ld-version-script.txt | 4 + libsycl/src/usm_functions.cpp | 130 ++++++++ libsycl/test/usm/alloc_functions.cpp | 124 ++++++++ 13 files changed, 638 insertions(+) create mode 100644 libsycl/include/sycl/__impl/usm_alloc_type.hpp create mode 100644 libsycl/include/sycl/__impl/usm_functions.hpp create mode 100644 libsycl/src/usm_functions.cpp create mode 100644 libsycl/test/usm/alloc_functions.cpp diff --git a/libsycl/docs/index.rst b/libsycl/docs/index.rst index 7a0d1aa406f61..512b1f8cb6195 100644 --- a/libsycl/docs/index.rst +++ b/libsycl/docs/index.rst @@ -108,3 +108,8 @@ TODO for added SYCL classes * ``context``: to implement get_info, properties & public constructors once context support is added to liboffload * ``queue``: to implement USM methods, to implement synchronization methods, to implement submit & copy with accessors (low priority), get_info & properties, ctors that accepts context (blocked by lack of liboffload support) * ``property_list``: to fully implement and integrate to 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) diff --git a/libsycl/include/sycl/__impl/usm_alloc_type.hpp b/libsycl/include/sycl/__impl/usm_alloc_type.hpp new file mode 100644 index 0000000000000..5455202754d0e --- /dev/null +++ b/libsycl/include/sycl/__impl/usm_alloc_type.hpp @@ -0,0 +1,25 @@ +//===----------------------------------------------------------------------===// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#ifndef _LIBSYCL___IMPL_USM_ALLOC_TYPE_HPP +#define _LIBSYCL___IMPL_USM_ALLOC_TYPE_HPP + +#include <sycl/__impl/detail/config.hpp> + +_LIBSYCL_BEGIN_NAMESPACE_SYCL + +namespace usm { + +// SYCL 2020 4.8.2. Kinds of unified shared memory. +enum class alloc : char { host = 0, device = 1, shared = 2, unknown = 3 }; + +} // namespace usm + +_LIBSYCL_END_NAMESPACE_SYCL + +#endif // _LIBSYCL___IMPL_USM_ALLOC_TYPE_HPP diff --git a/libsycl/include/sycl/__impl/usm_functions.hpp b/libsycl/include/sycl/__impl/usm_functions.hpp new file mode 100644 index 0000000000000..c4bba0c2b144c --- /dev/null +++ b/libsycl/include/sycl/__impl/usm_functions.hpp @@ -0,0 +1,300 @@ +//===----------------------------------------------------------------------===// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#ifndef _LIBSYCL___IMPL_USM_FUNCTIONS_HPP +#define _LIBSYCL___IMPL_USM_FUNCTIONS_HPP + +#include <sycl/__impl/detail/config.hpp> + +#include <sycl/__impl/context.hpp> +#include <sycl/__impl/queue.hpp> +#include <sycl/__impl/usm_alloc_type.hpp> + +_LIBSYCL_BEGIN_NAMESPACE_SYCL + +// SYCL 2020 4.8.3.2. Device allocation functions. + +/// Allocates device USM. +/// +/// \param numBytes allocation size that is specified in bytes. +/// \param syclDevice device that is used for allocation. +/// \param syclContext context that contains syclDevice or its parent device if +/// syclDevice is a subdevice. +/// \param propList properties for the memory 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 *_LIBSYCL_EXPORT malloc_device(std::size_t numBytes, + const device &syclDevice, + const context &syclContext, + const property_list &propList = {}); + +/// Allocates device USM. +/// +/// \param count allocation size that is specified in number of elements of +/// type T. +/// \param syclDevice device that is used for allocation. +/// \param syclContext context that contains syclDevice or its parent device if +/// syclDevice is a subdevice. +/// \param propList properties for the memory 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 = {}) { + // TODO: to rewrite with aligned_malloc_device once it's supported in + // liboffload. + return static_cast<T *>( + malloc_device(count * sizeof(T), syclDevice, syclContext, propList)); +} + +/// Allocates device USM. +/// +/// \param numBytes allocation size that is specified in bytes. +/// \param syclQueue queue that provides the device and context. +/// \param propList properties for the memory 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 *_LIBSYCL_EXPORT malloc_device(std::size_t numBytes, + const queue &syclQueue, + const property_list &propList = {}); + +/// Allocates device USM. +/// +/// \param count allocation size that is specified in number of elements of +/// type T. +/// \param syclQueue queue that provides the device and context. +/// \param propList properties for the memory 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); +} + +// SYCL 2020 4.8.3.3. Host allocation functions. + +/// Allocates host USM. +/// +/// \param numBytes allocation size that is specified in bytes. +/// \param syclContext context that should have access to the allocated memory. +/// \param propList properties for the memory 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 *_LIBSYCL_EXPORT malloc_host(std::size_t numBytes, + const context &syclContext, + const property_list &propList = {}); + +/// Allocates host USM. +/// +/// \param count allocation size that is specified in number of elements of +/// type T. +/// \param syclContext context that should have access to the allocated memory. +/// \param propList properties for the memory 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_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)); +} + +/// Allocates host USM. +/// +/// \param numBytes allocation size that is specified in bytes. +/// \param syclQueue queue that provides the context. +/// \param propList properties for the memory 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 *_LIBSYCL_EXPORT malloc_host(std::size_t numBytes, const queue &syclQueue, + const property_list &propList = {}); + +/// Allocates host USM. +/// +/// \param count allocation size that is specified in number of elements of +/// type T. +/// \param syclQueue queue that provides the context. +/// \param propList properties for the memory 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_host(std::size_t count, const queue &syclQueue, + const property_list &propList = {}) { + return malloc_host<T>(count, syclQueue.get_context(), propList); +} + +// SYCL 2020 4.8.3.4. Shared allocation functions. + +/// Allocates shared USM. +/// +/// \param numBytes allocation size that is specified in bytes. +/// \param syclDevice device that is used for allocation. +/// \param syclContext context that contains syclDevice or its parent device if +/// syclDevice is a subdevice. +/// \param propList properties for the memory 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 *_LIBSYCL_EXPORT malloc_shared(std::size_t numBytes, + const device &syclDevice, + const context &syclContext, + const property_list &propList = {}); + +/// Allocates shared USM. +/// +/// \param count allocation size that is specified in number of elements of +/// type T. +/// \param syclDevice device that is used for allocation. +/// \param syclContext context that contains syclDevice or its parent device if +/// syclDevice is a subdevice. +/// \param propList properties for the memory 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 + // liboffload. + return static_cast<T *>( + malloc_shared(count * sizeof(T), syclDevice, syclContext, propList)); +} + +/// Allocates shared USM. +/// +/// \param numBytes allocation size that is specified in bytes. +/// \param syclQueue queue that provides the device and context. +/// \param propList properties for the memory 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 *_LIBSYCL_EXPORT malloc_shared(std::size_t numBytes, + const queue &syclQueue, + const property_list &propList = {}); + +/// Allocates shared USM. +/// +/// \param count allocation size that is specified in number of elements of +/// type T. +/// \param syclQueue queue that provides the device and context. +/// \param propList properties for the memory 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); +} + +// SYCL 2020 4.8.3.5. Parameterized allocation functions + +/// Allocates USM of type `kind`. +/// +/// \param numBytes allocation size that is specified in bytes. +/// \param syclDevice device that is used for allocation. The syclDevice +/// parameter is ignored if kind is usm::alloc::host. +/// \param syclContext context that contains syclDevice or its parent device if +/// syclDevice is a subdevice. +/// \param kind type of memory to allocate. +/// \param propList properties for the memory 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. +void *_LIBSYCL_EXPORT 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 allocation size that is specified in number of elements of +/// type T. +/// \param syclDevice device that is used for allocation. The syclDevice +/// parameter is ignored if kind is usm::alloc::host. +/// \param syclContext context that contains syclDevice or its parent device if +/// syclDevice is a subdevice. +/// \param kind type of memory to allocate. +/// \param propList properties for the memory 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 allocation size that is specified in bytes. +/// \param syclQueue queue that provides the device and context. +/// \param kind type of memory to allocate. +/// \param propList properties for the memory 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. +void *_LIBSYCL_EXPORT malloc(std::size_t numBytes, const queue &syclQueue, + usm::alloc kind, + const property_list &propList = {}); + +/// Allocates USM of type `kind`. +/// +/// \param count allocation size that is specified in number of elements of +/// type T. +/// \param syclQueue queue that provides the device and context. +/// \param kind type of memory to allocate. +/// \param propList properties for the memory 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); +} + +// SYCL 2020 4.8.3.6. Memory deallocation functions + +/// Deallocate USM of any kind. +/// +/// \param ptr pointer that satisfies the following preconditions: points to +/// memory allocated against ctxt using one of the USM allocation routines, or +/// is a null pointer, ptr has not previously been deallocated; there are no +/// in-progress or enqueued commands using the memory pointed to by ptr. +/// \param ctxt context that is associated with ptr. +void _LIBSYCL_EXPORT free(void *ptr, const context &ctxt); + +/// Deallocate USM of any kind. +/// +/// Equivalent to free(ptr, q.get_context()). +/// +/// \param ptr pointer that satisfies the following preconditions: points to +/// memory allocated against ctxt using one of the USM allocation routines, or +/// is a null pointer, ptr has not previously been deallocated; there are no +/// in-progress or enqueued commands using the memory pointed to by ptr. +/// \param q queue to determine the context associated with ptr. +void _LIBSYCL_EXPORT free(void *ptr, const queue &q); + +_LIBSYCL_END_NAMESPACE_SYCL + +#endif // _LIBSYCL___IMPL_USM_FUNCTIONS_HPP diff --git a/libsycl/include/sycl/sycl.hpp b/libsycl/include/sycl/sycl.hpp index e1bd55e361561..3fcf088f45535 100644 --- a/libsycl/include/sycl/sycl.hpp +++ b/libsycl/include/sycl/sycl.hpp @@ -20,5 +20,6 @@ #include <sycl/__impl/exception.hpp> #include <sycl/__impl/platform.hpp> #include <sycl/__impl/queue.hpp> +#include <sycl/__impl/usm_functions.hpp> #endif // _LIBSYCL_SYCL_HPP diff --git a/libsycl/src/CMakeLists.txt b/libsycl/src/CMakeLists.txt index 1e4e4178bd66d..67ba7d28968de 100644 --- a/libsycl/src/CMakeLists.txt +++ b/libsycl/src/CMakeLists.txt @@ -88,6 +88,7 @@ set(LIBSYCL_SOURCES "device_selector.cpp" "platform.cpp" "queue.cpp" + "usm_functions.cpp" "detail/context_impl.cpp" "detail/device_impl.cpp" "detail/global_objects.cpp" diff --git a/libsycl/src/detail/device_impl.cpp b/libsycl/src/detail/device_impl.cpp index d12f97d0db864..4efc4d458c37e 100644 --- a/libsycl/src/detail/device_impl.cpp +++ b/libsycl/src/detail/device_impl.cpp @@ -25,6 +25,13 @@ bool DeviceImpl::has(aspect Aspect) const { case (aspect::emulated): case (aspect::host_debuggable): return false; + case (aspect::usm_device_allocations): + case (aspect::usm_host_allocations): + case (aspect::usm_shared_allocations): + // liboffload works with USM only and has no query to check support. We + // assume that USM is always supported otherwise lifoffload won't be able to + // work with device at all. + return true; default: // Other aspects are not implemented yet return false; diff --git a/libsycl/src/detail/global_objects.cpp b/libsycl/src/detail/global_objects.cpp index 35e32985e7cbb..d80be710268f8 100644 --- a/libsycl/src/detail/global_objects.cpp +++ b/libsycl/src/detail/global_objects.cpp @@ -53,3 +53,6 @@ std::vector<PlatformImplUPtr> &getPlatformCache() { } // namespace detail _LIBSYCL_END_NAMESPACE_SYCL + +extern "C" void __sycl_register_lib(void *) {} +extern "C" void __sycl_unregister_lib(void *) {} diff --git a/libsycl/src/detail/global_objects.hpp b/libsycl/src/detail/global_objects.hpp index 4535a254c6609..008cb01f4f355 100644 --- a/libsycl/src/detail/global_objects.hpp +++ b/libsycl/src/detail/global_objects.hpp @@ -16,6 +16,19 @@ #include <mutex> #include <vector> +// +++ Entry points referenced by the offload wrapper object { + +/// Executed as a part of current module's (.exe, .dll) static initialization. +/// Registers device executable images with the runtime. +extern "C" _LIBSYCL_EXPORT void __sycl_register_lib(void *); + +/// Executed as a part of current module's (.exe, .dll) static +/// de-initialization. +/// Unregisters device executable images with the runtime. +extern "C" _LIBSYCL_EXPORT void __sycl_unregister_lib(void *); + +// +++ } + _LIBSYCL_BEGIN_NAMESPACE_SYCL namespace detail { diff --git a/libsycl/src/detail/offload/offload_utils.cpp b/libsycl/src/detail/offload/offload_utils.cpp index 9a2609daddcee..e4e68eb83747e 100644 --- a/libsycl/src/detail/offload/offload_utils.cpp +++ b/libsycl/src/detail/offload/offload_utils.cpp @@ -88,5 +88,22 @@ info::device_type convertDeviceTypeToSYCL(ol_device_type_t DeviceType) { } } +ol_alloc_type_t convertUSMTypeToOL(usm::alloc USMType) { + switch (USMType) { + case usm::alloc::host: + return OL_ALLOC_TYPE_HOST; + case usm::alloc::device: + return OL_ALLOC_TYPE_DEVICE; + case usm::alloc::shared: + return OL_ALLOC_TYPE_MANAGED; + default: + // usm::alloc::unknown can be returned to user from get_pointer_type but it + // can't be converted to a valid backend type and there is no need to do + // that. + throw exception(sycl::make_error_code(sycl::errc::runtime), + "USM type is not supported"); + } +} + } // namespace detail _LIBSYCL_END_NAMESPACE_SYCL diff --git a/libsycl/src/detail/offload/offload_utils.hpp b/libsycl/src/detail/offload/offload_utils.hpp index e849ee137337f..1fa9d6d6f11e2 100644 --- a/libsycl/src/detail/offload/offload_utils.hpp +++ b/libsycl/src/detail/offload/offload_utils.hpp @@ -13,6 +13,7 @@ #include <sycl/__impl/detail/config.hpp> #include <sycl/__impl/exception.hpp> #include <sycl/__impl/info/device_type.hpp> +#include <sycl/__impl/usm_alloc_type.hpp> #include <OffloadAPI.h> @@ -102,6 +103,13 @@ ol_device_type_t convertDeviceTypeToOL(info::device_type DeviceType); /// \returns SYCL device type matching specified liboffload device type. info::device_type convertDeviceTypeToSYCL(ol_device_type_t DeviceType); +/// Converts SYCL USM type to liboffload type. +/// +/// \param DeviceType SYCL USM type. +/// +/// \returns ol_alloc_type_t matching specified SYCL USM type. +ol_alloc_type_t convertUSMTypeToOL(usm::alloc USMType); + /// Helper to map SYCL information descriptors to OL_<HANDLE>_INFO_<SMTH>. /// /// Typical usage: diff --git a/libsycl/src/ld-version-script.txt b/libsycl/src/ld-version-script.txt index a347d202a367f..eeb78e2cf59bf 100644 --- a/libsycl/src/ld-version-script.txt +++ b/libsycl/src/ld-version-script.txt @@ -15,6 +15,10 @@ _ZTSN4sycl*; /* typeinfo name */ _ZTVN4sycl*; /* vtable */ + /* Export offload image hooks */ + __sycl_register_lib; + __sycl_unregister_lib; + local: *; }; diff --git a/libsycl/src/usm_functions.cpp b/libsycl/src/usm_functions.cpp new file mode 100644 index 0000000000000..8bc525509f177 --- /dev/null +++ b/libsycl/src/usm_functions.cpp @@ -0,0 +1,130 @@ +//===----------------------------------------------------------------------===// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// + +#include <sycl/__impl/usm_functions.hpp> + +#include <detail/device_impl.hpp> +#include <detail/offload/offload_utils.hpp> + +#include <OffloadAPI.h> + +#include <algorithm> + +_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); +} + +// 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(), + [](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."); + return malloc(numBytes, ContextDevices[0], syclContext, usm::alloc::host, + 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 *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) { + switch (kind) { + case usm::alloc::host: + return aspect::usm_host_allocations; + case usm::alloc::device: + return aspect::usm_device_allocations; + case usm::alloc::shared: + return aspect::usm_shared_allocations; + default: + assert(false && + "Must be unreachable, usm::unknown allocation can't be requested"); + // usm::alloc::unknown can be returned to user from get_pointer_type but + // it can't be converted to a valid backend type and there is no need to + // do that. + throw exception(sycl::make_error_code(sycl::errc::runtime), + "USM type is not supported"); + } +} + +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 Result = detail::callNoCheck( + olMemAlloc, detail::getSyclObjImpl(syclDevice)->getOLHandle(), + detail::convertUSMTypeToOL(kind), numBytes, &Ptr); + assert(!!Result != !!Ptr && "Successful USM allocation can't return nullptr"); + 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); +} + +// SYCL 2020 4.8.3.6. Memory deallocation functions + +void free(void *ptr, const context &ctxt) { + std::ignore = ctxt; + detail::callAndThrow(olMemFree, ptr); +} + +void free(void *ptr, const queue &q) { return free(ptr, q.get_context()); } + +_LIBSYCL_END_NAMESPACE_SYCL diff --git a/libsycl/test/usm/alloc_functions.cpp b/libsycl/test/usm/alloc_functions.cpp new file mode 100644 index 0000000000000..f3ce8441ab580 --- /dev/null +++ b/libsycl/test/usm/alloc_functions.cpp @@ -0,0 +1,124 @@ +// REQUIRES: any-device +// RUN: %clangxx %sycl_options %s -o %t.out +// RUN: %t.out + +#include <sycl/sycl.hpp> + +#include <cstddef> +#include <iostream> +#include <tuple> + +using namespace sycl; + +constexpr size_t Align = 256; + +struct alignas(Align) Aligned { + int x; +}; + +int main() { + queue q; + context ctx = q.get_context(); + device d = q.get_device(); + + auto check = [&q](size_t Alignment, auto AllocFn, int Line = __builtin_LINE(), + int Case = 0) { + // First allocation might naturally be over-aligned. Do several of them to + // do the verification; + decltype(AllocFn()) Arr[10]; + for (auto *&Elem : Arr) + Elem = AllocFn(); + for (auto *Ptr : Arr) { + auto v = reinterpret_cast<uintptr_t>(Ptr); + if ((v & (Alignment - 1)) != 0) { + std::cout << "Failed at line " << Line << ", case " << Case + << std::endl; + assert(false && "Not properly aligned!"); + break; // To be used with commented out assert above. + } + } + for (auto *Ptr : Arr) + free(Ptr, q); + }; + + // The strictest (largest) fundamental alignment of any type is the alignment + // of max_align_t. This is, however, smaller than the minimal alignment + // returned by the underlyging runtime as of now. + constexpr size_t FAlign = alignof(std::max_align_t); + + auto CheckAll = [&](size_t Expected, auto Funcs, + int Line = __builtin_LINE()) { + std::apply( + [&](auto... Fs) { + int Case = 0; + (void)std::initializer_list<int>{ + (check(Expected, Fs, Line, Case++), 0)...}; + }, + Funcs); + }; + + auto MDevice = [&](auto... args) { + return malloc_device(sizeof(std::max_align_t), args...); + }; + CheckAll(FAlign, + std::tuple{[&]() { return MDevice(q); }, + [&]() { return MDevice(d, ctx); }, + [&]() { return MDevice(q, property_list{}); }, + [&]() { return MDevice(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{}); }}); + + if (d.has(aspect::usm_shared_allocations)) { + auto MShared = [&](auto... args) { + return malloc_shared(sizeof(std::max_align_t), args...); + }; + + CheckAll(FAlign, + std::tuple{[&]() { return MShared(q); }, + [&]() { return MShared(d, ctx); }, + [&]() { return MShared(q, property_list{}); }, + [&]() { return MShared(d, ctx, property_list{}); }}); + } + + auto TDevice = [&](auto... args) { + return malloc_device<Aligned>(1, args...); + }; + CheckAll(Align, std::tuple{[&]() { return TDevice(q); }, + [&]() { return TDevice(d, ctx); }}); + + auto THost = [&](auto... args) { return malloc_host<Aligned>(1, args...); }; + CheckAll(Align, std::tuple{[&]() { return THost(q); }, + [&]() { return THost(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 Malloc = [&](auto... args) { + return malloc(sizeof(std::max_align_t), args...); + }; + CheckAll( + FAlign, + std::tuple{ + [&]() { return Malloc(q, usm::alloc::host); }, + [&]() { return Malloc(d, ctx, usm::alloc::host); }, + [&]() { return Malloc(q, usm::alloc::host, property_list{}); }, + [&]() { return Malloc(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); }}); + + return 0; +} _______________________________________________ llvm-branch-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-branch-commits
