diff --git a/sycl/cmake/modules/FetchUnifiedRuntime.cmake b/sycl/cmake/modules/FetchUnifiedRuntime.cmake index ee17573ba6bb7..058f89f5bbc16 100644 --- a/sycl/cmake/modules/FetchUnifiedRuntime.cmake +++ b/sycl/cmake/modules/FetchUnifiedRuntime.cmake @@ -116,14 +116,8 @@ if(SYCL_UR_USE_FETCH_CONTENT) CACHE PATH "Path to external '${name}' adapter source dir" FORCE) endfunction() - set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git") - # commit 0247d0966ca8c5d1e3245f375e48e6c997bed9af - # Merge: 675dd292 04ffc909 - # Author: aarongreig - # Date: Tue Oct 1 17:10:58 2024 +0100 - # Merge pull request #2154 from npmiller/fix-graph-exce - # [CUDA][HIP] Fix exceptions throwing from adapter - set(UNIFIED_RUNTIME_TAG 0247d0966ca8c5d1e3245f375e48e6c997bed9af) + set(UNIFIED_RUNTIME_REPO "https://github.com/Bensuo/unified-runtime.git") + set(UNIFIED_RUNTIME_TAG ewan/ur_dyn_events) set(UMF_BUILD_EXAMPLES OFF CACHE INTERNAL "EXAMPLES") # Due to the use of dependentloadflag and no installer for UMF and hwloc we need diff --git a/sycl/source/detail/memory_manager.cpp b/sycl/source/detail/memory_manager.cpp index 5c86e5e21d216..5df6f61c45d7c 100644 --- a/sycl/source/detail/memory_manager.cpp +++ b/sycl/source/detail/memory_manager.cpp @@ -1343,11 +1343,11 @@ void MemoryManager::ext_oneapi_copyD2D_cmd_buffer( } if (1 == DimDst && 1 == DimSrc) { - Adapter->call( - CommandBuffer, sycl::detail::ur::cast(SrcMem), - sycl::detail::ur::cast(DstMem), SrcXOffBytes, - DstXOffBytes, SrcAccessRangeWidthBytes, Deps.size(), Deps.data(), - OutSyncPoint); + Adapter->call(urCommandBufferAppendMemBufferCopyExp, CommandBuffer, + sycl::detail::ur::cast(SrcMem), + sycl::detail::ur::cast(DstMem), SrcXOffBytes, + DstXOffBytes, SrcAccessRangeWidthBytes, Deps.size(), + Deps.data(), 0, nullptr, OutSyncPoint, nullptr, nullptr); } else { // passing 0 for pitches not allowed. Because clEnqueueCopyBufferRect will // calculate both src and dest pitch using region[0], which is not correct @@ -1369,11 +1369,12 @@ void MemoryManager::ext_oneapi_copyD2D_cmd_buffer( SrcAccessRange[SrcPos.YTerm], SrcAccessRange[SrcPos.ZTerm]}; - Adapter->call( - CommandBuffer, sycl::detail::ur::cast(SrcMem), - sycl::detail::ur::cast(DstMem), SrcOrigin, DstOrigin, - Region, SrcRowPitch, SrcSlicePitch, DstRowPitch, DstSlicePitch, - Deps.size(), Deps.data(), OutSyncPoint); + Adapter->call(urCommandBufferAppendMemBufferCopyRectExp, CommandBuffer, + sycl::detail::ur::cast(SrcMem), + sycl::detail::ur::cast(DstMem), SrcOrigin, + DstOrigin, Region, SrcRowPitch, SrcSlicePitch, DstRowPitch, + DstSlicePitch, Deps.size(), Deps.data(), 0, nullptr, + OutSyncPoint, nullptr, nullptr); } } @@ -1407,11 +1408,11 @@ void MemoryManager::ext_oneapi_copyD2H_cmd_buffer( } if (1 == DimDst && 1 == DimSrc) { - ur_result_t Result = - Adapter->call_nocheck( - CommandBuffer, sycl::detail::ur::cast(SrcMem), - SrcXOffBytes, SrcAccessRangeWidthBytes, DstMem + DstXOffBytes, - Deps.size(), Deps.data(), OutSyncPoint); + ur_result_t Result = Adapter->call_nocheck( + urCommandBufferAppendMemBufferReadExp, CommandBuffer, + sycl::detail::ur::cast(SrcMem), SrcXOffBytes, + SrcAccessRangeWidthBytes, DstMem + DstXOffBytes, Deps.size(), + Deps.data(), 0, nullptr, OutSyncPoint, nullptr, nullptr); if (Result == UR_RESULT_ERROR_UNSUPPORTED_FEATURE) { throw sycl::exception( @@ -1436,12 +1437,12 @@ void MemoryManager::ext_oneapi_copyD2H_cmd_buffer( SrcAccessRange[SrcPos.YTerm], SrcAccessRange[SrcPos.ZTerm]}; - ur_result_t Result = Adapter->call_nocheck< - UrApiKind::urCommandBufferAppendMemBufferReadRectExp>( - CommandBuffer, sycl::detail::ur::cast(SrcMem), - BufferOffset, HostOffset, RectRegion, BufferRowPitch, BufferSlicePitch, - HostRowPitch, HostSlicePitch, DstMem, Deps.size(), Deps.data(), - OutSyncPoint); + ur_result_t Result = Adapter->call_nocheck( + urCommandBufferAppendMemBufferReadRectExp, CommandBuffer, + sycl::detail::ur::cast(SrcMem), BufferOffset, + HostOffset, RectRegion, BufferRowPitch, BufferSlicePitch, HostRowPitch, + HostSlicePitch, DstMem, Deps.size(), Deps.data(), 0, nullptr, + OutSyncPoint, nullptr, nullptr); if (Result == UR_RESULT_ERROR_UNSUPPORTED_FEATURE) { throw sycl::exception( sycl::make_error_code(sycl::errc::feature_not_supported), @@ -1482,12 +1483,11 @@ void MemoryManager::ext_oneapi_copyH2D_cmd_buffer( } if (1 == DimDst && 1 == DimSrc) { - ur_result_t Result = - Adapter - ->call_nocheck( - CommandBuffer, sycl::detail::ur::cast(DstMem), - DstXOffBytes, DstAccessRangeWidthBytes, SrcMem + SrcXOffBytes, - Deps.size(), Deps.data(), OutSyncPoint); + ur_result_t Result = Adapter->call_nocheck( + urCommandBufferAppendMemBufferWriteExp, CommandBuffer, + sycl::detail::ur::cast(DstMem), DstXOffBytes, + DstAccessRangeWidthBytes, SrcMem + SrcXOffBytes, Deps.size(), + Deps.data(), 0, nullptr, OutSyncPoint, nullptr, nullptr); if (Result == UR_RESULT_ERROR_UNSUPPORTED_FEATURE) { throw sycl::exception( @@ -1512,12 +1512,12 @@ void MemoryManager::ext_oneapi_copyH2D_cmd_buffer( DstAccessRange[DstPos.YTerm], DstAccessRange[DstPos.ZTerm]}; - ur_result_t Result = Adapter->call_nocheck< - UrApiKind::urCommandBufferAppendMemBufferWriteRectExp>( - CommandBuffer, sycl::detail::ur::cast(DstMem), - BufferOffset, HostOffset, RectRegion, BufferRowPitch, BufferSlicePitch, - HostRowPitch, HostSlicePitch, SrcMem, Deps.size(), Deps.data(), - OutSyncPoint); + ur_result_t Result = Adapter->call_nocheck( + urCommandBufferAppendMemBufferWriteRectExp, CommandBuffer, + sycl::detail::ur::cast(DstMem), BufferOffset, + HostOffset, RectRegion, BufferRowPitch, BufferSlicePitch, HostRowPitch, + HostSlicePitch, SrcMem, Deps.size(), Deps.data(), 0, nullptr, + OutSyncPoint, nullptr, nullptr); if (Result == UR_RESULT_ERROR_UNSUPPORTED_FEATURE) { throw sycl::exception( @@ -1539,10 +1539,9 @@ void MemoryManager::ext_oneapi_copy_usm_cmd_buffer( "NULL pointer argument in memory copy operation."); const AdapterPtr &Adapter = Context->getAdapter(); - ur_result_t Result = - Adapter->call_nocheck( - CommandBuffer, DstMem, SrcMem, Len, Deps.size(), Deps.data(), - OutSyncPoint); + ur_result_t Result = Adapter->call_nocheck( + urCommandBufferAppendUSMMemcpyExp, CommandBuffer, DstMem, SrcMem, Len, + Deps.size(), Deps.data(), 0, nullptr, OutSyncPoint, nullptr, nullptr); if (Result == UR_RESULT_ERROR_UNSUPPORTED_FEATURE) { throw sycl::exception( sycl::make_error_code(sycl::errc::feature_not_supported), @@ -1564,9 +1563,9 @@ void MemoryManager::ext_oneapi_fill_usm_cmd_buffer( "NULL pointer argument in memory fill operation."); const AdapterPtr &Adapter = Context->getAdapter(); - Adapter->call( - CommandBuffer, DstMem, Pattern.data(), Pattern.size(), Len, Deps.size(), - Deps.data(), OutSyncPoint); + Adapter->call(urCommandBufferAppendUSMFillExp, CommandBuffer, DstMem, + Pattern.data(), Pattern.size(), Len, Deps.size(), Deps.data(), 0, + nullptr, OutSyncPoint, nullptr, nullptr); } void MemoryManager::ext_oneapi_fill_cmd_buffer( @@ -1594,10 +1593,11 @@ void MemoryManager::ext_oneapi_fill_cmd_buffer( size_t RangeMultiplier = AccessRange[0] * AccessRange[1] * AccessRange[2]; if (RangesUsable && OffsetUsable) { - Adapter->call( - CommandBuffer, ur::cast(Mem), Pattern, PatternSize, - AccessOffset[0] * ElementSize, RangeMultiplier * ElementSize, - Deps.size(), Deps.data(), OutSyncPoint); + Adapter->call(urCommandBufferAppendMemBufferFillExp, CommandBuffer, + ur::cast(Mem), Pattern, PatternSize, + AccessOffset[0] * ElementSize, RangeMultiplier * ElementSize, + Deps.size(), Deps.data(), 0, nullptr, OutSyncPoint, nullptr, + nullptr); return; } // The sycl::handler uses a parallel_for kernel in the case of unusable @@ -1612,9 +1612,9 @@ void MemoryManager::ext_oneapi_prefetch_usm_cmd_buffer( std::vector Deps, ur_exp_command_buffer_sync_point_t *OutSyncPoint) { const AdapterPtr &Adapter = Context->getAdapter(); - Adapter->call( - CommandBuffer, Mem, Length, ur_usm_migration_flags_t(0), Deps.size(), - Deps.data(), OutSyncPoint); + Adapter->call(urCommandBufferAppendUSMPrefetchExp, CommandBuffer, Mem, Length, + ur_usm_migration_flags_t(0), Deps.size(), Deps.data(), 0, + nullptr, OutSyncPoint, nullptr, nullptr); } void MemoryManager::ext_oneapi_advise_usm_cmd_buffer( @@ -1624,9 +1624,9 @@ void MemoryManager::ext_oneapi_advise_usm_cmd_buffer( std::vector Deps, ur_exp_command_buffer_sync_point_t *OutSyncPoint) { const AdapterPtr &Adapter = Context->getAdapter(); - Adapter->call( - CommandBuffer, Mem, Length, Advice, Deps.size(), Deps.data(), - OutSyncPoint); + Adapter->call(urCommandBufferAppendUSMAdviseExp, CommandBuffer, Mem, Length, + Advice, Deps.size(), Deps.data(), 0, nullptr, OutSyncPoint, + nullptr, nullptr); } void MemoryManager::copy_image_bindless( diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp index 41570dd11a3c1..84c2501d0ee45 100644 --- a/sycl/source/detail/scheduler/commands.cpp +++ b/sycl/source/detail/scheduler/commands.cpp @@ -2535,12 +2535,11 @@ ur_result_t enqueueImpCommandBufferKernel( LocalSize = RequiredWGSize; } - ur_result_t Res = - Adapter->call_nocheck( - CommandBuffer, UrKernel, NDRDesc.Dims, &NDRDesc.GlobalOffset[0], - &NDRDesc.GlobalSize[0], LocalSize, 0, nullptr, SyncPoints.size(), - SyncPoints.size() ? SyncPoints.data() : nullptr, OutSyncPoint, - OutCommand); + ur_result_t Res = Adapter->call_nocheck( + urCommandBufferAppendKernelLaunchExp, CommandBuffer, UrKernel, + NDRDesc.Dims, &NDRDesc.GlobalOffset[0], &NDRDesc.GlobalSize[0], LocalSize, + SyncPoints.size(), SyncPoints.size() ? SyncPoints.data() : nullptr, 0, + nullptr, OutSyncPoint, nullptr, OutCommand); if (!SyclKernelImpl && !Kernel) { Adapter->call(UrKernel);