[DEBUG] Try to run Proton tests on windows

.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

.github/workflows/build-test.yml

@@ -44,10 +44,6 @@ on:
         type: boolean
         default: false
 
-  pull_request:
-    branches:
-      - main
-      - release/**
   push:
     branches:
       - main

.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

.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

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__':

third_party/intel/backend/proton_utils.cpp

@@ -2,14 +2,20 @@
 #include <level_zero/ze_api.h>
 #include <sycl/sycl.hpp>
 
-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<sycl::queue *>(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<std::vector<std::array<uint8_t, 16>> *>(deviceUUIDsPtr);
   if (!deviceUUIDs_->empty()) {
@@ -57,10 +63,10 @@ void check(ze_result_t ret, const char *functionName) {
 // FIXME: rewrite with
 // sycl::device.get_info<sycl::ext::intel::info::device::architecture>; 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

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<void *>(sycl_queue);
         auto sessionId = SessionManager::instance().addSession(

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;

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-0.dll";
+#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;

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;
   }

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-0.dll'))):  # pylint: disable=W0640
                         os.environ["TRITON_XPUPTI_LIB_PATH"] = str(pathlib.Path(f.locate()).parent.resolve())
                         break
         except importlib.metadata.PackageNotFoundError: