From b211e3b63bd0996f5a5c03cb32662572844a5d17 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Mon, 17 Jul 2023 10:53:33 -0400 Subject: [PATCH 01/12] Added make_device for HIP. Signed-off-by: Jack Kirk --- .../sycl/detail/backend_traits_hip.hpp | 12 ++++ sycl/include/sycl/ext/oneapi/backend/hip.hpp | 29 ++++++++-- .../ur/adapters/hip/device.cpp | 56 ++++++++++++++++++- sycl/source/backend.cpp | 2 + 4 files changed, 92 insertions(+), 7 deletions(-) diff --git a/sycl/include/sycl/detail/backend_traits_hip.hpp b/sycl/include/sycl/detail/backend_traits_hip.hpp index 90a5913ac0774..a6495b0c69b4d 100644 --- a/sycl/include/sycl/detail/backend_traits_hip.hpp +++ b/sycl/include/sycl/detail/backend_traits_hip.hpp @@ -95,6 +95,18 @@ template <> struct BackendReturn { using type = HIPstream; }; +template <> struct InteropFeatureSupportMap { + static constexpr bool MakePlatform = false; + static constexpr bool MakeDevice = true; + static constexpr bool MakeContext = false; + static constexpr bool MakeQueue = false; + static constexpr bool MakeEvent = false; + static constexpr bool MakeBuffer = false; + static constexpr bool MakeKernel = false; + static constexpr bool MakeKernelBundle = false; + static constexpr bool MakeImage = false; +}; + } // namespace detail } // __SYCL_INLINE_VER_NAMESPACE(_V1) } // namespace sycl diff --git a/sycl/include/sycl/ext/oneapi/backend/hip.hpp b/sycl/include/sycl/ext/oneapi/backend/hip.hpp index 35a8ae078e2a8..b402f2807a1ac 100644 --- a/sycl/include/sycl/ext/oneapi/backend/hip.hpp +++ b/sycl/include/sycl/ext/oneapi/backend/hip.hpp @@ -15,11 +15,12 @@ __SYCL_INLINE_VER_NAMESPACE(_V1) { template <> inline backend_return_t -get_native(const device &Obj) { - // TODO use SYCL 2020 exception when implemented - if (Obj.get_backend() != backend::ext_oneapi_hip) { - throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch", - PI_ERROR_INVALID_OPERATION); +get_native(const device &Obj) +{ + if (Obj.get_backend() != backend::ext_oneapi_hip) + { + throw sycl::exception(make_error_code(errc::backend_mismatch), + "Backends mismatch"); } // HIP uses a 32-bit int instead of an opaque pointer like other backends, // so we need a specialization with static_cast instead of reinterpret_cast. @@ -27,5 +28,23 @@ get_native(const device &Obj) { Obj.getNative()); } +template <> +inline device make_device( + const backend_input_t &BackendObject) +{ + auto devs = device::get_devices(info::device_type::gpu); + for (auto &dev : devs) + { + if (dev.get_backend() == backend::ext_oneapi_hip && + BackendObject == get_native(dev)) + { + return dev; + } + } + pi_native_handle NativeHandle = static_cast(BackendObject); + return detail::make_device(NativeHandle, + backend::ext_oneapi_hip); +} + } // __SYCL_INLINE_VER_NAMESPACE(_V1) } // namespace sycl diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/device.cpp b/sycl/plugins/unified_runtime/ur/adapters/hip/device.cpp index 866819ca3c07f..067dabd286eb1 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/device.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/device.cpp @@ -921,8 +921,60 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetNativeHandle( } UR_APIEXPORT ur_result_t UR_APICALL urDeviceCreateWithNativeHandle( - ur_native_handle_t, ur_platform_handle_t, - const ur_device_native_properties_t *, ur_device_handle_t *) { + ur_native_handle_t hNativeDevice, ur_platform_handle_t hPlatform, + const ur_device_native_properties_t *pProperties, + ur_device_handle_t *phDevice) +{ + std::ignore = pProperties; + + // We can't cast between ur_native_handle_t and hipDevice_t, so memcpy the bits + // instead + hipDevice_t HIPDevice = 0; + memcpy(&HIPDevice, &hNativeDevice, sizeof(hipDevice_t)); + + auto IsDevice = [=](std::unique_ptr &Dev) + { + return Dev->get() == HIPDevice; + }; + + // If a platform is provided just check if the device is in it + if (hPlatform) + { + auto SearchRes = std::find_if(begin(hPlatform->Devices), + end(hPlatform->Devices), IsDevice); + if (SearchRes != end(hPlatform->Devices)) + { + *phDevice = SearchRes->get(); + return UR_RESULT_SUCCESS; + } + } + + // Get list of platforms + uint32_t NumPlatforms = 0; + ur_result_t Result = urPlatformGet(0, nullptr, &NumPlatforms); + if (Result != UR_RESULT_SUCCESS) + return Result; + + ur_platform_handle_t *Plat = static_cast( + malloc(NumPlatforms * sizeof(ur_platform_handle_t))); + Result = urPlatformGet(NumPlatforms, Plat, nullptr); + if (Result != UR_RESULT_SUCCESS) + return Result; + + // Iterate through platforms to find device that matches nativeHandle + for (uint32_t j = 0; j < NumPlatforms; ++j) + { + auto SearchRes = + std::find_if(begin(Plat[j]->Devices), end(Plat[j]->Devices), IsDevice); + if (SearchRes != end(Plat[j]->Devices)) + { + *phDevice = static_cast((*SearchRes).get()); + return UR_RESULT_SUCCESS; + } + } + + // If the provided nativeHandle cannot be matched to an + // existing device return error return UR_RESULT_ERROR_INVALID_OPERATION; } diff --git a/sycl/source/backend.cpp b/sycl/source/backend.cpp index a187fcccc3150..4f4e147693de8 100644 --- a/sycl/source/backend.cpp +++ b/sycl/source/backend.cpp @@ -37,6 +37,8 @@ static const PluginPtr &getPlugin(backend Backend) { return pi::getPlugin(); case backend::ext_oneapi_cuda: return pi::getPlugin(); + case backend::ext_oneapi_hip: + return pi::getPlugin(); default: throw sycl::exception(sycl::make_error_code(sycl::errc::runtime), "getPlugin: Unsupported backend " + From b747278902cd5c6e39354626dfe7e9fce78a9958 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Thu, 20 Jul 2023 12:56:01 -0400 Subject: [PATCH 02/12] make_queue impl for hip. Signed-off-by: Jack Kirk --- .../sycl/detail/backend_traits_hip.hpp | 2 +- sycl/plugins/unified_runtime/pi2ur.hpp | 1 - .../unified_runtime/ur/adapters/hip/queue.cpp | 48 +++++++++++++++---- .../unified_runtime/ur/adapters/hip/queue.hpp | 8 +++- 4 files changed, 45 insertions(+), 14 deletions(-) diff --git a/sycl/include/sycl/detail/backend_traits_hip.hpp b/sycl/include/sycl/detail/backend_traits_hip.hpp index a6495b0c69b4d..cfc3bd776d1fa 100644 --- a/sycl/include/sycl/detail/backend_traits_hip.hpp +++ b/sycl/include/sycl/detail/backend_traits_hip.hpp @@ -99,7 +99,7 @@ template <> struct InteropFeatureSupportMap { static constexpr bool MakePlatform = false; static constexpr bool MakeDevice = true; static constexpr bool MakeContext = false; - static constexpr bool MakeQueue = false; + static constexpr bool MakeQueue = true; static constexpr bool MakeEvent = false; static constexpr bool MakeBuffer = false; static constexpr bool MakeKernel = false; diff --git a/sycl/plugins/unified_runtime/pi2ur.hpp b/sycl/plugins/unified_runtime/pi2ur.hpp index eea9855860bd9..27a6ca11584e5 100644 --- a/sycl/plugins/unified_runtime/pi2ur.hpp +++ b/sycl/plugins/unified_runtime/pi2ur.hpp @@ -1520,7 +1520,6 @@ inline pi_result piextQueueCreateWithNativeHandle( PI_ASSERT(Context, PI_ERROR_INVALID_CONTEXT); PI_ASSERT(NativeHandle, PI_ERROR_INVALID_VALUE); PI_ASSERT(Queue, PI_ERROR_INVALID_QUEUE); - PI_ASSERT(Device, PI_ERROR_INVALID_DEVICE); ur_context_handle_t UrContext = reinterpret_cast(Context); diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp b/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp index 19447bcf8ae93..24b16f5e61367 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp @@ -193,6 +193,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueRelease(ur_queue_handle_t hQueue) { try { std::unique_ptr QueueImpl(hQueue); + if (!hQueue->backendHasOwnership()) + return UR_RESULT_SUCCESS; + ScopedContext Active(hQueue->getContext()); hQueue->forEachStream([](hipStream_t S) { @@ -252,19 +255,44 @@ urQueueGetNativeHandle(ur_queue_handle_t hQueue, ur_queue_native_desc_t *, } /// Created a UR queue object from a HIP queue handle. -/// TODO: Implement this. -/// NOTE: The created UR object takes ownership of the native handle. +/// NOTE: The created UR object doesn't takes ownership of the native handle. /// /// \param[in] hNativeQueue The native handle to create UR queue object from. /// \param[in] hContext is the UR context of the queue. /// \param[out] phQueue Set to the UR queue object created from native handle. -/// \param pProperties->isNativeHandleOwned tells if SYCL RT should assume the -/// ownership of -/// the native handle, if it can. -/// -/// \return UR_RESULT_ERROR_UNSUPPORTED_FEATURE UR_APIEXPORT ur_result_t UR_APICALL urQueueCreateWithNativeHandle( - ur_native_handle_t, ur_context_handle_t, ur_device_handle_t, - const ur_queue_native_properties_t *, ur_queue_handle_t *) { - return UR_RESULT_ERROR_UNSUPPORTED_FEATURE; + ur_native_handle_t hNativeQueue, ur_context_handle_t hContext, + ur_device_handle_t hDevice, const ur_queue_native_properties_t *pProperties, + ur_queue_handle_t *phQueue) { + (void)pProperties; + (void)hDevice; + + unsigned int HIPFlags; + hipStream_t HIPStream = reinterpret_cast(hNativeQueue); + + auto Return = UR_CHECK_ERROR(hipStreamGetFlags(HIPStream, &HIPFlags)); + + ur_queue_flags_t Flags = 0; + if (HIPFlags == hipStreamDefault) + Flags = UR_QUEUE_FLAG_USE_DEFAULT_STREAM; + else if (HIPFlags == hipStreamNonBlocking) + Flags = UR_QUEUE_FLAG_SYNC_WITH_DEFAULT_STREAM; + else + detail::ur::die("Unknown hip stream"); + + std::vector ComputeHIPStreams(1, HIPStream); + std::vector TransferHIPStreams(0); + + // Create queue and set num_compute_streams to 1, as computeHIPStreams has + // valid stream + *phQueue = new ur_queue_handle_t_{std::move(ComputeHIPStreams), + std::move(TransferHIPStreams), + hContext, + hContext->getDevice(), + HIPFlags, + Flags, + /*backend_owns*/ false}; + (*phQueue)->NumComputeStreams = 1; + + return Return; } diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/queue.hpp b/sycl/plugins/unified_runtime/ur/adapters/hip/queue.hpp index ac8aeaf37c373..0d2d69f27c8fe 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/queue.hpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/queue.hpp @@ -49,11 +49,13 @@ struct ur_queue_handle_t_ { std::mutex ComputeStreamMutex; std::mutex TransferStreamMutex; std::mutex BarrierMutex; + bool HasOwnership; ur_queue_handle_t_(std::vector &&ComputeStreams, std::vector &&TransferStreams, ur_context_handle_t Context, ur_device_handle_t Device, - unsigned int Flags, ur_queue_flags_t URFlags) + unsigned int Flags, ur_queue_flags_t URFlags, + bool BackendOwns = true) : ComputeStreams{std::move(ComputeStreams)}, TransferStreams{std::move(TransferStreams)}, DelayCompute(this->ComputeStreams.size(), false), @@ -62,7 +64,7 @@ struct ur_queue_handle_t_ { Device{Device}, RefCount{1}, EventCount{0}, ComputeStreamIdx{0}, TransferStreamIdx{0}, NumComputeStreams{0}, NumTransferStreams{0}, LastSyncComputeStreams{0}, LastSyncTransferStreams{0}, Flags(Flags), - URFlags(URFlags) { + URFlags(URFlags), HasOwnership{BackendOwns} { urContextRetain(Context); urDeviceRetain(Device); } @@ -235,4 +237,6 @@ struct ur_queue_handle_t_ { uint32_t getReferenceCount() const noexcept { return RefCount; } uint32_t getNextEventId() noexcept { return ++EventCount; } + + bool backendHasOwnership() const noexcept { return HasOwnership; } }; From e515de5afb2f00a691e328d4da5e73c3c8e9c203 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Fri, 21 Jul 2023 07:45:20 -0400 Subject: [PATCH 03/12] Added make_event hip impl. Signed-off-by: Jack Kirk --- .../sycl/detail/backend_traits_hip.hpp | 2 +- .../unified_runtime/ur/adapters/hip/event.cpp | 26 ++++++++++++++++--- .../unified_runtime/ur/adapters/hip/event.hpp | 13 ++++++++++ 3 files changed, 36 insertions(+), 5 deletions(-) diff --git a/sycl/include/sycl/detail/backend_traits_hip.hpp b/sycl/include/sycl/detail/backend_traits_hip.hpp index cfc3bd776d1fa..efed103d16f05 100644 --- a/sycl/include/sycl/detail/backend_traits_hip.hpp +++ b/sycl/include/sycl/detail/backend_traits_hip.hpp @@ -100,7 +100,7 @@ template <> struct InteropFeatureSupportMap { static constexpr bool MakeDevice = true; static constexpr bool MakeContext = false; static constexpr bool MakeQueue = true; - static constexpr bool MakeEvent = false; + static constexpr bool MakeEvent = true; static constexpr bool MakeBuffer = false; static constexpr bool MakeKernel = false; static constexpr bool MakeKernelBundle = false; diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp b/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp index 93faf2def0ac5..cfaaa7b55cd12 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp @@ -15,7 +15,7 @@ ur_event_handle_t_::ur_event_handle_t_(ur_command_t Type, ur_context_handle_t Context, ur_queue_handle_t Queue, hipStream_t Stream, uint32_t StreamToken) - : CommandType{Type}, RefCount{1}, HasBeenWaitedOn{false}, IsRecorded{false}, + : CommandType{Type}, RefCount{1}, HasOwnership{true}, HasBeenWaitedOn{false}, IsRecorded{false}, IsStarted{false}, StreamToken{StreamToken}, EvEnd{nullptr}, EvStart{nullptr}, EvQueued{nullptr}, Queue{Queue}, Stream{Stream}, Context{Context} { @@ -36,6 +36,15 @@ ur_event_handle_t_::ur_event_handle_t_(ur_command_t Type, urContextRetain(Context); } +ur_event_handle_t_::ur_event_handle_t_(ur_context_handle_t Context, + hipEvent_t EventNative) + : CommandType{UR_COMMAND_EVENTS_WAIT}, RefCount{1}, HasOwnership{false}, + HasBeenWaitedOn{false}, IsRecorded{false}, IsStarted{false}, + StreamToken{std::numeric_limits::max()}, EvEnd{EventNative}, + EvStart{nullptr}, EvQueued{nullptr}, Queue{nullptr}, Context{Context} { + urContextRetain(Context); +} + ur_event_handle_t_::~ur_event_handle_t_() { if (Queue != nullptr) { urQueueRelease(Queue); @@ -160,6 +169,9 @@ ur_result_t ur_event_handle_t_::wait() { } ur_result_t ur_event_handle_t_::release() { + if (!backendHasOwnership()) + return UR_RESULT_SUCCESS; + assert(Queue != nullptr); UR_CHECK_ERROR(hipEventDestroy(EvEnd)); @@ -310,7 +322,13 @@ UR_APIEXPORT ur_result_t UR_APICALL urEventGetNativeHandle( /// /// \return UR_RESULT_ERROR_UNSUPPORTED_FEATURE UR_APIEXPORT ur_result_t UR_APICALL urEventCreateWithNativeHandle( - ur_native_handle_t, ur_context_handle_t, - const ur_event_native_properties_t *, ur_event_handle_t *) { - return UR_RESULT_ERROR_UNSUPPORTED_FEATURE; + ur_native_handle_t hNativeEvent, ur_context_handle_t hContext, + const ur_event_native_properties_t *pProperties, + ur_event_handle_t *phEvent) { + std::ignore = pProperties; + + *phEvent = ur_event_handle_t_::makeWithNative( + hContext, reinterpret_cast(hNativeEvent)); + + return UR_RESULT_SUCCESS; } diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/event.hpp b/sycl/plugins/unified_runtime/ur/adapters/hip/event.hpp index 5960f384cdfd5..c9eed93d1c279 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/event.hpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/event.hpp @@ -60,6 +60,8 @@ struct ur_event_handle_t_ { uint32_t getEventId() const noexcept { return EventId; } + bool backendHasOwnership() const noexcept { return HasOwnership; } + // Returns the counter time when the associated command(s) were enqueued uint64_t getQueuedTime() const; @@ -77,6 +79,11 @@ struct ur_event_handle_t_ { StreamToken); } + static ur_event_handle_t makeWithNative(ur_context_handle_t context, + hipEvent_t eventNative) { + return new ur_event_handle_t_(context, eventNative); + } + ur_result_t release(); ~ur_event_handle_t_(); @@ -88,10 +95,16 @@ struct ur_event_handle_t_ { ur_queue_handle_t Queue, hipStream_t Stream, uint32_t StreamToken); + // This constructor is private to force programmers to use the + // makeWithNative for event interop + ur_event_handle_t_(ur_context_handle_t Context, hipEvent_t EventNative); + ur_command_t CommandType; // The type of command associated with event. std::atomic_uint32_t RefCount; // Event reference count. + bool HasOwnership; // Signifies if event owns the native type. + bool HasBeenWaitedOn; // Signifies whether the event has been waited // on through a call to wait(), which implies // that it has completed. From b0f8f7ce5d8237f9375327f9d001b92b9d6960d4 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Fri, 21 Jul 2023 07:54:34 -0400 Subject: [PATCH 04/12] Update interop-hip.cpp test. Signed-off-by: Jack Kirk --- sycl/test/basic_tests/interop-hip.cpp | 4 ++++ 1 file changed, 4 insertions(+) diff --git a/sycl/test/basic_tests/interop-hip.cpp b/sycl/test/basic_tests/interop-hip.cpp index 58255c163536d..2c036a4ab0329 100644 --- a/sycl/test/basic_tests/interop-hip.cpp +++ b/sycl/test/basic_tests/interop-hip.cpp @@ -56,5 +56,9 @@ int main() { hip_event = get_native(Event); hip_queue = get_native(Queue); +device InteropDevice = make_device(hip_device); +event InteropEvent = make_event(hip_event, Context); +queue InteropQueue = make_queue(hip_queue, Context); + return 0; } From 1376d577c4b986f86ecab17c1ae3f26043083e6e Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Fri, 21 Jul 2023 11:39:33 -0400 Subject: [PATCH 05/12] Reverted urDeviceCreateWithNativeHandle change. urDeviceCreateWithNativeHandle can never be reached in the hip backend (see comment in code) so I removed the impl I added earlier. Signed-off-by: Jack Kirk --- sycl/include/sycl/ext/oneapi/backend/hip.hpp | 9 ++- .../ur/adapters/hip/device.cpp | 56 +------------------ 2 files changed, 8 insertions(+), 57 deletions(-) diff --git a/sycl/include/sycl/ext/oneapi/backend/hip.hpp b/sycl/include/sycl/ext/oneapi/backend/hip.hpp index b402f2807a1ac..890c1d4ca59d2 100644 --- a/sycl/include/sycl/ext/oneapi/backend/hip.hpp +++ b/sycl/include/sycl/ext/oneapi/backend/hip.hpp @@ -41,9 +41,12 @@ inline device make_device( return dev; } } - pi_native_handle NativeHandle = static_cast(BackendObject); - return detail::make_device(NativeHandle, - backend::ext_oneapi_hip); + // The ext_oneapi_hip platform(s) adds all n available devices where n + // is returned from call to `hipGetDeviceCount`. + // Hence if this code is reached then the requested device ordinal must + // not be visible to the driver. + throw sycl::exception(make_error_code(errc::invalid), + "Native device has an invalid ordinal."); } } // __SYCL_INLINE_VER_NAMESPACE(_V1) diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/device.cpp b/sycl/plugins/unified_runtime/ur/adapters/hip/device.cpp index 067dabd286eb1..866819ca3c07f 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/device.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/device.cpp @@ -921,60 +921,8 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetNativeHandle( } UR_APIEXPORT ur_result_t UR_APICALL urDeviceCreateWithNativeHandle( - ur_native_handle_t hNativeDevice, ur_platform_handle_t hPlatform, - const ur_device_native_properties_t *pProperties, - ur_device_handle_t *phDevice) -{ - std::ignore = pProperties; - - // We can't cast between ur_native_handle_t and hipDevice_t, so memcpy the bits - // instead - hipDevice_t HIPDevice = 0; - memcpy(&HIPDevice, &hNativeDevice, sizeof(hipDevice_t)); - - auto IsDevice = [=](std::unique_ptr &Dev) - { - return Dev->get() == HIPDevice; - }; - - // If a platform is provided just check if the device is in it - if (hPlatform) - { - auto SearchRes = std::find_if(begin(hPlatform->Devices), - end(hPlatform->Devices), IsDevice); - if (SearchRes != end(hPlatform->Devices)) - { - *phDevice = SearchRes->get(); - return UR_RESULT_SUCCESS; - } - } - - // Get list of platforms - uint32_t NumPlatforms = 0; - ur_result_t Result = urPlatformGet(0, nullptr, &NumPlatforms); - if (Result != UR_RESULT_SUCCESS) - return Result; - - ur_platform_handle_t *Plat = static_cast( - malloc(NumPlatforms * sizeof(ur_platform_handle_t))); - Result = urPlatformGet(NumPlatforms, Plat, nullptr); - if (Result != UR_RESULT_SUCCESS) - return Result; - - // Iterate through platforms to find device that matches nativeHandle - for (uint32_t j = 0; j < NumPlatforms; ++j) - { - auto SearchRes = - std::find_if(begin(Plat[j]->Devices), end(Plat[j]->Devices), IsDevice); - if (SearchRes != end(Plat[j]->Devices)) - { - *phDevice = static_cast((*SearchRes).get()); - return UR_RESULT_SUCCESS; - } - } - - // If the provided nativeHandle cannot be matched to an - // existing device return error + ur_native_handle_t, ur_platform_handle_t, + const ur_device_native_properties_t *, ur_device_handle_t *) { return UR_RESULT_ERROR_INVALID_OPERATION; } From 11a7f7457071367e9b1839e6c2304f5012c3a535 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Fri, 21 Jul 2023 12:07:01 -0400 Subject: [PATCH 06/12] Fix format. Signed-off-by: Jack Kirk --- sycl/include/sycl/ext/oneapi/backend/hip.hpp | 19 +++++++------------ .../unified_runtime/ur/adapters/hip/event.cpp | 8 ++++---- sycl/test/basic_tests/interop-hip.cpp | 6 +++--- 3 files changed, 14 insertions(+), 19 deletions(-) diff --git a/sycl/include/sycl/ext/oneapi/backend/hip.hpp b/sycl/include/sycl/ext/oneapi/backend/hip.hpp index 890c1d4ca59d2..3b767763d6ec9 100644 --- a/sycl/include/sycl/ext/oneapi/backend/hip.hpp +++ b/sycl/include/sycl/ext/oneapi/backend/hip.hpp @@ -15,10 +15,8 @@ __SYCL_INLINE_VER_NAMESPACE(_V1) { template <> inline backend_return_t -get_native(const device &Obj) -{ - if (Obj.get_backend() != backend::ext_oneapi_hip) - { +get_native(const device &Obj) { + if (Obj.get_backend() != backend::ext_oneapi_hip) { throw sycl::exception(make_error_code(errc::backend_mismatch), "Backends mismatch"); } @@ -30,23 +28,20 @@ get_native(const device &Obj) template <> inline device make_device( - const backend_input_t &BackendObject) -{ + const backend_input_t &BackendObject) { auto devs = device::get_devices(info::device_type::gpu); - for (auto &dev : devs) - { + for (auto &dev : devs) { if (dev.get_backend() == backend::ext_oneapi_hip && - BackendObject == get_native(dev)) - { + BackendObject == get_native(dev)) { return dev; } } // The ext_oneapi_hip platform(s) adds all n available devices where n - // is returned from call to `hipGetDeviceCount`. + // is returned from call to `hipGetDeviceCount`. // Hence if this code is reached then the requested device ordinal must // not be visible to the driver. throw sycl::exception(make_error_code(errc::invalid), - "Native device has an invalid ordinal."); + "Native device has an invalid ordinal."); } } // __SYCL_INLINE_VER_NAMESPACE(_V1) diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp b/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp index cfaaa7b55cd12..49753104dda70 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp @@ -15,10 +15,10 @@ ur_event_handle_t_::ur_event_handle_t_(ur_command_t Type, ur_context_handle_t Context, ur_queue_handle_t Queue, hipStream_t Stream, uint32_t StreamToken) - : CommandType{Type}, RefCount{1}, HasOwnership{true}, HasBeenWaitedOn{false}, IsRecorded{false}, - IsStarted{false}, StreamToken{StreamToken}, EvEnd{nullptr}, - EvStart{nullptr}, EvQueued{nullptr}, Queue{Queue}, Stream{Stream}, - Context{Context} { + : CommandType{Type}, RefCount{1}, HasOwnership{true}, + HasBeenWaitedOn{false}, IsRecorded{false}, IsStarted{false}, + StreamToken{StreamToken}, EvEnd{nullptr}, EvStart{nullptr}, + EvQueued{nullptr}, Queue{Queue}, Stream{Stream}, Context{Context} { bool ProfilingEnabled = Queue->URFlags & UR_QUEUE_FLAG_PROFILING_ENABLE; diff --git a/sycl/test/basic_tests/interop-hip.cpp b/sycl/test/basic_tests/interop-hip.cpp index 2c036a4ab0329..1077b76a81f88 100644 --- a/sycl/test/basic_tests/interop-hip.cpp +++ b/sycl/test/basic_tests/interop-hip.cpp @@ -56,9 +56,9 @@ int main() { hip_event = get_native(Event); hip_queue = get_native(Queue); -device InteropDevice = make_device(hip_device); -event InteropEvent = make_event(hip_event, Context); -queue InteropQueue = make_queue(hip_queue, Context); + device InteropDevice = make_device(hip_device); + event InteropEvent = make_event(hip_event, Context); + queue InteropQueue = make_queue(hip_queue, Context); return 0; } From 7ba37c510c17a76b2ff6d0b7e53224f3ac51bdb8 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Fri, 21 Jul 2023 12:12:04 -0400 Subject: [PATCH 07/12] Fix format. Signed-off-by: Jack Kirk --- sycl/include/sycl/ext/oneapi/backend/hip.hpp | 2 +- sycl/test/basic_tests/interop-hip.cpp | 6 +++--- 2 files changed, 4 insertions(+), 4 deletions(-) diff --git a/sycl/include/sycl/ext/oneapi/backend/hip.hpp b/sycl/include/sycl/ext/oneapi/backend/hip.hpp index 3b767763d6ec9..112595c7a2b31 100644 --- a/sycl/include/sycl/ext/oneapi/backend/hip.hpp +++ b/sycl/include/sycl/ext/oneapi/backend/hip.hpp @@ -41,7 +41,7 @@ inline device make_device( // Hence if this code is reached then the requested device ordinal must // not be visible to the driver. throw sycl::exception(make_error_code(errc::invalid), - "Native device has an invalid ordinal."); + "Native device has an invalid ordinal."); } } // __SYCL_INLINE_VER_NAMESPACE(_V1) diff --git a/sycl/test/basic_tests/interop-hip.cpp b/sycl/test/basic_tests/interop-hip.cpp index 1077b76a81f88..579139ff4300b 100644 --- a/sycl/test/basic_tests/interop-hip.cpp +++ b/sycl/test/basic_tests/interop-hip.cpp @@ -56,9 +56,9 @@ int main() { hip_event = get_native(Event); hip_queue = get_native(Queue); - device InteropDevice = make_device(hip_device); - event InteropEvent = make_event(hip_event, Context); - queue InteropQueue = make_queue(hip_queue, Context); + device InteropDevice = make_device(hip_device); + event InteropEvent = make_event(hip_event, Context); + queue InteropQueue = make_queue(hip_queue, Context); return 0; } From b593ecb590f228288eb5e518d26e84a11ea0bea2 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Tue, 25 Jul 2023 14:41:19 -0400 Subject: [PATCH 08/12] Revert sycl exception change. Update comments. Signed-off-by: Jack Kirk --- sycl/include/sycl/ext/oneapi/backend/hip.hpp | 5 +++-- sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp | 5 +---- 2 files changed, 4 insertions(+), 6 deletions(-) diff --git a/sycl/include/sycl/ext/oneapi/backend/hip.hpp b/sycl/include/sycl/ext/oneapi/backend/hip.hpp index 5b63c9055ce2d..20526932fcbd6 100644 --- a/sycl/include/sycl/ext/oneapi/backend/hip.hpp +++ b/sycl/include/sycl/ext/oneapi/backend/hip.hpp @@ -16,9 +16,10 @@ inline namespace _V1 { template <> inline backend_return_t get_native(const device &Obj) { + // TODO swap with SYCL 2020 exception when in ABI-break window if (Obj.get_backend() != backend::ext_oneapi_hip) { - throw sycl::exception(make_error_code(errc::backend_mismatch), - "Backends mismatch"); + throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch", + PI_ERROR_INVALID_OPERATION); } // HIP uses a 32-bit int instead of an opaque pointer like other backends, // so we need a specialization with static_cast instead of reinterpret_cast. diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp b/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp index 49753104dda70..0467ede65b32d 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/event.cpp @@ -314,13 +314,10 @@ UR_APIEXPORT ur_result_t UR_APICALL urEventGetNativeHandle( } /// Created a UR event object from a HIP event handle. -/// TODO: Implement this. -/// NOTE: The created UR object takes ownership of the native handle. +/// NOTE: The created UR object doesn't take ownership of the native handle. /// /// \param[in] hNativeEvent The native handle to create UR event object from. /// \param[out] phEvent Set to the UR event object created from native handle. -/// -/// \return UR_RESULT_ERROR_UNSUPPORTED_FEATURE UR_APIEXPORT ur_result_t UR_APICALL urEventCreateWithNativeHandle( ur_native_handle_t hNativeEvent, ur_context_handle_t hContext, const ur_event_native_properties_t *pProperties, From 1f0216338d6c55c0e9bceff4cf095dae0285ac30 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Mon, 28 Aug 2023 06:55:18 -0400 Subject: [PATCH 09/12] DPC++ runtime decides native ownership. Signed-off-by: Jack Kirk --- sycl/include/sycl/backend.hpp | 3 ++- sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp | 3 +-- sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp | 3 +-- 3 files changed, 4 insertions(+), 5 deletions(-) diff --git a/sycl/include/sycl/backend.hpp b/sycl/include/sycl/backend.hpp index 445a44f59acda..ee4153e59962a 100644 --- a/sycl/include/sycl/backend.hpp +++ b/sycl/include/sycl/backend.hpp @@ -298,8 +298,9 @@ std::enable_if_t::MakeQueue == true, make_queue(const typename backend_traits::template input_type &BackendObject, const context &TargetContext, const async_handler Handler = {}) { + auto KeepOwnership = (Backend == backend::ext_oneapi_cuda || Backend == backend::ext_oneapi_hip) ? true : false; return detail::make_queue(detail::pi::cast(BackendObject), - false, TargetContext, nullptr, false, {}, Handler, + false, TargetContext, nullptr, KeepOwnership, {}, Handler, Backend); } diff --git a/sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp b/sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp index 4127059c58f66..e39f7b974a56a 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp @@ -242,7 +242,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueCreateWithNativeHandle( ur_native_handle_t hNativeQueue, ur_context_handle_t hContext, ur_device_handle_t hDevice, const ur_queue_native_properties_t *pProperties, ur_queue_handle_t *phQueue) { - (void)pProperties; (void)hDevice; unsigned int CuFlags; @@ -269,7 +268,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueCreateWithNativeHandle( hContext->getDevice(), CuFlags, Flags, - /*backend_owns*/ false}; + /*backend_owns*/ pProperties->isNativeHandleOwned}; (*phQueue)->NumComputeStreams = 1; return Return; diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp b/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp index 766fe5d1a8b86..4c38467f3dcb9 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp @@ -264,7 +264,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueCreateWithNativeHandle( ur_native_handle_t hNativeQueue, ur_context_handle_t hContext, ur_device_handle_t hDevice, const ur_queue_native_properties_t *pProperties, ur_queue_handle_t *phQueue) { - (void)pProperties; (void)hDevice; unsigned int HIPFlags; @@ -291,7 +290,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueCreateWithNativeHandle( hContext->getDevice(), HIPFlags, Flags, - /*backend_owns*/ false}; + /*backend_owns*/ pProperties->isNativeHandleOwned}; (*phQueue)->NumComputeStreams = 1; return Return; From d1a164e78930b2f849cbcc68884157146e9d7f80 Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Mon, 28 Aug 2023 07:21:13 -0400 Subject: [PATCH 10/12] Lint. Signed-off-by: Jack Kirk --- sycl/include/sycl/backend.hpp | 9 ++++++--- .../unified_runtime/ur/adapters/cuda/queue.cpp | 15 ++++++++------- .../unified_runtime/ur/adapters/hip/queue.cpp | 15 ++++++++------- 3 files changed, 22 insertions(+), 17 deletions(-) diff --git a/sycl/include/sycl/backend.hpp b/sycl/include/sycl/backend.hpp index ee4153e59962a..fa758ae049ae6 100644 --- a/sycl/include/sycl/backend.hpp +++ b/sycl/include/sycl/backend.hpp @@ -298,10 +298,13 @@ std::enable_if_t::MakeQueue == true, make_queue(const typename backend_traits::template input_type &BackendObject, const context &TargetContext, const async_handler Handler = {}) { - auto KeepOwnership = (Backend == backend::ext_oneapi_cuda || Backend == backend::ext_oneapi_hip) ? true : false; + auto KeepOwnership = (Backend == backend::ext_oneapi_cuda || + Backend == backend::ext_oneapi_hip) + ? true + : false; return detail::make_queue(detail::pi::cast(BackendObject), - false, TargetContext, nullptr, KeepOwnership, {}, Handler, - Backend); + false, TargetContext, nullptr, KeepOwnership, {}, + Handler, Backend); } template diff --git a/sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp b/sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp index e39f7b974a56a..2252d996b3a67 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/cuda/queue.cpp @@ -262,13 +262,14 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueCreateWithNativeHandle( // Create queue and set num_compute_streams to 1, as computeCuStreams has // valid stream - *phQueue = new ur_queue_handle_t_{std::move(ComputeCuStreams), - std::move(TransferCuStreams), - hContext, - hContext->getDevice(), - CuFlags, - Flags, - /*backend_owns*/ pProperties->isNativeHandleOwned}; + *phQueue = + new ur_queue_handle_t_{std::move(ComputeCuStreams), + std::move(TransferCuStreams), + hContext, + hContext->getDevice(), + CuFlags, + Flags, + /*backend_owns*/ pProperties->isNativeHandleOwned}; (*phQueue)->NumComputeStreams = 1; return Return; diff --git a/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp b/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp index 4c38467f3dcb9..d83b8016ee300 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/hip/queue.cpp @@ -284,13 +284,14 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueCreateWithNativeHandle( // Create queue and set num_compute_streams to 1, as computeHIPStreams has // valid stream - *phQueue = new ur_queue_handle_t_{std::move(ComputeHIPStreams), - std::move(TransferHIPStreams), - hContext, - hContext->getDevice(), - HIPFlags, - Flags, - /*backend_owns*/ pProperties->isNativeHandleOwned}; + *phQueue = + new ur_queue_handle_t_{std::move(ComputeHIPStreams), + std::move(TransferHIPStreams), + hContext, + hContext->getDevice(), + HIPFlags, + Flags, + /*backend_owns*/ pProperties->isNativeHandleOwned}; (*phQueue)->NumComputeStreams = 1; return Return; From 1ba72f51cc690e1338dd254dd29775a1237ebce7 Mon Sep 17 00:00:00 2001 From: JackAKirk Date: Tue, 29 Aug 2023 15:37:28 +0100 Subject: [PATCH 11/12] Update sycl/include/sycl/backend.hpp Co-authored-by: Sergey Semenov --- sycl/include/sycl/backend.hpp | 6 ++---- 1 file changed, 2 insertions(+), 4 deletions(-) diff --git a/sycl/include/sycl/backend.hpp b/sycl/include/sycl/backend.hpp index fa758ae049ae6..aa10f738b86d8 100644 --- a/sycl/include/sycl/backend.hpp +++ b/sycl/include/sycl/backend.hpp @@ -298,10 +298,8 @@ std::enable_if_t::MakeQueue == true, make_queue(const typename backend_traits::template input_type &BackendObject, const context &TargetContext, const async_handler Handler = {}) { - auto KeepOwnership = (Backend == backend::ext_oneapi_cuda || - Backend == backend::ext_oneapi_hip) - ? true - : false; + auto KeepOwnership = Backend == backend::ext_oneapi_cuda || + Backend == backend::ext_oneapi_hip; return detail::make_queue(detail::pi::cast(BackendObject), false, TargetContext, nullptr, KeepOwnership, {}, Handler, Backend); From 828f54cab32e9c4b611de40229cb81cc4bfb02ee Mon Sep 17 00:00:00 2001 From: Jack Kirk Date: Tue, 29 Aug 2023 10:54:53 -0400 Subject: [PATCH 12/12] Format. Signed-off-by: Jack Kirk --- sycl/include/sycl/backend.hpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/include/sycl/backend.hpp b/sycl/include/sycl/backend.hpp index aa10f738b86d8..c47f077b4319a 100644 --- a/sycl/include/sycl/backend.hpp +++ b/sycl/include/sycl/backend.hpp @@ -298,8 +298,8 @@ std::enable_if_t::MakeQueue == true, make_queue(const typename backend_traits::template input_type &BackendObject, const context &TargetContext, const async_handler Handler = {}) { - auto KeepOwnership = Backend == backend::ext_oneapi_cuda || - Backend == backend::ext_oneapi_hip; + auto KeepOwnership = + Backend == backend::ext_oneapi_cuda || Backend == backend::ext_oneapi_hip; return detail::make_queue(detail::pi::cast(BackendObject), false, TargetContext, nullptr, KeepOwnership, {}, Handler, Backend);