From 31637b1edec9c1600021f0389e6f6bb3fb5c22e4 Mon Sep 17 00:00:00 2001 From: Anatoly Myachev Date: Sun, 23 Nov 2025 15:44:34 +0000 Subject: [PATCH 1/6] [DEBUG] Try to run Proton tests on windows Signed-off-by: Anatoly Myachev --- .github/workflows/build-test-windows.yml | 7 +++++++ .github/workflows/build-test.yml | 4 ---- .github/workflows/pip-test-windows.yml | 7 +++++++ .github/workflows/pip-test.yml | 5 +++++ 4 files changed, 19 insertions(+), 4 deletions(-) diff --git a/.github/workflows/build-test-windows.yml b/.github/workflows/build-test-windows.yml index beadb7f8a4..b873224658 100644 --- a/.github/workflows/build-test-windows.yml +++ b/.github/workflows/build-test-windows.yml @@ -152,6 +152,13 @@ jobs: cd ${{ env.NEW_WORKSPACE }} ${{ env.TRITON_TEST_CMD }} --unit + - name: Run Proton tests + run: | + .venv\Scripts\activate.ps1 + Invoke-BatchFile "C:\Program Files (x86)\Intel\oneAPI\setvars.bat" + cd ${{ env.NEW_WORKSPACE }}\third_party\proton\test + pytest test_api.py test_cmd.py test_lib.py test_profile.py test_viewer.py --device xpu -s -v + - name: Run core tests run: | .venv\Scripts\activate.ps1 diff --git a/.github/workflows/build-test.yml b/.github/workflows/build-test.yml index ca75d54fdd..ebab991b6a 100644 --- a/.github/workflows/build-test.yml +++ b/.github/workflows/build-test.yml @@ -44,10 +44,6 @@ on: type: boolean default: false - pull_request: - branches: - - main - - release/** push: branches: - main diff --git a/.github/workflows/pip-test-windows.yml b/.github/workflows/pip-test-windows.yml index 5d1a0d4974..a1acb82408 100644 --- a/.github/workflows/pip-test-windows.yml +++ b/.github/workflows/pip-test-windows.yml @@ -129,6 +129,13 @@ jobs: cd ${{ env.NEW_WORKSPACE }} ${{ env.TRITON_TEST_CMD }} --unit + - name: Run Proton tests + run: | + .venv\Scripts\activate.ps1 + Invoke-BatchFile "C:\Program Files\Microsoft Visual Studio\2022\Community\VC\Auxiliary\Build\vcvarsall.bat" x64 + cd ${{ env.NEW_WORKSPACE }}\third_party\proton\test + pytest test_api.py test_cmd.py test_lib.py test_profile.py test_viewer.py --device xpu -s -v + - name: Run core tests run: | .venv\Scripts\activate.ps1 diff --git a/.github/workflows/pip-test.yml b/.github/workflows/pip-test.yml index 9f6f5b1d5a..2b8babc6c8 100644 --- a/.github/workflows/pip-test.yml +++ b/.github/workflows/pip-test.yml @@ -70,6 +70,11 @@ jobs: pip install -r /tmp/requirements.txt pip install transformers==4.54.0 + - name: Run Proton tests + run: | + cd third_party/proton/test + pytest test_api.py test_cmd.py test_lib.py test_profile.py test_viewer.py --device xpu -s -v + - name: Run core tests run: | ${{ env.TRITON_TEST_CMD }} --core From 89201943e37225f8a3c473523c86c0db86f9a61a Mon Sep 17 00:00:00 2001 From: Anatoly Myachev Date: Wed, 26 Nov 2025 18:08:21 +0000 Subject: [PATCH 2/6] change type: long -> uint64_t Signed-off-by: Anatoly Myachev --- third_party/proton/csrc/Proton.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/third_party/proton/csrc/Proton.cpp b/third_party/proton/csrc/Proton.cpp index 811f794394..1acfcf49c5 100644 --- a/third_party/proton/csrc/Proton.cpp +++ b/third_party/proton/csrc/Proton.cpp @@ -18,7 +18,7 @@ static void initProton(pybind11::module &&m) { "start", [](const std::string &path, const std::string &contextSourceName, const std::string &dataName, const std::string &profilerName, - const std::string &mode, long sycl_queue, + const std::string &mode, uint64_t sycl_queue, const std::string &utils_cache_path) { void *queue = reinterpret_cast(sycl_queue); auto sessionId = SessionManager::instance().addSession( From bf7fd8a99197ccfba7b42bed689a11f9dbbed678 Mon Sep 17 00:00:00 2001 From: Anatoly Myachev Date: Wed, 26 Nov 2025 18:43:22 +0000 Subject: [PATCH 3/6] add pti_view.lib Signed-off-by: Anatoly Myachev --- third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp | 4 ++++ third_party/proton/proton/profile.py | 2 +- 2 files changed, 5 insertions(+), 1 deletion(-) diff --git a/third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp b/third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp index c618661372..fbe6633c86 100644 --- a/third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp +++ b/third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp @@ -8,7 +8,11 @@ namespace xpupti { struct ExternLibXpupti : public ExternLibBase { using RetType = pti_result; +#if defined(_WIN32) + static constexpr const char *name = "pti_view.lib"; +#else static constexpr const char *name = "libpti_view.so"; +#endif static constexpr const char *defaultDir = ""; static constexpr const char *pathEnv = "TRITON_XPUPTI_LIB_PATH"; static constexpr RetType success = PTI_SUCCESS; diff --git a/third_party/proton/proton/profile.py b/third_party/proton/proton/profile.py index 1bada6b324..6a4f4b0fa2 100644 --- a/third_party/proton/proton/profile.py +++ b/third_party/proton/proton/profile.py @@ -27,7 +27,7 @@ def _select_backend() -> str: try: if (files := importlib.metadata.files('intel-pti')) is not None: for f in files: - if 'libpti_view.so' in f.name: + if any(map(lambda el: el in f.name, ('libpti_view.so', 'pti_view.lib'))): # pylint: disable=W0640 os.environ["TRITON_XPUPTI_LIB_PATH"] = str(pathlib.Path(f.locate()).parent.resolve()) break except importlib.metadata.PackageNotFoundError: From 350f7886e4900cd403fe8f559c2afec539e261e4 Mon Sep 17 00:00:00 2001 From: Anatoly Myachev Date: Thu, 27 Nov 2025 13:15:43 +0000 Subject: [PATCH 4/6] use 'pti_view-0.dll' Signed-off-by: Anatoly Myachev --- scripts/pti_lib.py | 6 +++--- third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp | 2 +- third_party/proton/proton/profile.py | 2 +- 3 files changed, 5 insertions(+), 5 deletions(-) diff --git a/scripts/pti_lib.py b/scripts/pti_lib.py index 8ce4343791..7c4fad90a1 100644 --- a/scripts/pti_lib.py +++ b/scripts/pti_lib.py @@ -9,13 +9,13 @@ def get_pti_lib_path() -> pathlib.Path: Raises: importlib.metadata.PackageNotFoundError: if 'intel-pti' not installed. - AssertionError: if libpti_view.so not found. + AssertionError: if libpti_view.so/pti_view-0.dll not found. """ files = importlib.metadata.files('intel-pti') or [] for f in files: - if any(map(lambda el: el in f.name, ('libpti_view.so', 'pti_view.lib'))): # pylint: disable=W0640 + if any(map(lambda el: el in f.name, ('libpti_view.so', 'pti_view-0.dll'))): # pylint: disable=W0640 return pathlib.Path(f.locate()).parent.resolve() - raise AssertionError('libpti_view.so not found') + raise AssertionError('libpti_view.so/pti_view-0.dll not found') if __name__ == '__main__': diff --git a/third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp b/third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp index fbe6633c86..04da9e65f2 100644 --- a/third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp +++ b/third_party/proton/csrc/lib/Driver/GPU/XpuptiApi.cpp @@ -9,7 +9,7 @@ namespace xpupti { struct ExternLibXpupti : public ExternLibBase { using RetType = pti_result; #if defined(_WIN32) - static constexpr const char *name = "pti_view.lib"; + static constexpr const char *name = "pti_view-0.dll"; #else static constexpr const char *name = "libpti_view.so"; #endif diff --git a/third_party/proton/proton/profile.py b/third_party/proton/proton/profile.py index 6a4f4b0fa2..51cf9377e0 100644 --- a/third_party/proton/proton/profile.py +++ b/third_party/proton/proton/profile.py @@ -27,7 +27,7 @@ def _select_backend() -> str: try: if (files := importlib.metadata.files('intel-pti')) is not None: for f in files: - if any(map(lambda el: el in f.name, ('libpti_view.so', 'pti_view.lib'))): # pylint: disable=W0640 + if any(map(lambda el: el in f.name, ('libpti_view.so', 'pti_view-0.dll'))): # pylint: disable=W0640 os.environ["TRITON_XPUPTI_LIB_PATH"] = str(pathlib.Path(f.locate()).parent.resolve()) break except importlib.metadata.PackageNotFoundError: From ecd6930c6fcf7a9495c867c8b8739e126012a5e7 Mon Sep 17 00:00:00 2001 From: Anatoly Myachev Date: Thu, 27 Nov 2025 16:15:55 +0100 Subject: [PATCH 5/6] improve error messages Signed-off-by: Anatoly Myachev --- .../proton/csrc/lib/Driver/GPU/XpuApi.cpp | 20 +++++++++------- .../lib/Profiler/Xpupti/XpuptiProfiler.cpp | 24 ++++++++++++------- 2 files changed, 28 insertions(+), 16 deletions(-) diff --git a/third_party/proton/csrc/lib/Driver/GPU/XpuApi.cpp b/third_party/proton/csrc/lib/Driver/GPU/XpuApi.cpp index 35ba093b7d..7fa0275381 100644 --- a/third_party/proton/csrc/lib/Driver/GPU/XpuApi.cpp +++ b/third_party/proton/csrc/lib/Driver/GPU/XpuApi.cpp @@ -24,8 +24,9 @@ Device getDevice(uint64_t index) { HMODULE handle = LoadLibrary(PROTON_UTILS.data()); if (!handle) { long err = GetLastError(); - throw std::runtime_error(std::string("Failed to load library code:") + - std::to_string(err)); + throw std::runtime_error( + std::string("Failed to load 'PROTON_UTILS' library:") + + std::to_string(err)); } GetLastError(); @@ -34,8 +35,9 @@ Device getDevice(uint64_t index) { long err = GetLastError(); if (err) { FreeLibrary(handle); - throw std::runtime_error(std::string("Failed to load function code:") + - std::to_string(err)); + throw std::runtime_error( + std::string("Failed to load 'getDeviceProperties' function") + + std::to_string(err)); } uint32_t clockRate = 0; @@ -55,8 +57,9 @@ Device getDevice(uint64_t index) { void *handle = dlopen(PROTON_UTILS.data(), RTLD_LAZY); if (!handle) { const char *dlopen_error = dlerror(); - throw std::runtime_error(std::string("Failed to load library: ") + - std::string(dlopen_error)); + throw std::runtime_error( + std::string("Failed to load 'PROTON_UTILS' library: ") + + std::string(dlopen_error)); } dlerror(); @@ -65,8 +68,9 @@ Device getDevice(uint64_t index) { const char *dlsym_error = dlerror(); if (dlsym_error) { dlclose(handle); - throw std::runtime_error(std::string("Failed to load function: ") + - std::string(dlsym_error)); + throw std::runtime_error( + std::string("Failed to load 'getDeviceProperties' function: ") + + std::string(dlsym_error)); } uint32_t clockRate = 0; diff --git a/third_party/proton/csrc/lib/Profiler/Xpupti/XpuptiProfiler.cpp b/third_party/proton/csrc/lib/Profiler/Xpupti/XpuptiProfiler.cpp index 341d735494..d4ac0dd7ab 100644 --- a/third_party/proton/csrc/lib/Profiler/Xpupti/XpuptiProfiler.cpp +++ b/third_party/proton/csrc/lib/Profiler/Xpupti/XpuptiProfiler.cpp @@ -273,7 +273,8 @@ typedef void (*EnumDeviceUUIDsFunc)(void *); int callEnumDeviceUUIDs(const std::string &utils_cache_path) { HMODULE handle = LoadLibrary(xpu::PROTON_UTILS.data()); if (!handle) { - std::cerr << "Failed to load library: " << GetLastError() << std::endl; + std::cerr << "Failed to load 'PROTON_UTILS' library: " << GetLastError() + << std::endl; return 1; } @@ -282,7 +283,8 @@ int callEnumDeviceUUIDs(const std::string &utils_cache_path) { (EnumDeviceUUIDsFunc)GetProcAddress(handle, "enumDeviceUUIDs"); long dlsym_error = GetLastError(); if (dlsym_error) { - std::cerr << "Failed to load function: " << dlsym_error << std::endl; + std::cerr << "Failed to load 'enumDeviceUUIDs' function: " << dlsym_error + << std::endl; FreeLibrary(handle); return 1; } @@ -296,7 +298,8 @@ int callEnumDeviceUUIDs(const std::string &utils_cache_path) { int callEnumDeviceUUIDs(const std::string &utils_cache_path) { void *handle = dlopen(xpu::PROTON_UTILS.data(), RTLD_LAZY); if (!handle) { - std::cerr << "Failed to load library: " << dlerror() << std::endl; + std::cerr << "Failed to load 'PROTON_UTILS' library: " << dlerror() + << std::endl; return 1; } @@ -305,7 +308,8 @@ int callEnumDeviceUUIDs(const std::string &utils_cache_path) { (EnumDeviceUUIDsFunc)dlsym(handle, "enumDeviceUUIDs"); const char *dlsym_error = dlerror(); if (dlsym_error) { - std::cerr << "Failed to load function: " << dlsym_error << std::endl; + std::cerr << "Failed to load 'enumDeviceUUIDs' function: " << dlsym_error + << std::endl; dlclose(handle); return 1; } @@ -323,7 +327,8 @@ typedef void (*WaitOnSyclQueueFunc)(void *); int callWaitOnSyclQueue(void *syclQueue) { HMODULE handle = LoadLibrary(xpu::PROTON_UTILS.data()); if (!handle) { - std::cerr << "Failed to load library: " << GetLastError() << std::endl; + std::cerr << "Failed to load 'PROTON_UTILS' library: " << GetLastError() + << std::endl; return 1; } @@ -332,7 +337,8 @@ int callWaitOnSyclQueue(void *syclQueue) { (WaitOnSyclQueueFunc)GetProcAddress(handle, "waitOnSyclQueue"); long dlsym_error = GetLastError(); if (dlsym_error) { - std::cerr << "Failed to load function: " << dlsym_error << std::endl; + std::cerr << "Failed to load 'waitOnSyclQueue' function: " << dlsym_error + << std::endl; FreeLibrary(handle); return 1; } @@ -346,7 +352,8 @@ int callWaitOnSyclQueue(void *syclQueue) { int callWaitOnSyclQueue(void *syclQueue) { void *handle = dlopen(xpu::PROTON_UTILS.data(), RTLD_LAZY); if (!handle) { - std::cerr << "Failed to load library: " << dlerror() << std::endl; + std::cerr << "Failed to load 'PROTON_UTILS' library: " << dlerror() + << std::endl; return 1; } @@ -355,7 +362,8 @@ int callWaitOnSyclQueue(void *syclQueue) { (WaitOnSyclQueueFunc)dlsym(handle, "waitOnSyclQueue"); const char *dlsym_error = dlerror(); if (dlsym_error) { - std::cerr << "Failed to load function: " << dlsym_error << std::endl; + std::cerr << "Failed to load 'waitOnSyclQueue' function: " << dlsym_error + << std::endl; dlclose(handle); return 1; } From fac73a1a52b66a6d8630055cc32dad89f79622bf Mon Sep 17 00:00:00 2001 From: Anatoly Myachev Date: Thu, 27 Nov 2025 17:13:06 +0100 Subject: [PATCH 6/6] add 'EXPORT_FUNC' macros Signed-off-by: Anatoly Myachev --- third_party/intel/backend/proton_utils.cpp | 18 ++++++++++++------ 1 file changed, 12 insertions(+), 6 deletions(-) diff --git a/third_party/intel/backend/proton_utils.cpp b/third_party/intel/backend/proton_utils.cpp index 165b6b1667..a3a8740f41 100644 --- a/third_party/intel/backend/proton_utils.cpp +++ b/third_party/intel/backend/proton_utils.cpp @@ -2,14 +2,20 @@ #include #include -extern "C" void waitOnSyclQueue(void *syclQueue) { +#if defined(_WIN32) +#define EXPORT_FUNC __declspec(dllexport) +#else +#define EXPORT_FUNC __attribute__((visibility("default"))) +#endif + +extern "C" EXPORT_FUNC void waitOnSyclQueue(void *syclQueue) { sycl::queue *queue = static_cast(syclQueue); queue->wait(); } // FIXME: Should it be in DeviceInfo class? // Inspired by Kineto: `XpuptiActivityProfiler.cpp` -extern "C" void enumDeviceUUIDs(void *deviceUUIDsPtr) { +extern "C" EXPORT_FUNC void enumDeviceUUIDs(void *deviceUUIDsPtr) { auto *deviceUUIDs_ = reinterpret_cast> *>(deviceUUIDsPtr); if (!deviceUUIDs_->empty()) { @@ -57,10 +63,10 @@ void check(ze_result_t ret, const char *functionName) { // FIXME: rewrite with // sycl::device.get_info; cache // the result -extern "C" void getDeviceProperties(uint64_t index, uint32_t *clockRate, - uint32_t *memoryClockRate, - uint32_t *busWidth, uint32_t *numSms, - char arch[256]) { +extern "C" EXPORT_FUNC void +getDeviceProperties(uint64_t index, uint32_t *clockRate, + uint32_t *memoryClockRate, uint32_t *busWidth, + uint32_t *numSms, char arch[256]) { // ref: getDeviceProperties // FIXME: double check that initialization is needed