[Feature] Add full Windows support and fix related cross-platform issues (#2093)

* [Feature] Add Windows support with cross-platform minimal diff

* [Feature] Add ccache support for Windows

* [Feature] Add parallel autotune support for Windows

* [Feature] Add nvrtc support for Windows

* [Bugfix] Fix tilelibrary support for Windows

* [Bugfix] Fix alignment for Windows

* [Feature] Add clang-cl support for Windows

* [Bugfix] Fix numeric limit for Windows

* [Bugfix] Fix rand for Windows

* [Misc] Use dynamic test workload to avoid hang

* [Bugfix] Fix TMA alignment for Windows and sm120

* [Bugfix] Add unix-like hex formatter for Windows

* [CI] Add Windows wheel job

* [Feature] Flexible CUDA loader

* fix rebase

---------

Co-authored-by: SiriusNEO <chaofan@deepseek.com>

Author: sepcnt

.github/workflows/dist.yml

@@ -81,7 +81,7 @@ jobs:
         uses: hendrikmuhs/ccache-action@v1
         with:
           max-size: "200MB"
-          create-symlink: true
+          create-symlink: ${{ runner.os != 'Windows' }}
           evict-old-files: "7d"
           append-timestamp: false
           key: sdist-${{ runner.os }}-${{ runner.arch }}-${{ hashFiles('**/*.cc') }}
@@ -108,19 +108,19 @@ jobs:
           if-no-files-found: error
 
   build-wheels:
-    name: Build wheels for Python ${{ matrix.python-version }} on ${{ matrix.target.runner }} with ${{ matrix.target.toolkit }}
+    name: Build wheels for Python ${{ matrix.target.python_version }} on ${{ matrix.target.runner }} with ${{ matrix.target.toolkit }}
     if: |
       github.repository_owner == 'tile-ai' &&
       (github.event_name != 'pull_request' || !github.event.pull_request.draft)
     strategy:
       matrix:
         target:
-          - { runner: ubuntu-latest, toolkit: "CUDA-12.8", test_backends: "cu118 cu130" }
-          - { runner: ubuntu-24.04-arm, toolkit: "CUDA-12.8", test_backends: "cu126 cu130" }
-          - { runner: macos-latest, toolkit: "Metal" }
-        python-version:
-          # Build wheels for different Python ABIs
-          - "3.9"
+          # Build wheels for different Python ABIs.
+          # Windows CUDA 13.0 uses cp310 because PyTorch cu130 does not publish cp39 wheels.
+          - { runner: ubuntu-latest, toolkit: "CUDA-12.8", test_backends: "cu118 cu130", python_version: "3.9" }
+          - { runner: ubuntu-24.04-arm, toolkit: "CUDA-12.8", test_backends: "cu126 cu130", python_version: "3.9" }
+          - { runner: windows-latest, toolkit: "CUDA-13.0", test_backends: "cu130", python_version: "3.10" }
+          - { runner: macos-latest, toolkit: "Metal", python_version: "3.9" }
           # - "3.14t"  # let user to build from source for now
           # TODO: Add cp315-abi3.abi3t after PEP 803
       fail-fast: false
@@ -131,6 +131,11 @@ jobs:
       NO_VERSION_LABEL: "OFF"
 
     steps:
+      - name: Enable git long paths
+        if: runner.os == 'Windows'
+        shell: pwsh
+        run: git config --global core.longpaths true
+
       - name: Checkout repository
         uses: actions/checkout@v6
         with:
@@ -141,18 +146,19 @@ jobs:
         uses: hendrikmuhs/ccache-action@v1
         with:
           max-size: "200MB"
-          create-symlink: true
+          create-symlink: ${{ runner.os != 'Windows' }}
           evict-old-files: "7d"
           append-timestamp: false
-          key: wheel-${{ runner.os }}-${{ runner.arch }}-${{ matrix.python-version }}-${{ hashFiles('**/*.cc') }}
+          key: wheel-${{ runner.os }}-${{ runner.arch }}-${{ matrix.target.python_version }}-${{ hashFiles('**/*.cc') }}
           restore-keys: |
-            wheel-${{ runner.os }}-${{ runner.arch }}-${{ matrix.python-version }}-${{ hashFiles('**/*.cc') }}
-            wheel-${{ runner.os }}-${{ runner.arch }}-${{ matrix.python-version }}
+            wheel-${{ runner.os }}-${{ runner.arch }}-${{ matrix.target.python_version }}-${{ hashFiles('**/*.cc') }}
+            wheel-${{ runner.os }}-${{ runner.arch }}-${{ matrix.target.python_version }}
             wheel-${{ runner.os }}-${{ runner.arch }}
 
       - name: Set CIBW_BUILD
+        if: runner.os != 'Windows'
         run: |
-          PYTHON_VERSION="${{ matrix.python-version }}"
+          PYTHON_VERSION="${{ matrix.target.python_version }}"
           PYTHON_VERSION_MAJMIN="$(echo "${PYTHON_VERSION}" | cut -d '.' -f-2)"
           PYTHON_VERSION_MAJMIN_NODOT="${PYTHON_VERSION_MAJMIN//./}"
           echo "CIBW_BUILD=cp${PYTHON_VERSION_MAJMIN_NODOT}-*" | tee -a "${GITHUB_ENV}"
@@ -175,6 +181,30 @@ jobs:
             echo "CIBW_BEFORE_BUILD_LINUX=dnf install -y ccache && ccache -o cache_dir=/host${HOST_CCACHE_DIR}" | tee -a "${GITHUB_ENV}"
           fi
 
+      - name: Set CIBW_BUILD (Windows)
+        if: runner.os == 'Windows'
+        shell: pwsh
+        run: |
+          $pythonVersion = "${{ matrix.target.python_version }}"
+          $pythonVersionMajMin = ($pythonVersion -split '\.')[0..1] -join '.'
+          $pythonVersionMajMinNoDot = $pythonVersionMajMin -replace '\.', ''
+          "CIBW_BUILD=cp${pythonVersionMajMinNoDot}-*" | Tee-Object -FilePath $env:GITHUB_ENV -Append
+          "CIBW_CONFIG_SETTINGS=cmake.build-type=Release" | Tee-Object -FilePath $env:GITHUB_ENV -Append
+          "CMAKE_BUILD_TYPE=Release" | Tee-Object -FilePath $env:GITHUB_ENV -Append
+
+          if ("${{ matrix.target.toolkit }}" -like "*CUDA*") {
+            $cudaVersion = "${{ matrix.target.toolkit }}".Split('-')[-1]
+            $cudaVersionMajMin = ($cudaVersion -split '\.')[0..1] -join '.'
+            $cudaVersionMajMinNoDot = $cudaVersionMajMin -replace '\.', ''
+            "CUDA_VERSION=$cudaVersion" | Tee-Object -FilePath $env:GITHUB_ENV -Append
+            "UV_TORCH_BACKEND=cu$cudaVersionMajMinNoDot" | Tee-Object -FilePath $env:GITHUB_ENV -Append
+            "PIP_EXTRA_INDEX_URL=https://download.pytorch.org/whl/cu$cudaVersionMajMinNoDot" | Tee-Object -FilePath $env:GITHUB_ENV -Append
+          }
+
+          if ("${{ env.IS_RELEASE }}" -eq "true") {
+            "NO_VERSION_LABEL=ON" | Tee-Object -FilePath $env:GITHUB_ENV -Append
+          }
+
       - name: Build wheels
         uses: pypa/cibuildwheel@v3.4
         with:
@@ -186,10 +216,11 @@ jobs:
         id: setup-uv
         uses: astral-sh/setup-uv@v7
         with:
-          python-version: ${{ matrix.python-version }}
+          python-version: ${{ matrix.target.python_version }}
           activate-environment: true
 
       - name: Test built wheels
+        if: runner.os != 'Windows'
         run: |
           for WHEEL in wheelhouse/*.whl; do
             echo "Testing wheel: ${WHEEL}"
@@ -212,8 +243,34 @@ jobs:
             )
           done
 
+      - name: Test built wheels (Windows)
+        if: runner.os == 'Windows'
+        shell: pwsh
+        run: |
+          Get-ChildItem wheelhouse -Filter *.whl | ForEach-Object {
+            $wheel = $_.FullName
+            Write-Host "Testing wheel: $wheel"
+            if ($_.Name -like "*abi3*") {
+              uvx abi3audit --verbose --strict $wheel
+            }
+            if (Test-Path test-venv) {
+              Remove-Item -Recurse -Force test-venv
+            }
+            uv venv test-venv
+            $python = Join-Path (Resolve-Path test-venv).Path "Scripts\python.exe"
+
+            uv pip install -v --python $python $wheel
+            Push-Location $env:RUNNER_TEMP
+            try {
+              & $python -c "import tilelang; print(tilelang.__version__)"
+            } finally {
+              Pop-Location
+            }
+            Remove-Item -Recurse -Force test-venv
+          }
+
       - name: Test built wheels with different CUDA
-        if: contains(matrix.target.toolkit, 'CUDA')
+        if: contains(matrix.target.toolkit, 'CUDA') && runner.os != 'Windows'
         run: |
           export UV_PYTHON=3.12
           VERSION_LOG=$(mktemp)
@@ -239,12 +296,48 @@ jobs:
           done
           cat "$VERSION_LOG"
 
+      - name: Test built wheels with CUDA (Windows)
+        if: contains(matrix.target.toolkit, 'CUDA') && runner.os == 'Windows'
+        shell: pwsh
+        run: |
+          $env:UV_PYTHON = "3.12"
+          $versionLog = Join-Path $env:RUNNER_TEMP "tilelang-wheel-versions.txt"
+          if (Test-Path $versionLog) {
+            Remove-Item -Force $versionLog
+          }
+
+          foreach ($backend in ("${{ matrix.target.test_backends }}" -split '\s+')) {
+            if ($backend -eq "") {
+              continue
+            }
+            $env:UV_TORCH_BACKEND = $backend
+            Get-ChildItem wheelhouse -Filter *.whl | ForEach-Object {
+              $wheel = $_.FullName
+              Write-Host "Testing wheel: $wheel"
+              if (Test-Path test-venv) {
+                Remove-Item -Recurse -Force test-venv
+              }
+              uv venv test-venv
+              $python = Join-Path (Resolve-Path test-venv).Path "Scripts\python.exe"
+
+              uv pip install -v --python $python $wheel
+              Push-Location $env:RUNNER_TEMP
+              try {
+                & $python -c "import tilelang, torch; print(tilelang.__version__, torch.__version__)" | Out-File -Append -FilePath $versionLog -Encoding utf8
+              } finally {
+                Pop-Location
+              }
+              Remove-Item -Recurse -Force test-venv
+            }
+          }
+          Get-Content $versionLog
+
       - name: Upload wheels
         # Not PR to save artifact storage, as wheels are only needed for releases.
         if: github.event_name != 'pull_request' || contains(github.event.pull_request.title, '[Release]')
         uses: actions/upload-artifact@v7
         with:
-          name: wheels-${{ matrix.python-version }}-${{ runner.os }}-${{ runner.arch }}-${{ matrix.target.toolkit }}
+          name: wheels-${{ matrix.target.python_version }}-${{ runner.os }}-${{ runner.arch }}-${{ matrix.target.toolkit }}
           path: wheelhouse/*.whl
           if-no-files-found: error
 

.gitignore

@@ -124,3 +124,6 @@ maint/host_checks/logs/*
 
 # agent skills folder
 .humanize*
+
+# uv Lockfile
+uv.lock