From edec3880c188df90700eef04a004e90c229c0c34 Mon Sep 17 00:00:00 2001 From: "Kornev, Nikita" Date: Wed, 22 Apr 2026 18:15:38 +0200 Subject: [PATCH 1/7] [SYCL][NFC] Enable IPC tests UMF issue was fixed. --- sycl/test-e2e/Experimental/ipc_memory.cpp | 3 --- sycl/test-e2e/Experimental/ipc_put_after_free.cpp | 3 --- 2 files changed, 6 deletions(-) diff --git a/sycl/test-e2e/Experimental/ipc_memory.cpp b/sycl/test-e2e/Experimental/ipc_memory.cpp index 7366935a0cba4..8221144a001df 100644 --- a/sycl/test-e2e/Experimental/ipc_memory.cpp +++ b/sycl/test-e2e/Experimental/ipc_memory.cpp @@ -1,8 +1,5 @@ // REQUIRES: aspect-usm_device_allocations && aspect-ext_oneapi_ipc_memory -// UNSUPPORTED: level_zero && windows -// UNSUPPORTED-TRACKER: UMFW-348 - // DEFINE: %{cpp20} = %if cl_options %{/clang:-std=c++20%} %else %{-std=c++20%} // RUN: %{build} -o %t.out diff --git a/sycl/test-e2e/Experimental/ipc_put_after_free.cpp b/sycl/test-e2e/Experimental/ipc_put_after_free.cpp index 2851a2589eb7f..f376e4b0314fa 100644 --- a/sycl/test-e2e/Experimental/ipc_put_after_free.cpp +++ b/sycl/test-e2e/Experimental/ipc_put_after_free.cpp @@ -1,8 +1,5 @@ // REQUIRES: aspect-usm_device_allocations && aspect-ext_oneapi_ipc_memory -// UNSUPPORTED: level_zero && windows -// UNSUPPORTED-TRACKER: UMFW-348 - // RUN: %{build} -o %t.out // RUN: %{run} %t.out From 151b5f59704243a3520928266dcc5341de177749 Mon Sep 17 00:00:00 2001 From: "Kornev, Nikita" Date: Wed, 22 Apr 2026 19:08:28 +0200 Subject: [PATCH 2/7] update-level-zero-adapter --- unified-runtime/source/adapters/level_zero/device.cpp | 5 ----- 1 file changed, 5 deletions(-) diff --git a/unified-runtime/source/adapters/level_zero/device.cpp b/unified-runtime/source/adapters/level_zero/device.cpp index ed3e9a8e88b9d..4dcf56970fd1b 100644 --- a/unified-runtime/source/adapters/level_zero/device.cpp +++ b/unified-runtime/source/adapters/level_zero/device.cpp @@ -1367,12 +1367,7 @@ ur_result_t urDeviceGetInfo( #endif } case UR_DEVICE_INFO_IPC_MEMORY_SUPPORT_EXP: -#ifdef _WIN32 - // TODO: Remove when IPC memory works in UMF on Windows. - return ReturnValue(false); -#else return ReturnValue(true); -#endif case UR_DEVICE_INFO_ASYNC_BARRIER: return ReturnValue(false); case UR_DEVICE_INFO_HOST_PIPE_READ_WRITE_SUPPORT: From a2504afed1fe4e297fc372a78a85084ebae468b7 Mon Sep 17 00:00:00 2001 From: "Kornev, Nikita" Date: Thu, 23 Apr 2026 12:03:33 +0200 Subject: [PATCH 3/7] log --- sycl/test-e2e/Experimental/ipc_memory.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test-e2e/Experimental/ipc_memory.cpp b/sycl/test-e2e/Experimental/ipc_memory.cpp index 8221144a001df..86729585bfdbc 100644 --- a/sycl/test-e2e/Experimental/ipc_memory.cpp +++ b/sycl/test-e2e/Experimental/ipc_memory.cpp @@ -3,7 +3,7 @@ // DEFINE: %{cpp20} = %if cl_options %{/clang:-std=c++20%} %else %{-std=c++20%} // RUN: %{build} -o %t.out -// RUN: %{run} %t.out +// RUN: %{run} SYCL_UR_TRACE=-1 %t.out // RUN: %{build} -DUSE_VIEW %{cpp20} -o %t.view.out // RUN: %{run} %t.view.out From ba9921f90f3e2eedd34b8ce1b568026a5c3431ad Mon Sep 17 00:00:00 2001 From: Rafal Rudnicki Date: Wed, 13 May 2026 08:06:36 +0000 Subject: [PATCH 4/7] x --- .github/workflows/sycl-windows-build.yml | 7 +- .github/workflows/sycl-windows-precommit.yml | 2 + sycl/test-e2e/Experimental/ipc_memory.cpp | 104 ++++++++++++++++-- unified-runtime/scripts/core/LEVEL_ZERO.rst | 4 + .../source/adapters/cuda/memory.cpp | 4 + .../source/adapters/level_zero/memory.cpp | 23 ++++ .../source/adapters/level_zero/v2/memory.cpp | 49 ++++++++- .../source/adapters/level_zero/v2/usm.cpp | 25 +++++ 8 files changed, 207 insertions(+), 11 deletions(-) diff --git a/.github/workflows/sycl-windows-build.yml b/.github/workflows/sycl-windows-build.yml index d2ff6df42e5c9..f49ebc5d01326 100644 --- a/.github/workflows/sycl-windows-build.yml +++ b/.github/workflows/sycl-windows-build.yml @@ -50,6 +50,11 @@ on: type: string required: false + e2e_extra_lit_opts: + type: string + required: false + default: '' + e2e_binaries_new_offload_model_artifact: type: string required: false @@ -251,7 +256,7 @@ jobs: testing_mode: build-only target_devices: all binaries_artifact: ${{ inputs.e2e_binaries_artifact }} - extra_lit_opts: --param sycl_build_targets="spir" + extra_lit_opts: --param sycl_build_targets="spir" ${{ inputs.e2e_extra_lit_opts }} cxx: ${{ inputs.cxx }} - name: Build E2E tests with New Offload Model diff --git a/.github/workflows/sycl-windows-precommit.yml b/.github/workflows/sycl-windows-precommit.yml index 4c747545df5bf..c964d13168de6 100644 --- a/.github/workflows/sycl-windows-precommit.yml +++ b/.github/workflows/sycl-windows-precommit.yml @@ -58,6 +58,7 @@ jobs: with: changes: ${{ needs.detect_changes.outputs.filters }} e2e_binaries_artifact: sycl_windows_e2ebin + e2e_extra_lit_opts: --filter=Experimental/ipc_memory\.cpp$ e2e_binaries_new_offload_model_artifact: sycl_windows_e2ebin_with_new_offload_model run_prebuilt_e2e_tests: @@ -84,6 +85,7 @@ jobs: toolchain_artifact_filename: ${{ needs.build.outputs.toolchain_artifact_filename }} testing_mode: run-only binaries_artifact: sycl_windows_e2ebin + extra_lit_opts: --filter=Experimental/ipc_memory\.cpp$ run_prebuilt_e2e_with_new_offload_model_tests: needs: build diff --git a/sycl/test-e2e/Experimental/ipc_memory.cpp b/sycl/test-e2e/Experimental/ipc_memory.cpp index 86729585bfdbc..5c874596591d0 100644 --- a/sycl/test-e2e/Experimental/ipc_memory.cpp +++ b/sycl/test-e2e/Experimental/ipc_memory.cpp @@ -3,22 +3,30 @@ // DEFINE: %{cpp20} = %if cl_options %{/clang:-std=c++20%} %else %{-std=c++20%} // RUN: %{build} -o %t.out -// RUN: %{run} SYCL_UR_TRACE=-1 %t.out +// RUN: env %if windows %{UR_L0_V2_ENABLE_WINDOWS_IPC_WA=1 %} SYCL_UR_TRACE=-1 UR_LOG_LOADER="level:debug;output:stdout;flush:debug" UR_LOG_LEVEL_ZERO="level:debug;output:stdout;flush:debug" UMF_LOG="level:debug;flush:debug;output:stdout;pid:yes" %{run} %t.out // RUN: %{build} -DUSE_VIEW %{cpp20} -o %t.view.out -// RUN: %{run} %t.view.out +// RUN: env %if windows %{UR_L0_V2_ENABLE_WINDOWS_IPC_WA=1 %} SYCL_UR_TRACE=-1 UR_LOG_LOADER="level:debug;output:stdout;flush:debug" UR_LOG_LEVEL_ZERO="level:debug;output:stdout;flush:debug" UMF_LOG="level:debug;flush:debug;output:stdout;pid:yes" %{run} %t.view.out #include #include #include +#include #include #include +#include #include +#include +#include +#include +#include #if defined(__linux__) #include #include #include +#elif defined(__WIN32__) || defined(_WIN32) +#include #endif // defined(__linux__) namespace syclexp = sycl::ext::oneapi::experimental; @@ -26,9 +34,80 @@ namespace syclexp = sycl::ext::oneapi::experimental; constexpr size_t N = 32; constexpr const char *CommsFile = "ipc_comms.txt"; -int spawner(int argc, char *argv[]) { +static void print_env(const char *Name) { + const char *Value = std::getenv(Name); + std::cout << Name << '=' << (Value ? Value : "") << std::endl; +} + +static void print_runtime_diagnostics(const char *Role, sycl::queue &Q) { + std::cout << '[' << Role << "] backend=" << static_cast(Q.get_backend()) + << " device=" << Q.get_device().get_info() + << std::endl; + print_env("UR_L0_V2_ENABLE_WINDOWS_IPC_WA"); + print_env("SYCL_UR_TRACE"); + print_env("UR_LOG_LOADER"); + print_env("UR_LOG_LEVEL_ZERO"); + print_env("UMF_LOG"); +} + +void spawn_and_sync(std::string Exe) { + std::string Cmd = '"' + Exe + '"' + " 1"; + std::cout << "Spawning: " << Cmd << std::endl; +#if defined(__WIN32__) || defined(_WIN32) + STARTUPINFO StartupInfo; + PROCESS_INFORMATION ProcInfo; + + std::memset(&ProcInfo, 0, sizeof(ProcInfo)); + std::memset(&StartupInfo, 0, sizeof(StartupInfo)); + StartupInfo.cb = sizeof(StartupInfo); + BOOL Created = + CreateProcessA(NULL, const_cast(Cmd.c_str()), NULL, NULL, TRUE, 0, + NULL, NULL, &StartupInfo, &ProcInfo); + std::cout << "CreateProcessA result: " << Created << std::endl; + if (!Created) { + std::cout << "CreateProcessA GetLastError: " << GetLastError() << std::endl; + throw std::runtime_error("CreateProcessA failed"); + } + + DWORD WaitStatus = WaitForSingleObject(ProcInfo.hProcess, 30000); + std::cout << "WaitForSingleObject result: " << WaitStatus << std::endl; + if (WaitStatus == WAIT_FAILED) { + std::cout << "WaitForSingleObject GetLastError: " << GetLastError() + << std::endl; + CloseHandle(ProcInfo.hProcess); + CloseHandle(ProcInfo.hThread); + throw std::runtime_error("WaitForSingleObject failed"); + } + + if (WaitStatus == WAIT_TIMEOUT) { + CloseHandle(ProcInfo.hProcess); + CloseHandle(ProcInfo.hThread); + throw std::runtime_error("Child process timed out"); + } + + DWORD ExitCode = 0; + if (!GetExitCodeProcess(ProcInfo.hProcess, &ExitCode)) { + std::cout << "GetExitCodeProcess GetLastError: " << GetLastError() + << std::endl; + CloseHandle(ProcInfo.hProcess); + CloseHandle(ProcInfo.hThread); + throw std::runtime_error("GetExitCodeProcess failed"); + } + std::cout << "Child exit code: " << ExitCode << std::endl; + CloseHandle(ProcInfo.hProcess); + CloseHandle(ProcInfo.hThread); + if (ExitCode != 0) + throw std::runtime_error("Child process returned non-zero exit code"); +#else + std::system(Cmd.c_str()); +#endif +} + +int spawner(int argc, char *argv[]) try { + std::cout << "Running spanwer..." << std::endl; assert(argc == 1); sycl::queue Q; + print_runtime_diagnostics("spawner", Q); #if defined(__linux__) // UMF currently requires ptrace permissions to be set for the spawner. As @@ -57,6 +136,7 @@ int spawner(int argc, char *argv[]) { syclexp::ipc_memory::handle_data_t HandleData = Handle.data(); #endif size_t HandleDataSize = HandleData.size(); + std::cout << "Spawner handle size: " << HandleDataSize << std::endl; std::fstream FS(CommsFile, std::ios_base::out | std::ios_base::binary); FS.write(reinterpret_cast(&HandleDataSize), sizeof(size_t)); FS.write(reinterpret_cast(HandleData.data()), @@ -64,9 +144,7 @@ int spawner(int argc, char *argv[]) { } // Spawn other process with an argument. - std::string Cmd = std::string{argv[0]} + " 1"; - std::cout << "Spawning: " << Cmd << std::endl; - std::system(Cmd.c_str()); + spawn_and_sync(std::string{argv[0]}); } int Failures = 0; @@ -81,15 +159,21 @@ int spawner(int argc, char *argv[]) { } sycl::free(DataPtr, Q); return Failures; +} catch (sycl::exception &e) { + std::cout << "Spawner failed: " << e.what() << std::endl; + throw; } -int consumer() { +int consumer() try { + std::cout << "Running consumer..." << std::endl; sycl::queue Q; + print_runtime_diagnostics("consumer", Q); // Read the handle data. std::fstream FS(CommsFile, std::ios_base::in | std::ios_base::binary); size_t HandleSize = 0; FS.read(reinterpret_cast(&HandleSize), sizeof(size_t)); + std::cout << "Consumer handle size: " << HandleSize << std::endl; std::unique_ptr HandleData{new std::byte[HandleSize]}; FS.read(reinterpret_cast(HandleData.get()), HandleSize); @@ -102,6 +186,8 @@ int consumer() { #endif int *DataPtr = reinterpret_cast( syclexp::ipc_memory::open(Handle, Q.get_context(), Q.get_device())); + std::cout << "Consumer open succeeded: " << static_cast(DataPtr) + << std::endl; // Test the data already in the USM pointer. int Failures = 0; @@ -121,8 +207,12 @@ int consumer() { // Close the IPC pointer. syclexp::ipc_memory::close(DataPtr, Q.get_context()); + std::cout << "Consumer close succeeded" << std::endl; return Failures; +} catch (sycl::exception &e) { + std::cout << "Consumer failed: " << e.what() << std::endl; + throw; } int main(int argc, char *argv[]) { diff --git a/unified-runtime/scripts/core/LEVEL_ZERO.rst b/unified-runtime/scripts/core/LEVEL_ZERO.rst index ad71d48cb34bf..3b05ca32a32b1 100644 --- a/unified-runtime/scripts/core/LEVEL_ZERO.rst +++ b/unified-runtime/scripts/core/LEVEL_ZERO.rst @@ -145,6 +145,10 @@ Environment Variables | UR_L0_DISABLE_USM_ALLOCATOR | Controls the use of the USM allocator. | "0": USM allocator is enabled. | "0" | | | | Any other value: USM allocator is disabled. | | +---------------------------------------------+--------------------------------------------------------------+--------------------------------------------------------------+------------------+ +| UR_L0_V2_ENABLE_WINDOWS_IPC_WA | Enables the Windows UMF IPC workaround during Level Zero v2 | "0" or unset: The workaround is disabled. | "0" | +| | provider creation. This must be enabled before provider/pool | "1": The workaround is enabled. | | +| | creation so UMF import/export IPC support is configured. | | | ++---------------------------------------------+--------------------------------------------------------------+--------------------------------------------------------------+------------------+ | UR_L0_CMD_BUFFER_USE_IMMEDIATE_APPEND_PATH | Controls which command-buffer implementation path is used. | "1": the immediate append path will always be enabled as | Unset | | | The paths rely on different APIs to enqueue command-buffers. | long as the pre-requisites are met. | | | | The immediate append path relies on | "0": the immediate append path will always be disabled. | | diff --git a/unified-runtime/source/adapters/cuda/memory.cpp b/unified-runtime/source/adapters/cuda/memory.cpp index 6eb0c8e12ad3a..7172fabcab2ae 100644 --- a/unified-runtime/source/adapters/cuda/memory.cpp +++ b/unified-runtime/source/adapters/cuda/memory.cpp @@ -10,6 +10,10 @@ #include +#ifdef _WIN32 +#include +#endif + #include "common.hpp" #include "context.hpp" #include "enqueue.hpp" diff --git a/unified-runtime/source/adapters/level_zero/memory.cpp b/unified-runtime/source/adapters/level_zero/memory.cpp index 1a31be1e57595..bd02e9a2b02d9 100644 --- a/unified-runtime/source/adapters/level_zero/memory.cpp +++ b/unified-runtime/source/adapters/level_zero/memory.cpp @@ -13,6 +13,10 @@ #include #include +#ifdef _WIN32 +#include +#endif + #include "context.hpp" #include "event.hpp" #include "helpers/memory_helpers.hpp" @@ -1952,14 +1956,31 @@ ur_result_t urEnqueueWriteHostPipe( return UR_RESULT_ERROR_UNSUPPORTED_FEATURE; } +inline ur_result_t enableWindowsUMFIPCWorkaround( + [[maybe_unused]] umf_memory_pool_handle_t umfPool) { +#ifdef _WIN32 + // UMF on Windows currently requires a workaround for IPC to work. + umf_memory_provider_handle_t umfProvider = nullptr; + UR_CALL(umf::umf2urResult(umfPoolGetMemoryProvider(umfPool, &umfProvider))); + int useImportExportForIPC = 1; + UR_CALL(umf::umf2urResult(umfCtlSet( + "umf.provider.by_handle.{}.LEVEL_ZERO.params.use_import_export_for_IPC", + &useImportExportForIPC, sizeof(useImportExportForIPC), umfProvider))); +#endif + return UR_RESULT_SUCCESS; +} + ur_result_t urIPCGetMemHandleExp(ur_context_handle_t, void *pMem, void **ppIPCMemHandleData, size_t *pIPCMemHandleDataSizeRet) { + umf_memory_pool_handle_t umfPool; auto urRet = umf::umf2urResult(umfPoolByPtr(pMem, &umfPool)); if (urRet) return urRet; + UR_CALL(enableWindowsUMFIPCWorkaround(umfPool)); + // Fast path for returning the size of the handle only. if (!ppIPCMemHandleData) return umf::umf2urResult( @@ -1989,6 +2010,8 @@ ur_result_t urIPCOpenMemHandleExp(ur_context_handle_t hContext, return UR_RESULT_ERROR_INVALID_CONTEXT; umf_memory_pool_handle_t umfPool = pool->UmfPool.get(); + UR_CALL(enableWindowsUMFIPCWorkaround(umfPool)); + size_t umfHandleSize = 0; auto urRet = umf::umf2urResult(umfPoolGetIPCHandleSize(umfPool, &umfHandleSize)); diff --git a/unified-runtime/source/adapters/level_zero/v2/memory.cpp b/unified-runtime/source/adapters/level_zero/v2/memory.cpp index e91348ad257ce..cb1a802b7df91 100644 --- a/unified-runtime/source/adapters/level_zero/v2/memory.cpp +++ b/unified-runtime/source/adapters/level_zero/v2/memory.cpp @@ -8,6 +8,10 @@ // //===----------------------------------------------------------------------===// +#ifdef _WIN32 +#include +#endif + #include "memory.hpp" #include "../ur_interface_loader.hpp" @@ -860,14 +864,44 @@ ur_result_t urMemImageGetInfo(ur_mem_handle_t /*hMemory*/, return UR_RESULT_ERROR_UNSUPPORTED_FEATURE; } +inline ur_result_t enableWindowsUMFIPCWorkaround( + [[maybe_unused]] umf_memory_pool_handle_t umfPool) { +#ifdef _WIN32 + if (!getenv_tobool("UR_L0_V2_ENABLE_WINDOWS_IPC_WA")) + return UR_RESULT_SUCCESS; + + umf_memory_provider_handle_t umfProvider = nullptr; + auto urRet = + umf::umf2urResult(umfPoolGetMemoryProvider(umfPool, &umfProvider)); + if (urRet) + return urRet; + + int useImportExportForIPC = 1; + UR_LOG(INFO, + "Applying Windows IPC workaround to provider handle {} before IPC " + "operation", + static_cast(umfProvider)); + return umf::umf2urResult(umfCtlSet( + "umf.provider.by_handle.{}.LEVEL_ZERO.params.use_import_export_for_IPC", + &useImportExportForIPC, sizeof(useImportExportForIPC), umfProvider)); +#endif + return UR_RESULT_SUCCESS; +} + ur_result_t urIPCGetMemHandleExp(ur_context_handle_t, void *pMem, void **ppIPCMemHandleData, size_t *pIPCMemHandleDataSizeRet) { + UR_LOG(INFO, "urIPCGetMemHandleExp: mem={} size_only={}", pMem, + ppIPCMemHandleData == nullptr); umf_memory_pool_handle_t umfPool; auto urRet = umf::umf2urResult(umfPoolByPtr(pMem, &umfPool)); if (urRet) return urRet; + urRet = enableWindowsUMFIPCWorkaround(umfPool); + if (urRet) + return urRet; + // Fast path for returning the size of the handle only. if (!ppIPCMemHandleData) return umf::umf2urResult( @@ -891,6 +925,8 @@ ur_result_t urIPCOpenMemHandleExp(ur_context_handle_t hContext, ur_device_handle_t hDevice, void *pIPCMemHandleData, size_t ipcMemHandleDataSize, void **ppMem) { + UR_LOG(INFO, "urIPCOpenMemHandleExp: handle={} size={}", pIPCMemHandleData, + ipcMemHandleDataSize); auto *pool = hContext->getDefaultUSMPool()->getPool( usm::pool_descriptor{hContext->getDefaultUSMPool(), hContext, hDevice, UR_USM_TYPE_DEVICE, false}); @@ -898,9 +934,12 @@ ur_result_t urIPCOpenMemHandleExp(ur_context_handle_t hContext, return UR_RESULT_ERROR_INVALID_CONTEXT; umf_memory_pool_handle_t umfPool = pool->umfPool.get(); + auto urRet = enableWindowsUMFIPCWorkaround(umfPool); + if (urRet) + return urRet; + size_t umfHandleSize = 0; - auto urRet = - umf::umf2urResult(umfPoolGetIPCHandleSize(umfPool, &umfHandleSize)); + urRet = umf::umf2urResult(umfPoolGetIPCHandleSize(umfPool, &umfHandleSize)); if (urRet) return urRet; @@ -912,12 +951,16 @@ ur_result_t urIPCOpenMemHandleExp(ur_context_handle_t hContext, if (urRet) return urRet; - return umf::umf2urResult(umfOpenIPCHandle( + urRet = umf::umf2urResult(umfOpenIPCHandle( umfIPCHandler, reinterpret_cast(pIPCMemHandleData), ppMem)); + if (!urRet) + UR_LOG(INFO, "urIPCOpenMemHandleExp succeeded: mem={}", *ppMem); + return urRet; } ur_result_t urIPCCloseMemHandleExp(ur_context_handle_t, void *pMem) { + UR_LOG(INFO, "urIPCCloseMemHandleExp: mem={}", pMem); return umf::umf2urResult(umfCloseIPCHandle(pMem)); } diff --git a/unified-runtime/source/adapters/level_zero/v2/usm.cpp b/unified-runtime/source/adapters/level_zero/v2/usm.cpp index e0f23b695cc94..3f98190636d97 100644 --- a/unified-runtime/source/adapters/level_zero/v2/usm.cpp +++ b/unified-runtime/source/adapters/level_zero/v2/usm.cpp @@ -93,8 +93,33 @@ descToDisjoinPoolMemType(const usm::pool_descriptor &desc) { UR_FFAILURE("invalid memory type: " << desc.type); } +static inline void enableWindowsUMFIPCWorkaroundAtProviderCreation() { +#ifdef _WIN32 + const bool Enabled = getenv_tobool("UR_L0_V2_ENABLE_WINDOWS_IPC_WA"); + UR_LOG(INFO, + "Windows IPC workaround at provider creation: flag={}, applying " + "before provider creation", + Enabled); + if (!Enabled) { + return; + } + + // UMF IPC on Windows must be configured before the Level Zero provider is + // created, otherwise the pool can end up with an IPC handler built without + // import/export support. + int useImportExportForIPC = 1; + UMF_CALL_THROWS(umfCtlSet( + "umf.provider.default.LEVEL_ZERO.params.use_import_export_for_IPC", + &useImportExportForIPC, sizeof(useImportExportForIPC))); + UR_LOG(INFO, + "Windows IPC workaround configured for default Level Zero provider"); +#endif +} + static umf::provider_unique_handle_t makeProvider(usm::pool_descriptor poolDescriptor) { + enableWindowsUMFIPCWorkaroundAtProviderCreation(); + umf_level_zero_memory_provider_params_handle_t hParams; UMF_CALL_THROWS(umfLevelZeroMemoryProviderParamsCreate(&hParams)); std::unique_ptr Date: Wed, 13 May 2026 08:31:56 +0000 Subject: [PATCH 5/7] [SYCL][E2E] avoid Windows lit env wrapper --- sycl/test-e2e/Experimental/ipc_memory.cpp | 28 +++++++++++++++++++++-- 1 file changed, 26 insertions(+), 2 deletions(-) diff --git a/sycl/test-e2e/Experimental/ipc_memory.cpp b/sycl/test-e2e/Experimental/ipc_memory.cpp index 5c874596591d0..c87ed4c4f27d0 100644 --- a/sycl/test-e2e/Experimental/ipc_memory.cpp +++ b/sycl/test-e2e/Experimental/ipc_memory.cpp @@ -3,9 +3,9 @@ // DEFINE: %{cpp20} = %if cl_options %{/clang:-std=c++20%} %else %{-std=c++20%} // RUN: %{build} -o %t.out -// RUN: env %if windows %{UR_L0_V2_ENABLE_WINDOWS_IPC_WA=1 %} SYCL_UR_TRACE=-1 UR_LOG_LOADER="level:debug;output:stdout;flush:debug" UR_LOG_LEVEL_ZERO="level:debug;output:stdout;flush:debug" UMF_LOG="level:debug;flush:debug;output:stdout;pid:yes" %{run} %t.out +// RUN: %{run} %t.out // RUN: %{build} -DUSE_VIEW %{cpp20} -o %t.view.out -// RUN: env %if windows %{UR_L0_V2_ENABLE_WINDOWS_IPC_WA=1 %} SYCL_UR_TRACE=-1 UR_LOG_LOADER="level:debug;output:stdout;flush:debug" UR_LOG_LEVEL_ZERO="level:debug;output:stdout;flush:debug" UMF_LOG="level:debug;flush:debug;output:stdout;pid:yes" %{run} %t.view.out +// RUN: %{run} %t.view.out #include #include @@ -39,6 +39,28 @@ static void print_env(const char *Name) { std::cout << Name << '=' << (Value ? Value : "") << std::endl; } +static void set_env(const char *Name, const char *Value) { +#if defined(__WIN32__) || defined(_WIN32) + if (!SetEnvironmentVariableA(Name, Value)) { + std::cout << "SetEnvironmentVariableA failed for " << Name + << ": " << GetLastError() << std::endl; + throw std::runtime_error("SetEnvironmentVariableA failed"); + } +#else + if (setenv(Name, Value, 1) != 0) { + throw std::runtime_error("setenv failed"); + } +#endif +} + +static void configure_runtime_diagnostics_env() { + set_env("UR_L0_V2_ENABLE_WINDOWS_IPC_WA", "1"); + set_env("SYCL_UR_TRACE", "-1"); + set_env("UR_LOG_LOADER", "level:debug;output:stdout;flush:debug"); + set_env("UR_LOG_LEVEL_ZERO", "level:debug;output:stdout;flush:debug"); + set_env("UMF_LOG", "level:debug;flush:debug;output:stdout;pid:yes"); +} + static void print_runtime_diagnostics(const char *Role, sycl::queue &Q) { std::cout << '[' << Role << "] backend=" << static_cast(Q.get_backend()) << " device=" << Q.get_device().get_info() @@ -106,6 +128,7 @@ void spawn_and_sync(std::string Exe) { int spawner(int argc, char *argv[]) try { std::cout << "Running spanwer..." << std::endl; assert(argc == 1); + configure_runtime_diagnostics_env(); sycl::queue Q; print_runtime_diagnostics("spawner", Q); @@ -166,6 +189,7 @@ int spawner(int argc, char *argv[]) try { int consumer() try { std::cout << "Running consumer..." << std::endl; + configure_runtime_diagnostics_env(); sycl::queue Q; print_runtime_diagnostics("consumer", Q); From 3809c9a86751431f229fb43d7b6f17677438466d Mon Sep 17 00:00:00 2001 From: Rafal Rudnicki Date: Wed, 13 May 2026 08:51:25 +0000 Subject: [PATCH 6/7] [CI] add Windows IPC E2E diagnostics --- .../actions/run-tests/windows/e2e/action.yml | 39 ++++++++++++++++++- 1 file changed, 38 insertions(+), 1 deletion(-) diff --git a/devops/actions/run-tests/windows/e2e/action.yml b/devops/actions/run-tests/windows/e2e/action.yml index f4b9e54b67ab3..95c199c5e8341 100644 --- a/devops/actions/run-tests/windows/e2e/action.yml +++ b/devops/actions/run-tests/windows/e2e/action.yml @@ -76,7 +76,7 @@ runs: continue-on-error: true shell: bash env: - LIT_OPTS: -v --no-progress-bar --show-unsupported --show-pass --show-xfail --max-time ${{ inputs.testing_mode == 'run-only' && 1200 || 3600 }} --time-tests --param print_features=True --param test-mode=${{ inputs.testing_mode }} --param sycl_devices=${{ inputs.target_devices }} ${{ inputs.extra_lit_opts }} + LIT_OPTS: -vv --no-progress-bar --show-unsupported --show-pass --show-xfail --max-time ${{ inputs.testing_mode == 'run-only' && 1200 || 3600 }} --time-tests --param print_features=True --param test-mode=${{ inputs.testing_mode }} --param sycl_devices=${{ inputs.target_devices }} ${{ inputs.extra_lit_opts }} run: | cmake --build build-e2e --target check-sycl-e2e > e2e.log 2>&1 # Two steps below are duplicated between Lin/Win actions, updates must change both @@ -87,6 +87,43 @@ runs: echo "::group::Show Full E2E Log" cat e2e.log echo "::endgroup::" + - name: E2E failure diagnostics + if: steps.run_e2e.outcome != 'success' + shell: bash + run: | + echo "::group::IPC Failure Diagnostics" + echo "working-directory=$(pwd)" + if [ -f build-e2e/CMakeCache.txt ]; then + echo "::group::Relevant CMakeCache entries" + rg -n "CMAKE_CXX_COMPILER:|LLVM_LIT:|SYCL_E2E_TESTS_LIT_FLAGS:|CMAKE_GENERATOR:|LEVEL_ZERO_" build-e2e/CMakeCache.txt || true + echo "::endgroup::" + fi + if [ -f build-e2e/build.ninja ]; then + echo "::group::check-sycl-e2e target excerpt" + rg -n -C 3 "check-sycl-e2e|ipc_memory\.cpp" build-e2e/build.ninja || true + echo "::endgroup::" + fi + if [ -d build-e2e/Experimental/Output ]; then + echo "::group::Experimental Output listing" + find build-e2e/Experimental/Output -maxdepth 1 -type f | sort || true + echo "::endgroup::" + for path in build-e2e/Experimental/Output/ipc_memory.cpp*; do + [ -e "$path" ] || continue + echo "::group::Artifact $path" + ls -l "$path" + echo "::endgroup::" + done + for exe in build-e2e/Experimental/Output/ipc_memory.cpp*.out; do + [ -f "$exe" ] || continue + echo "::group::Direct rerun $exe" + "$(realpath "$exe")" 2>&1 || true + echo "::endgroup::" + done + fi + echo "::group::ipc_memory excerpt from e2e.log" + rg -n -C 8 "ipc_memory\.cpp|FAILED:|executed command|command stdout|command stderr|error:" e2e.log || true + echo "::endgroup::" + echo "::endgroup::" - name: Report E2E Failures if: steps.run_e2e.outcome != 'success' shell: bash From a446c4bfdd15babcafa7b8abeca0392c6c1792a6 Mon Sep 17 00:00:00 2001 From: Rafal Rudnicki Date: Wed, 13 May 2026 10:10:20 +0000 Subject: [PATCH 7/7] [SYCL] Fix Windows ipc_memory E2E run path --- devops/actions/run-tests/windows/e2e/action.yml | 6 +++--- sycl/test-e2e/Experimental/ipc_memory.cpp | 10 ++++++---- 2 files changed, 9 insertions(+), 7 deletions(-) diff --git a/devops/actions/run-tests/windows/e2e/action.yml b/devops/actions/run-tests/windows/e2e/action.yml index 95c199c5e8341..3d028745189c4 100644 --- a/devops/actions/run-tests/windows/e2e/action.yml +++ b/devops/actions/run-tests/windows/e2e/action.yml @@ -95,12 +95,12 @@ runs: echo "working-directory=$(pwd)" if [ -f build-e2e/CMakeCache.txt ]; then echo "::group::Relevant CMakeCache entries" - rg -n "CMAKE_CXX_COMPILER:|LLVM_LIT:|SYCL_E2E_TESTS_LIT_FLAGS:|CMAKE_GENERATOR:|LEVEL_ZERO_" build-e2e/CMakeCache.txt || true + grep -nE "CMAKE_CXX_COMPILER:|LLVM_LIT:|SYCL_E2E_TESTS_LIT_FLAGS:|CMAKE_GENERATOR:|LEVEL_ZERO_" build-e2e/CMakeCache.txt || true echo "::endgroup::" fi if [ -f build-e2e/build.ninja ]; then echo "::group::check-sycl-e2e target excerpt" - rg -n -C 3 "check-sycl-e2e|ipc_memory\.cpp" build-e2e/build.ninja || true + grep -nE "check-sycl-e2e|ipc_memory\.cpp" build-e2e/build.ninja || true echo "::endgroup::" fi if [ -d build-e2e/Experimental/Output ]; then @@ -121,7 +121,7 @@ runs: done fi echo "::group::ipc_memory excerpt from e2e.log" - rg -n -C 8 "ipc_memory\.cpp|FAILED:|executed command|command stdout|command stderr|error:" e2e.log || true + grep -nE "ipc_memory\.cpp|FAILED:|executed command|command stdout|command stderr|error:" e2e.log || true echo "::endgroup::" echo "::endgroup::" - name: Report E2E Failures diff --git a/sycl/test-e2e/Experimental/ipc_memory.cpp b/sycl/test-e2e/Experimental/ipc_memory.cpp index c87ed4c4f27d0..afef6bdd23862 100644 --- a/sycl/test-e2e/Experimental/ipc_memory.cpp +++ b/sycl/test-e2e/Experimental/ipc_memory.cpp @@ -3,9 +3,9 @@ // DEFINE: %{cpp20} = %if cl_options %{/clang:-std=c++20%} %else %{-std=c++20%} // RUN: %{build} -o %t.out -// RUN: %{run} %t.out +// RUN: %{run-unfiltered-devices} %t.out // RUN: %{build} -DUSE_VIEW %{cpp20} -o %t.view.out -// RUN: %{run} %t.view.out +// RUN: %{run-unfiltered-devices} %t.view.out #include #include @@ -42,8 +42,8 @@ static void print_env(const char *Name) { static void set_env(const char *Name, const char *Value) { #if defined(__WIN32__) || defined(_WIN32) if (!SetEnvironmentVariableA(Name, Value)) { - std::cout << "SetEnvironmentVariableA failed for " << Name - << ": " << GetLastError() << std::endl; + std::cout << "SetEnvironmentVariableA failed for " << Name << ": " + << GetLastError() << std::endl; throw std::runtime_error("SetEnvironmentVariableA failed"); } #else @@ -54,6 +54,7 @@ static void set_env(const char *Name, const char *Value) { } static void configure_runtime_diagnostics_env() { + set_env("ONEAPI_DEVICE_SELECTOR", "level_zero:gpu"); set_env("UR_L0_V2_ENABLE_WINDOWS_IPC_WA", "1"); set_env("SYCL_UR_TRACE", "-1"); set_env("UR_LOG_LOADER", "level:debug;output:stdout;flush:debug"); @@ -65,6 +66,7 @@ static void print_runtime_diagnostics(const char *Role, sycl::queue &Q) { std::cout << '[' << Role << "] backend=" << static_cast(Q.get_backend()) << " device=" << Q.get_device().get_info() << std::endl; + print_env("ONEAPI_DEVICE_SELECTOR"); print_env("UR_L0_V2_ENABLE_WINDOWS_IPC_WA"); print_env("SYCL_UR_TRACE"); print_env("UR_LOG_LOADER");