# ROCm on the Windows lane (`-Variant rocm`) `.\windows\Build-Buildkit.ps1 -Variant rocm` (amd64 only) builds the AMD image `:winamd64-rocm`. Since 2026-09-23 the ROCm layer sits in the **sdk slot**, like `Dockerfile.nvidia` on the GPU lane, so toolchain and media build on top of it and can turn on ROCm and AMD-GPU features. Every one of them is gated on `(Get-GpuEnvironment).HasRocm`: the cpu and nvidia lanes keep their flags and outputs. **Nothing on this page has been seen executing on a GPU.** Windows containers get DirectX only, and on this host GPU passthrough is additionally blocked by the 26100/26200 build skew ([windows-build-resources.md](windows-build-resources.md)). Every image check below is GPU-less: presence, linkage, listings, compile-only. GPU proof belongs on the bare host. ## What the rocm lane enables | Component | Enabled on the rocm lane | Mechanism | Needs the ROCm layer | | --- | --- | --- | --- | | GPU API loaders (sdk layer) | TheRock's `amdocl64.dll` registered as an OpenCL ICD; LunarG's Khronos `vulkan-1.dll` in System32 | OpenCL, Vulkan | OpenCL: TheRock's ICD | | GStreamer 1.29.2 | `hip`, `amfcodec`, `d3d11`, `d3d12` pinned `enabled` (already built by `auto` on every lane) | HIP (dlopen), AMF, D3D | hip: at run time | | FFmpeg n9.0.2 | AMF encoders, decoders, filters, `amf` hwdevice; Vulkan hwdevice, 9 hwaccels, 5 encoders, 18 filters, swscale's SPIR-V backend | AMF (driver), Vulkan (loader dlopened at run time) | no | | OpenCV 5.0.0 | OpenCL T-API (already on every lane); clBLAS/clFFT probes pinned off | OpenCL | TheRock's ICD loader + the registered `amdocl64.dll` ICD | | IREE | HIP HAL driver, `rocm` compiler target (gfx1201) | ROCm/HIP | at run time | | TVM | OpenCL runtime; ROCm codegen + runtime as a **spike** (`TVM_ROCM=1`), on a minimal LLVM of its own that carries AMDGPU | OpenCL, ROCm/HIP | spike: yes | | LiteRT-LM | GPU backend (WebGPU over Dawn on D3D12) | D3D12 | no | | ONNX Runtime | CPU + DirectML, plus the in-tree WebGPU EP as a **spike** (`ORT_WEBGPU=1`). ORT >= 1.23 has no ROCm EP | DirectML, WebGPU (Dawn on D3D12) | no | | PyTorch | torch 2.14.1+rocm10.1.0, torchvision 0.29.1+rocm10.1.0, built from source here for gfx1201 and gfx1200 | ROCm/HIP (AMD's runtime wheels) | no | | App venv LiteRT | `ai-edge-litert` with its WebGPU accelerator, built from source for the pinned Python (CON91; absent until that lands) | WebGPU (Dawn on D3D12) | no | | llama.cpp HIP | `ggml-hip` built from source at b11583 against the image's TheRock (gfx1200, gfx1201), with the same tag's official CPU tools, `C:\runtime\opt\llama.cpp-hip` | ROCm/HIP | yes: TheRock's clang, HIP runtime, hipBLAS/rocBLAS | | llama.cpp Vulkan | the same build's official Windows Vulkan zip (`ggml-vulkan`), `C:\runtime\opt\llama.cpp-vulkan` | Vulkan (the sdk layer's loader) | no | | MIGraphX + ORT plugin EP | MIGraphX 2.18.0 from source, `migraphx-ep.dll` as a **spike** | ROCm/HIP | yes | `-NoRocmSpikes` drops the three spikes: the migraphx stage, `TVM_ROCM` and `ORT_WEBGPU`. Every ONNX Runtime on this lane is the chain's own build, like on every other lane ([`onnxruntime-single-source.md`](onnxruntime-single-source.md)). The WebGPU EP is built into it, not installed from PyPI. **Not possible on Windows**, checked against upstream: the ORT ROCm EP (removed in 1.23), rocDecode / rocJPEG / rocAL (Linux only), any HIP path in FFmpeg, OpenCV or LiteRT, OpenCV's MIGraphX dnn backend (`NOT UNIX`-guarded), IREE's `amdgpu` HAL driver (needs HSA/ROCR), TVM with RCCL. ## The ROCm layer (`Dockerfile.rocm`) **Built by `Build-Buildkit.ps1 -Variant rocm`** (amd64 only). Since 2026-09-23 the ROCm layer sits in the **sdk slot**, where `Dockerfile.nvidia` sits on the GPU lane (owner directive 2026-09-23): ``` base -> sdk = rocm -> toolchain -> media (all branches + merge) -> migraphx -> llama -> torch -> final -> smoke ``` Toolchain and media now build on top of `GPU_TYPE=rocm`. That lets media turn on ROCm/AMD-GPU features behind `HasRocm` (below). The features stay optional: the cpu and nvidia lanes send the same configure/build flags as before. **Tags.** `bk-windows-base` is shared. Every tag from sdk on carries a `-rocm` infix (`Get-BkTag`): - `bk-windows-sdk-rocm` - `bk-windows-toolchain-rocm` - `bk-windows-media-core-{onnx,ffmpeg,opencv,hailo}-rocm`, `bk-windows-media-core-rocm` - `bk-windows-media-{litert,tvm}-rocm`, `bk-windows-media-rocm` - `bk-windows-media-migraphx-rocm`, `bk-windows-media-llama-rocm` - `bk-windows-torch-rocm` - `bk-winamd64-rocm` A rocm run therefore never overwrites a default image. The cpu and nvidia tags have not moved; nvidia still shares the default names. A golden table in `Driver.Variant.Tests.ps1` locks both lanes on amd64 and arm64. The old `bk-windows-rocm` tag is orphaned: remove it by hand (admin `nerdctl --namespace buildkit rmi`). Rocm runs from before this change wrote CPU-identical toolchain and media under the default tags; that is harmless and needs nothing. **Stages and switches.** - `-Stages` takes `base,sdk,toolchain,media,migraphx,llama,torch,final`. - `migraphx` and `llama` are rocm-only. Other variants drop them from the default list and refuse them when named. - `-Stages rocm` is gone. It is refused with a pointer to the sdk slot. - `migraphx` builds `Dockerfile.rocm-migraphx` (target `built`) FROM `bk-windows-media-rocm`. Its pins come from `Get-MediaBranchVersionArg -Branch rocm-migraphx`. - `llama` builds `Dockerfile.rocm-llama` (target `built`) FROM the migraphx image. - `torch` builds FROM the llama image. - `-NoRocmSpikes` drops `migraphx`, so llama builds FROM the merged media. It also passes `TVM_ROCM=0` instead of `1` to media-tvm and `ORT_WEBGPU=0` instead of `1` to the onnx stage of media-core. - These build-args are sent on the rocm lane only (`Get-BkRocmStageArg`): - `TVM_ROCM` to media-tvm; - `ORT_WEBGPU` plus the five `ORT_WEBGPU_WINDOWS_*` pins to `media-core-built-onnx` (the pins also go under `-NoRocmSpikes`, where nothing reads them); - `EXPECT_ROCM_SPIKES` to the smoke gate; - `TORCH_ROCM=1` plus every `TORCH_ROCM_WINDOWS_*` pin to `Dockerfile.torch`. The consumers declare them with default `0` or valueless, so the cpu and nvidia solves receive nothing new. The rocm sdk stage alone also gets `VULKAN_VERSION` and `VULKAN_RT_WINDOWS_ZIP_SHA256`, and the llama stage `LLAMA_CPP_VULKAN_SHA256`. - `-ConcurrentAux` children get `-Variant rocm` (and `-NoRocmSpikes`). Without it they would build litert/tvm FROM the default toolchain, under default tags. - `-Variant rocm -Stages torch,final` reuses `bk-windows-media-llama-rocm`. **Refused at launch:** - arm64; - `-Gpu` together with rocm; - a push tag other than `:winamd64-rocm` (and a non-rocm run pushing to it); - `-NoRocmSpikes` off the rocm lane; - `-Stages migraphx` together with `-NoRocmSpikes`; - a rocm `-Stages` list that skips a post-media stage between two it builds. Example: `media,torch` is refused, because torch would build FROM an earlier run's llama image (backlog #39). The merge-skip guard also refuses migraphx, llama, torch and final. **Refused in the stage:** `Install-Rocm.ps1` stops when its base already carries CUDA or ROCm (`GPU_TYPE`, `CUDA_ROOT`, `CUDA_PATH`). The rocm sdk builds FROM the plain base, never FROM an nvidia sdk and never on top of a second ROCm layer. **The `HasRocm` contract.** `Get-GpuEnvironment` returns `HasRocm` (true only when `GPU_TYPE=rocm`) and `RocmRoot`. It throws when `GPU_TYPE=rocm` but `HIP_PATH`/`ROCM_PATH` has no `lib\cmake\hip`, because a mis-plumbed layer would otherwise build CPU flags silently. Every rocm feature is gated on `HasRocm`, which the cpu and nvidia lanes can never enter. That includes features that need the AMD driver but not ROCm (AMF, OpenCL, D3D12, WebGPU), by owner decision. ORT builds the cpu lane's feature set, DirectML included, on this lane and logs `ROCm layer present: CPU+DML ORT`. The WebGPU spike adds its EP on top ([§ ONNX Runtime WebGPU EP](#onnx-runtime-webgpu-ep-rocm-lane-spike)). ORT >= 1.23 has no ROCm EP, and the dead `onnxruntime_USE_ROCM` branch was deleted. **The app venv's ORT is checked on every amd64 lane, rocm included.** Smoke section 21 requires that `C:\opt\OrchestrANT\.venv`'s onnxruntime lists `DmlExecutionProvider` and that its GenAI reports `is_dml_available()`. It also requires that the venv's onnxruntime is the chain wheel: the only owner of the package, the chain version, and every `.pyd`/`.dll` byte-identical to `C:\runtime\wheels\onnxruntime-*.whl`. Both are compiled-in facts, so no device is needed. The bytes matter because PyPI also ships onnxruntime 1.31.0 for cp314 win_amd64, without DirectML. **CMake isolation.** With `C:\TheRock\build\bin` on PATH, CMake treats `C:\TheRock\build` as a package prefix: a PATH entry ending in `bin` makes its parent a search prefix. TheRock ships `lib/cmake/{flatbuffers,llvm,clang,lld,hip,...}` and `share/cmake/nlohmann_json` there. ORT v1.31.0 asks for flatbuffers (`FIND_PACKAGE_ARGS 23.5.9`) and nlohmann_json (`3.10`) through `find_package` first. - `Invoke-CmakeConfigure` therefore appends `-DCMAKE_IGNORE_PREFIX_PATH=` on the rocm lane (`Get-CMakeRocmIsolationArgs`). Elsewhere it appends nothing, so cpu/nvidia command lines are unchanged. - A build that needs TheRock's packages (`find_package(hip)`, hipblas, ...) passes `-AllowRocmPrefix`. - Builds that bypass `Invoke-CmakeConfigure` must isolate themselves or show why they cannot leak. The toolchain's patched LLVM (raw cmake) cannot leak: - its runtimes find LLVM/Clang with `NO_DEFAULT_PATH` (llvmorg-23.1.1 `runtimes/CMakeLists.txt:89-90`); - TheRock's zlib/zstd sit only in `lib\rocm_sysdeps`; - there is no libxml2 and no python.exe in the tree. - A header-only leak (nlohmann_json) leaves no trace in a binary. To check a rocm run, grep its onnx stage log for `TheRock`: only the ignore-prefix line should match. **HIP device code** compiles only with TheRock's AMD clang (`C:\TheRock\build\lib\llvm\bin\clang++.exe`), passed by absolute path and never through PATH. The hub's patched clang-cl is built with `LLVM_TARGETS_TO_BUILD=AArch64;X86` and has no AMDGPU backend, so TVM's ROCm spike builds an LLVM of its own ([§ IREE and TVM](#iree-and-tvm-on-the-rocm-lane)). Host-only HIP consumers stay on clang-cl. **Smoke gate.** It runs the CPU suite and floor (170), then `Test-RocmImage.ps1` (`EXPECT_ROCM=1`): 1. the env contract; 2. no CUDA; 3. AMD's LLVM not shadowing `clang-cl`/`clang`/`lld-link`; 4. `.info\version` matching the pin; 5. device bitcode present; 6. a `hipcc` compile for gfx1201; 7. every `windows/scripts/build/rocm-checks/*.ps1`, sorted by name; 8. then MIGraphX presence and TVM's `TVM_ROCM` marker matching `EXPECT_ROCM_SPIKES`, the mode the driver passed (`OrtWebGpu.ps1` does the same for `ORT_WEBGPU`). The rules for the checks: - Each check writes zero or more finding strings (empty = pass) and needs no GPU. - A throw is a finding that names its file. - A missing or empty folder is a finding. - Any finding fails the gate. - Checks run under `Set-StrictMode -Version Latest` and `$ErrorActionPreference = 'Stop'`. - Their exit codes are ignored. - The gate bind-mounts `windows/scripts`, so a check edit needs no image rebuild. **Source.** AMD's Windows tar install (rocm.docs.amd.com, install → Windows → tar): `https://stable.repo.amd.com/rocm/core/tarball/therock-dist-windows--.tar.gz`. That is the same TheRock release, from the same host, that `linux/scripts/01-core/setup-rocm-repo.sh` installs on Linux as apt packages. `windows/scripts/host/Install-Rocm.ps1` downloads it, verifies it, and extracts it with System32 `tar.exe` into `C:\TheRock\build` (AMD's documented path). It then checks the layout and runs `hipcc --version` inside the container. That step now runs on the plain base, with no VS environment loaded. **No checksum from AMD.** Measured 2026-09-22: there is no `.sha256`, `.sig` or `.asc` sidecar, no `SHA256SUMS`, and no hash fragment on the pip index. `ROCM_WINDOWS_TARBALL_SHA256` is therefore self-measured and pinned, and the script refuses an empty or malformed pin. | Family | Covers | Bytes | SHA256 | | --- | --- | --- | --- | | `gfx120X-all` (pinned, 10.1.0, 2026-10-07) | RDNA4, incl. RX 9070 XT (gfx1201) | 2,244,477,973 | `e8d5acd522aa106d685485707e5085d491ead7d0d84a79ef74dc5995941403b2` | | `multiarch` (10.0.0, not re-measured for 10.1.0) | every supported GPU | 4,796,456,804 | `ebe454fe9ad663655177462187a4c86c72fd0537638f6cbea34660ddebf40056` | **Disk.** The 10.1.0 `gfx120X-all` tree unpacks to **9.65 GB** in 11,575 files (10.0.0: 9.56 GB, 10,544). At its peak the sdk stage holds the 2.24 GB tarball, the tree and its layer export at once, so `Get-StageDiskFloorGb` gives `Dockerfile.rocm` a 45 GB floor (the default is 40). That figure is calculated, not measured. The tree is stored once, as a parent layer, under every later rocm stage, so the toolchain and media floors are unchanged. What grows is the store itself: a second full toolchain + media set beside the default one. **Why the sdk slot, and what it costs (2026-09-23).** Until 2026-09-22 the layer forked after media, because nothing in media consumed ROCm. The owner moved it into the sdk slot, like nvidia, so media can build ROCm/AMD-GPU features behind `HasRocm`. The cost is that the rocm lane shares only `base` with the default image. Its first run rebuilds patched LLVM, all of media, the merge and torch. Cold figures from CPU amd64 run 20260919-021220: - patched-llvm 3749 s; - onnx / ffmpeg / opencv / core 1023 / 730 / 859 / 601 s; - litert 3535 s, tvm 2832 s; - merge 4314 s, torch 773 s; - about 5.3 h in total. sccache hashes no `PATH`/`HIP_PATH`/`ROCM_PATH` for C/C++, so unchanged compile commands may hit the shared cache; that is unmeasured. A ROCm pin bump, or an edit to `Install-Rocm.ps1` or `Dockerfile.rocm`, now rebuilds the whole rocm toolchain and media. The alternative placement, toolchain -> rocm -> media, would keep patched LLVM shared. It was not chosen. **PATH.** `Dockerfile.rocm` sets the variables AMD documents (`HIP_PATH`, `ROCM_PATH`, `HIP_PLATFORM`, `HIP_DEVICE_LIB_PATH`, `LLVM_PATH`) with one deliberate exception: AMD's page also adds `lib\llvm\bin` to PATH. That directory holds AMD's own `clang-cl.exe`, `clang.exe` and linker, and every compile here must use the patched clang-cl. So only `bin\` is added, and it goes **last**, because it also carries its own `flatc.exe`, `OpenCL.dll` and `amdocl64.dll`. Both rules now protect every toolchain and media compile, not just torch. `Rocm.Install.Tests.ps1` fails if either rule, or the Vulkan loader's placement below, is broken. **OpenCL ICD and Vulkan loader (sdk layer, since 2026-09-23).** Two runtime pieces make the AMD GPU APIs loadable in the image. Neither needs a GPU to check, and `rocm-checks/GpuLoaders.ps1` checks both. - **OpenCL ICD.** TheRock's `bin\OpenCL.dll` is the Khronos ICD loader (built from ROCm clr's `opencl/khronos/icd`). It loads the vendor ICDs listed under `HKLM\SOFTWARE\Khronos\OpenCL\Vendors`. `Install-Rocm.ps1` registers `C:\TheRock\build\bin\amdocl64.dll` there, as `REG_DWORD 0` in the 64-bit view. Until then nothing registered it, so the loader found no platform. The loader rules, read at the rocm-systems commit therock-10.0 pins (6b0e43f): - the value name is the DLL path, passed to `LoadLibraryA` as is, and only DWORD 0 counts (`icd_windows.c:117-151`); - `OCL_ICD_FILENAMES` is no substitute: it is read through `secure_getenv`, which returns nothing in an elevated process (`icd_windows_envvars.c`); - ICDs a display adapter registers come first, and the loader dedupes by file name (`icd.c:66-78`). With Adrenalin's `amdocl64.dll` exposed, the driver's copy wins. `amdocl64.dll` links PAL statically and imports only system DLLs. AMD's ICD always reports one platform, `AMD Accelerated Parallel Processing`, even with no device (clr `cl_icd.cpp:119-134`). Measured on this host with TheRock's DLL and no gfx120X GPU enabled: 1 platform, 0 devices, under 1 s. The layout gate now also requires `bin\OpenCL.dll` and `bin\amdocl64.dll`. - **Vulkan loader.** Server Core ships no `vulkan-1.dll`, so no Vulkan-linked binary could load in the image. The base image's Vulkan SDK cannot supply one: scoop installs it `copy_only=1`, its `Bin` has no loader, and `Helpers\VulkanRT.exe` is an installer that never runs. `windows/scripts/host/Install-VulkanLoader.ps1` therefore downloads LunarG's Runtime Components zip for `VULKAN_VERSION`: - pinned as `VULKAN_RT_WINDOWS_ZIP_SHA256`, which equals LunarG's published digest (`https://sdk.lunarg.com/sdk/sha//windows/.txt`); `bump_versions.py` `spec_vulkan` refreshes it with the version; - the signed x64 `vulkan-1.dll` goes into `C:\Windows\System32`, where `VulkanRT.exe` puts it on a real host. Its version must equal `VULKAN_VERSION`; - the verified copy and `VulkanRT-License.txt` go into `C:\vulkan-loader` (`VULKAN_LOADER_DIR`), which is not on PATH; - a System32 copy with other bytes is never overwritten: the build fails instead, so a base image that starts shipping its own loader is noticed. Why System32 and not a PATH directory: FFmpeg's Windows `dlopen` searches only the application directory and System32 (`compat/w32dlfcn.h` at n9.0.2), and Python 3.8+ resolves a `.pyd`'s imports without PATH, which is how TVM's Vulkan runtime loads. System32 comes before PATH in the standard search, so the copy also wins over any PATH copy; an app-local copy still wins over it. The loader layer comes after the ROCm RUN in `Dockerfile.rocm`, so a Vulkan bump does not re-download the 2.3 GB tarball. - **The default and nvidia images carry the same loader** since 2026-09-26 (BACKLOG CON25), installed by the final stage with `-SystemDir ''`: `C:\vulkan-loader` is appended to `PATH` and nothing is written to System32. That lets every statically linked consumer load (gstvulkan, IREE's Vulkan HAL, OpenCV vkcom, ggml-vulkan), not FFmpeg's `dlopen` or a `.pyd`. The rocm variant skips it: its `VULKAN_LOADER_DIR` is set. - **What the loader serves:** gstvulkan and amfcodec through it, llama.cpp's `ggml-vulkan.dll`, IREE's Vulkan HAL, OpenCV's vkcom backend, FFmpeg's `vulkan` hwdevice and TVM's Vulkan runtime. There is no Vulkan ICD in the container, so the loader reports zero devices there. That is the correct answer. - **Smoke: `rocm-checks/GpuLoaders.ps1`.** It sorts before `GStreamer.ps1`. It requires that `vulkan-1.dll` resolves to System32's copy and is byte-identical to the pinned one beside its licence, exports `vkGetInstanceProcAddr` and `vkEnumerateInstanceVersion`, and is not older than the SDK's `VK_HEADER_VERSION`. It requires the Vendors value (`REG_DWORD 0`, an existing file), `amdocl64.dll` loading with its ICD exports, and `clGetPlatformIDs` returning 0 or -1001 with an AMD platform listed. Zero platforms is a finding, because AMD's ICD always reports its platform. Both loaders run in a child pwsh with a 120 s timeout; device counts are only logged. Proven on 2026-09-25, when the rocm chain's smoke gate passed (`CHANGELOG.md`): `GpuLoaders.ps1` found the HKLM value in the final image, and AMD's ICD loaded there and listed its platform with no GPU, without a crash or a hang. If PAL ever misbehaves, dropping the `Register-OpenClIcd` call is the quick escape; `GpuLoaders.ps1` then reports the missing registration. **Licences.** See [Redistribution](#redistribution) below. **Not proven by anything here:** that HIP code runs. Windows containers get DirectX/DirectML only, so a GPU check needs the bare host (Windows 11 25H2, with the RX 9070 XT re-enabled outside any build window): - `hipInfo`; - the family's `-tests` tarball; - a torch matmul compared against the CPU. This one is required, because upstream TheRock#8379 reports torch+ROCm returning zeros on exactly that card. ### HIP compiles against MSVC 14.51 Under clang, the image's STL (VS 2026, MSVC 14.51) defines `isgreater`, `isgreaterequal`, `isless`, `islessequal`, `islessgreater` and `isunordered` as `constexpr` builtin wrappers. HIP makes those `__host__ __device__`, so clang's `__clang_cuda_math_forward_declares.h` and `__clang_hip_cmath.h` cannot declare their `__device__` versions. **Every** HIP compile in the image failed: the smoke gate's one-line `hipcc` kernel stopped with 20 errors on 2026-09-25, after 215 other assertions had passed. From then until 2026-10-11 the image shipped a fix inside the toolchain: an overlay of two headers at `C:\runtime\opt\hip-msvc-cmath`, each renaming the six names around an `#include_next` of the untouched original, loaded by a `clang.cfg` and `clang++.cfg` beside TheRock's `clang.exe` (TheRock ships no `*.cfg` of its own). A probe, `Test-HipMsvcCmath.ps1`, measured the choice in the rocm image: | Way to reach TheRock's clang | nothing | `HIPCC_COMPILE_FLAGS_APPEND` | `CPATH` | config beside `clang.exe` | | --- | --- | --- | --- | --- | | `hipcc` | 20 errors | compiles | compiles | compiles | | `clang -x hip` (so `amdclang++` and CMake's HIP language) | 20 errors | not reached | compiles | compiles | The config files were the only row-complete choice that no user environment could undo. **TheRock 10.1.0 no longer needs it, so the image ships TheRock's clang as AMD does.** With `--no-default-config`, which drops the overlay, the probe's kernel compiles under AMD clang 24.0.0git and MSVC 14.51.36231. This held through `hipcc`, `clang -x hip` and `amdclang++ -x hip`. First seen on 2026-10-07 with the tarball in a `:winamd64` container, then on the published `:winamd64-rocm` (`32c7cc51…`) on 2026-10-11. The overlay, both configs and the probe left the tree that day (CON78). The first rocm chain without them has to build MIGraphX, ggml-hip and torch. The smoke gate's one-line `hipcc` kernel stays the canary: when a toolset or TheRock bump brings the clash back, it fails with the same 20 errors, and the overlay comes back from git history. ## ONNX Runtime WebGPU EP (rocm lane, spike) The rocm lane's chain ORT (built from source by `Build-OnnxFromSource.ps1`) is built **with ORT's in-tree WebGPU execution provider**, over Dawn's D3D12 backend. It replaced the PyPI `onnxruntime-ep-webgpu` 0.4.0 plugin the app venv carried for part of 2026-09-23: that plugin was ORT code built by Microsoft, which the ORT single-source rule forbids ([`onnxruntime-single-source.md`](onnxruntime-single-source.md)). **Gating.** - It is a spike: on by default with `-Variant rocm`, off under `-NoRocmSpikes`. - The driver sends `ORT_WEBGPU=1|0` plus the five `ORT_WEBGPU_WINDOWS_*` pins to `media-core-built-onnx` (`Get-BkRocmStageArg`). cpu and nvidia get none of them: the ARGs are valueless, and their configure line is unchanged. - `ORT_WEBGPU=1` anywhere but the native rocm lane is a build error. - `ORT_ENABLE_WEBGPU` is the unrelated Linux knob; the Windows spike does not read it. **Why in-tree (`onnxruntime_USE_WEBGPU=ON`) and not the plugin build.** - The in-tree EP registers inside `onnxruntime.dll`, so `get_available_providers()` lists `WebGpuExecutionProvider` in every interpreter and GenAI sees it with no extra call. - The plugin build (`onnxruntime_USE_EP_API_ADAPTERS`) makes a separate provider DLL that each consumer must register with `register_execution_provider_library`, and GenAI would not see it without its own registration. The in-tree build needs none of that. - ORT's `onnxruntime_CUSTOM_DAWN_SRC_PATH` takes a pinned, pre-patched Dawn tree and turns Dawn's own dependency fetch off. **What is fetched, and how each fetch is pinned.** Nothing else is fetched. The configure gate fails if Dawn's `fetch_dawn_dependencies` runs. | Input | Source | Pin | | --- | --- | --- | | Dawn at the tag ORT's `cmake/deps.txt` names (`ORT_WEBGPU_WINDOWS_DAWN_VERSION`) | the GitHub tag archive | `ORT_WEBGPU_WINDOWS_DAWN_SHA256`, **and** a SHA1 equal to deps.txt's; the build refuses a deps.txt naming another tag | | ORT's six Dawn patches | ORT's own tree, read from `onnxruntime_external_deps.cmake` in PATCH_COMMAND order | ORT's pin. Applied with GNU `patch --binary --ignore-whitespace -p1` (Git's `usr\bin\patch.exe`), because five of them need fuzz against Dawn v20260916.214021 and `git apply` has none | | Dawn DEPS `jinja2`, `markupsafe`, `spirv-headers`, `spirv-tools` | as Dawn's `DEPS` names them | the commit read from DEPS; each checkout must equal it | | DXC release (`ORT_WEBGPU_WINDOWS_DXC_VERSION`, `_ASSET`) | the DirectXShaderCompiler GitHub release | `ORT_WEBGPU_WINDOWS_DXC_SHA256`; uses the x64 `dxcompiler.dll`, `dxcompiler.lib` and the three licence texts. `dxil.dll` stays in the zip: since ORT v1.31.0, Dawn creates the DXIL validator from `dxcompiler.dll` | ORT's other `cmake/deps.txt` downloads are unchanged: ORT fetches them itself, SHA1-pinned by ORT. Every rocm onnx build fetches these inputs again; there is no preseed yet. **Why DXC is a pinned release, not built.** DXC's `WinIncludes.h` includes `` under `_MSC_VER`, and the image has no ATL (the same C1083 the LLVM, IREE and TVM builds record). Its clang-cl branch also adds `-Wall`, which is `-Weverything` there, and sets `/Zi` in a way sccache cannot cache. So `Invoke-DawnPrebuiltDxcPatch` replaces Dawn's `AddSubdirectoryDXC()` with an imported `dxcompiler` target over the pinned release. `DAWN_USE_BUILT_DXC` stays ON, so Dawn keeps DXC (ShaderF16, subgroups) instead of falling back to FXC. The patch refuses to run if that block moves. **Dawn and Tint under clang-cl.** Both have upstream clang-cl branches, `DAWN_WERROR` is off by default, and no flags or patches are added. That was read from source, and the rocm chain proved it: the image it built by 2026-09-25 carries the spike and passed `OrtWebGpu.ps1`. That was Dawn v20260818.211311; ORT v1.31.0's v20260916.214021 waits for the next rocm chain. **What ships.** - `\bin\dxcompiler.dll`, beside `onnxruntime.dll`, which loads it at run time. The build refuses an `onnxruntime.dll` that imports it or `dxil.dll`. - The same DLL in the chain wheel's `onnxruntime\capi\`, checked byte for byte. - No `dxil.dll`. ORT v1.31.0 stopped shipping it, because Dawn v20260916.214021 creates the DXIL validator from `dxcompiler.dll` and loads nothing else. - `\ROCM-FEATURES.txt`, the marker: `ORT_WEBGPU=0|1` and, when on, `DAWN_VERSION`, `DXC_VERSION` and `DXCOMPILER_SHA256`. **Licences.** Dawn and Tint are linked statically into `onnxruntime.dll` and `onnxruntime_pybind11_state.pyd`; their BSD 3-Clause text is already ORT's own `dawn` entry in `ThirdPartyNotices.txt`. ORT's notices do not cover DXC, so the build copies DXC's three texts to `\licenses\directx-shader-compiler\` and appends them to the wheel's `onnxruntime\ThirdPartyNotices.txt` (`Add-OrtWebGpuWheelNotice`) before `bdist_wheel`. Both are `docs/deps/deps.json` rows, which `Rocm.OrtWebGpu.Tests.ps1` requires. **Build gates.** - After configure, `Get-OrtWebGpuConfigureFinding` requires `onnxruntime_USE_WEBGPU=ON`, `DAWN_FETCH_DEPENDENCIES=OFF`, `DAWN_USE_BUILT_DXC=ON`, D3D12 on, Vulkan off and the prebuilt-DXC log line, and throws otherwise. - After the wheel, `Get-OrtWebGpuWheelFinding` requires `WebGpuExecutionProvider` in the installed wheel, capi's `dxcompiler.dll` equal to the staged hash, no `dxil.dll` there, and the DXC notice entry. **Smoke (`rocm-checks\OrtWebGpu.ps1`, GPU-less).** It reads the marker and fails on a mode that does not match `EXPECT_ROCM_SPIKES` (a stale parent) or a DLL hash that does not match. With the spike off it requires no DXC leftovers. With it on, in the base `python` and in the app venv, it checks that: - `WebGpuExecutionProvider` is listed; - an identity-model session gets the WebGPU EP, or fails **only** with `Failed to get a WebGPU adapter`; - `onnxruntime_genai` loads that same `capi\onnxruntime.dll` and creates a WebGPU model, or fails for want of an adapter; - `dxcompiler.dll` loads from capi and exports `DxcCreateInstance`, and no `dxil.dll` sits in capi or beside `onnxruntime.dll`. Any other outcome fails. The whole-script test runs it against a real venv and a shim `python`, and each finding carries its `[base]` or `[venv]` label. **Proven on 2026-09-25**, when the rocm chain's smoke gate passed: Dawn and Tint compile under the image's clang-cl, and Server Core without a GPU gets as far as adapter selection. A failure before that has other error text, and the smoke goes red on it by design. Which of the two accepted outcomes it gave is not recorded here. **Not proven yet:** Dawn finding the Windows SDK's `d3dcompiler_47.dll` in the container, and the build time the onnx stage gains (probably large). **Bumping.** The Dawn pin follows `ONNXRUNTIME_VERSION`: when ORT moves, take the tag from its `cmake/deps.txt` `dawn` row and re-measure the archive's SHA256 (held with `bump:hold`). DXC is a report row: `bump_versions.py` (`spec_ort_webgpu_dxc`) proposes the newest release and, with `--write-all`, its single `dxc_YYYY_MM_DD.zip` name and SHA256. ## GStreamer on the rocm lane GStreamer 1.29.2 builds its AMD paths on **every** lane already. Meson `auto` needs only in-tree stubs, the vendored AMF headers and the Windows SDK. The CPU log `out/build-logs/rebuild-amd64-20260919-021219.log` installs these with no ROCm input: `gsthip.dll` + `gsthip-0.dll`, `gstamfcodec.dll`, and `gstd3d11`/`gstd3d12` (plugin + `-1.0-0` library). So the rocm lane adds pins, isolation and a proof, not new code. All of it is gated on `(Get-GpuEnvironment).HasRocm` in `windows/scripts/build/Build-GstreamerFromSource.ps1`. The cpu and nvidia meson lines are byte-identical (`SourceBuild.GstreamerRocm.Tests.ps1`). | Step | What | Why | | --- | --- | --- | | Pins | `-Dgst-plugins-bad:{hip,amfcodec,d3d11,d3d12}=enabled` | `auto` skips a plugin silently when a dependency goes missing; `enabled` fails meson setup. At 1.29.2 `enabled` only turns those skips into errors, and every dependency they need was found on the CPU lane. | | Isolation | Entries under the ROCm root are dropped from `PATH`, `PKG_CONFIG_PATH`, `PKG_CONFIG_LIBDIR`, `CMAKE_{PREFIX,INCLUDE,LIBRARY,PROGRAM}_PATH`, `INCLUDE` and `LIB`. This lasts from the pkg-config pre-flight through install. `PATH` gets TheRock's `bin` back, appended last, before the phase-9 plugin gate. | Meson does not configure through `Invoke-CmakeConfigure`. Its cmake dependency probe turns `C:\TheRock\build\bin` on `PATH` into the prefix `C:\TheRock\build`, which exposes `lib/cmake/{flatbuffers,llvm,clang,hip*,...}` and `share/cmake/nlohmann_json`. TheRock's `.pc` files (`flatbuffers.pc`, `nlohmann_json.pc`, `rocm_sysdeps` `zlib.pc`/`libzstd.pc`) are reachable through `PKG_CONFIG_PATH`. | | Proof | After meson setup, `build.ninja` and `meson-info/intro-dependencies.json` are scanned for the ROCm root in backslash, forward-slash and JSON-escaped spelling. A hit or a missing file throws. | A scrub nobody checks is an assumption. | **HIP is loaded at run time, not linked.** `gst-libs/gst/hip` compiles against in-tree stubs. At 1.29.2 the loader opens the first `HIP_PATH\bin\amdhip64_*.dll`, then `amdhip64_7.dll` by name. It opens hiprtc as `HIP_PATH\bin\hiprtc.dll`, which is `hiprtc0716.dll` for TheRock 10.1.0 (HIP 7.16; `hiprtc0715.dll` on 10.0.0). GLib opens both with `LoadLibraryW` by full path. So `amdhip64_7.dll`'s own imports, `rocm_kpack.dll` and `amd_comgr.dll`, resolve only through `PATH` or System32, never through `HIP_PATH\bin`. **TheRock's `bin` has to stay on the image `PATH`.** Measured against the real 10.0.0 gfx120X-all DLLs, and again against 10.1.0's on 2026-10-07 (`Get-GstHipRuntimeFinding`, no finding): amdhip64 exports all 37 names the loader resolves, including `hipGLGetDevices`/`hipGraphicsGLRegisterBuffer`, and hiprtc exports all 7. **Not enabled: `hip-amd-precompile`.** At 1.29.2, `sys/hip/meson.build` looks only for `hipcc.bin`, and TheRock ships `hipcc.exe`. Upstream fixed this after the tag (`4cdf9d9796`, MR 12331). The hiprtc JIT path already covers the converter kernels, so revisit at the next GStreamer bump. **Smoke check: `windows/scripts/build/rocm-checks/GStreamer.ps1`.** It is GPU-less and runs through `Test-RocmImage.ps1`. It asserts: - the HIP runtime above: exports, import closure, and `hiprtc-builtins.dll` on `PATH`; - per plugin (d3d11, d3d12, amfcodec, hip): the DLL exists; its import closure resolves on `PATH`/System32; no static link to `amdhip64`/`hiprtc`/`amfrt`; and `gst-inspect-1.0 ` loads it, with a private registry and a 300 s timeout. What it cannot assert: - **Elements** (`hipupload`, `hipconvert`, `amfh264enc`, `amfh265enc`). Without an AMD GPU both plugins register 0 elements and still load: hip finds no device, and amfcodec needs `amfrt64.dll` from the Adrenalin driver plus an AMD adapter. - **Loading hip in any Windows container, on any lane.** `gsthip` is built with GL interop, and `gstgl-1.0-0.dll` imports `OPENGL32.dll`, which Server Core does not ship. The load probe is skipped for that reason; everything else is still checked. - **Loading amfcodec on an image without the Vulkan loader.** amfcodec links `gstvulkan`, so the same skip applies there. The rocm image ships `vulkan-1.dll` in System32 (§ The ROCm layer), so amfcodec IS load-probed on it. At 1.29.2 its Windows `plugin_init` is D3D11-only and returns TRUE without `amfrt64.dll`, and gstvulkan registers its elements even when `vkCreateInstance` fails for want of an ICD. On the rocm image a missing loader is a `GpuLoaders.ps1` finding. Elements have to be proven on the bare host with the RX 9070 XT: `gst-inspect-1.0 hipconvert`, `amfh265enc`, `d3d12h265dec`. Never bake a GPU-less GStreamer registry into an image: these plugins call no `gst_plugin_add_dependency`. Not at 1.29.2: `onnxinference` `execution-provider=hip`/`migraphx`/`dml`. They exist only on GStreamer `main` (Aug–Sep 2026, untagged). Evidence: `subprojects/gst-plugins-bad/{meson.options,gst-libs/gst/hip/{meson.build,gsthiploader.cpp,gsthiprtc.cpp,gsthip-interop.cpp},sys/hip/{meson.build,plugin.cpp},sys/amfcodec/{meson.build,plugin.cpp,gstamfencoder.cpp},gst-libs/gst/d3d1{1,2}/meson.build,sys/d3d1{1,2}/meson.build}` at tag `1.29.2`; the TheRock 10.0.0 tarball listing; the CPU build log's install lines and import-walk summary. ## FFmpeg on the rocm lane: AMD AMF and Vulkan AMF is below; Vulkan has its own subsection, [§ Vulkan](#vulkan-rocm-lane-only-since-2026-09-23). **What AMF enables (rocm lane only).** - Encoders `h264_amf`, `hevc_amf`, `av1_amf`. - Decoders `h264_amf`, `hevc_amf`, `av1_amf`, `vp9_amf`. - Filters `vpp_amf` (scaling and colour), `sr_amf` (super resolution), `frc_amf` (frame-rate conversion), and the screen-capture source `vsrc_amf` (configured as `amf_capture`). - The `amf` hwdevice, which can take a D3D11 or D3D12 child device. The cpu and nvidia lanes are configured exactly as before. **None of it uses ROCm.** AMF is AMD's driver-level media SDK, and FFmpeg n9.0.2 has no HIP code at all (the configure script contains no `hip` token). It is tied to the rocm lane by owner decision, not by any technical need. The TheRock tarball carries no AMF files, and FFmpeg needs none from it. **How.** - `Build-FfmpegFromSource.ps1` downloads the AMF release's header-only asset, `AMF-headers-.tar.gz` (84 KB, never the 1.2 GB repo). It is verified against `AMF_HEADERS_SHA256` in `versions.env`, which equals the release asset's GitHub digest. - The headers are copied into `compat/amf/AMF` in the source tree and configure gets `--enable-amf --extra-cflags=-I/compat/amf`. This is the same route the ONNX Runtime headers already take on this lane. - `--enable-amf` is explicit, so a missing or too-old header stops configure with `amf requested but not found`. FFmpeg n9.0.2 requires AMF 1.5.2.0 or newer (configure:7902-7904), and the pin is `v1.5.2`. - After configure, two rocm-only gates fail the stage in minutes rather than at the smoke gate: - `ffbuild/config.mak` must enable all 12 AMF symbols; - `ffbuild/config.mak` must not name the ROCm tree in any spelling. FFmpeg is not a CMake build, so `-DCMAKE_IGNORE_PREFIX_PATH` does not apply. It is also safe by construction, because `Dockerfile.rocm` sets no `INCLUDE`/`LIB`/`PKG_CONFIG_PATH` and no TheRock executable is one of FFmpeg's build tools. - The headers are also installed to `C:\runtime\ffmpeg\include\AMF`. Every lane installs `libavutil/hwcontext_amf.h`, which includes ``; on the rocm lane that header now compiles for consumers. - Pins travel as build args: `AMF_HEADERS_VERSION` and `AMF_HEADERS_SHA256` go through the media-core map in `Get-MediaBranchVersionArg` into the `media-core-built-ffmpeg` stage. They are excluded from the merge stage. `bump_versions.py` refreshes the SHA whenever the tag moves. **Smoke check.** `windows/scripts/build/rocm-checks/FFmpeg.ps1`, run by `Test-RocmImage.ps1` under `EXPECT_ROCM=1`. It checks that the image's ffmpeg lists every AMF encoder, decoder, filter and the `amf` hwaccel, and that `--enable-amf` appears in its configuration line. It also checks that the AMF headers are installed and that no `amfrt*.dll` ships in the image. Since 2026-09-23 it checks the Vulkan components too (§ Vulkan below). Listing reads FFmpeg's static tables, so no GPU is needed. It does NOT prove that an AMF session opens. **Limits.** - **Runtime comes from the host driver.** `amfrt64.dll` belongs to the AMD Adrenalin driver and is never shipped. FFmpeg's Windows `dlopen` searches only the exe directory and System32 (`compat/w32dlfcn.h:134-138`), so PATH does not help. - **Inside a container: UNVERIFIED.** - Microsoft supports only DirectX in Windows containers and lists Server Core images as not supported for GPU acceleration. - On this host, container GPU passthrough is blocked by the 26100 image vs 26200 host build skew (docs/windows-build-resources.md). - The only community recipe copies the AMF DLLs from `HostDriverStore` into System32, and reports no encode test. - Verify on the bare host with the exported `C:\runtime\ffmpeg` instead: `ffmpeg -f lavfi -i testsrc2 -c:v hevc_amf -f null -`. - **Driver floors:** 10-bit encode needs AMF runtime 1.4.32 or newer (driver 23.30); the decoder's bit-depth detection expects 1.4.36. Per-codec coverage on RX 9070 XT (gfx1201) is UNVERIFIED. - **Generic encoder lookup differs per lane.** On the rocm lane, looking up an H.264/HEVC encoder by codec id, for example PyAV `add_stream('h264')` or `avcodec_find_encoder`, now returns `h264_amf`/`hevc_amf` instead of `*_d3d12va`, because the AMF encoders are registered first (allcodecs.c:864 before 867, 877 before 880). AV1 (`av1_d3d12va` is registered first) and all decoders are unchanged. The smoke asks for `mpeg4` by name, so it is unaffected. - **Already on every lane, so not made rocm-only:** D3D11VA/D3D12VA/DXVA2 hwaccels, the `*_d3d12va` encoders and the `*_d3d11`/`*_d3d12` filters. These come in through autodetect, and making them rocm-only would change the cpu and nvidia flags. The ONNX DNN backend's `device=dml` (DirectML) is also already built on every lane. - **Not enabled:** - OpenCL: never autodetected. It needs a statically built Khronos ICD loader, because TheRock ships no `OpenCL.lib` and no `cl_d3d11.h`. - HIP-native FFmpeg: does not exist upstream. **License.** The AMF headers are MIT; all 57 files carry the grant, plus AMD's codec-royalty notice. `amf` is in no nonfree or GPL list at n9.0.2 (configure:2029-2064, 2192-2195), so the image's FFmpeg stays `--enable-gpl --enable-version3`, "GPL version 3 or later", with no `--enable-nonfree`. **Evidence.** - FFmpeg n9.0.2 at https://raw.githubusercontent.com/FFmpeg/FFmpeg/n9.0.2/ : - `configure` (350, 2166, 3597, 3609-3730, 4257-4259, 4799-4801, 7902-7904, 8360-8361); - `libavcodec/allcodecs.c` (857-858, 864-865, 877-878, 915); - `libavfilter/allfilters.c` (444-446, 570); - `libavfilter/vsrc_amf.c` (387-388); - `libavutil/hwcontext.c` (91); - `compat/w32dlfcn.h` (134-138). - The AMF v1.5.2 release, https://github.com/GPUOpen-LibrariesAndSDKs/AMF/releases/tag/v1.5.2 : its asset `AMF-headers-v1.5.2.tar.gz` has digest sha256:d3c12eb3…11c99, and the tag peels to commit eadd00804d5f7e5cd8c85d540073198312870776. - A real FFmpeg 9.0.2 build with `--enable-amf` lists exactly these names, and the smoke check reports zero listing findings against it. ### Vulkan (rocm lane only, since 2026-09-23) **What it enables.** `--enable-vulkan`, built against the base image's Vulkan SDK (`VULKAN_VERSION`, headers at `VK_HEADER_VERSION` 363 since 2026-10-07; the list below was measured at 357): - the `vulkan` hwdevice, with `hwupload`/`hwmap`/`hwdownload` to and from it; - Vulkan Video decode hwaccels `av1`, `h264`, `hevc`, `vp9`, and compute-shader decode hwaccels `apv`, `dpx`, `ffv1`, `prores`, `prores_raw`; - encoders `h264_vulkan`, `hevc_vulkan`, `av1_vulkan` (Vulkan Video) and `ffv1_vulkan`, `prores_ks_vulkan` (compute); - 18 filters: `avgblur`, `blackdetect`, `blend`, `bwdif`, `chromaber`, `flip`, `gblur`, `hflip`, `interlace`, `nlmeans`, `overlay`, `scale`, `scdet`, `transpose`, `v360`, `vflip`, `xfade` (each `_vulkan`), plus the source `color_vulkan`; - swscale's Vulkan/SPIR-V backend (`libswscale/vulkan`). `vp9_vulkan` and `av1_vulkan` also need `vulkan_1_4`, that is headers >= 1.4.317 (configure:3547, 3625, 7849). None of it uses ROCm; like AMF it is on this lane by owner decision. The cpu and nvidia configure lines are byte-identical to before, which `SourceBuild.FfmpegRocm.Tests.ps1` asserts. **Why it was off.** The SDK ships no `vulkan.pc`, so `check_pkg_config_header_only` fails. The fallback `check_cpp_condition` (configure:7806) sees only CFLAGS, and n9.0.2 never assigns `$vulkan_incflags`. The SDK's `Include` missing from INCLUDE was not the cause. **How.** - `Get-FfmpegVulkanPlan` exists only when the AMF plan does, so `HasRocm` is still read in one place. It fails closed when `VULKAN_SDK` is unset or contains whitespace (configure runs the glslc probe unquoted), or when `Include\vulkan\vulkan.h`, `Include\spirv-headers\spirv.h` or `Bin\glslc.exe` is missing. - configure gets `--enable-vulkan --extra-cflags=-I/Include --glslc=/Bin/glslc.exe`. `--glslc` pins the SDK's compiler; otherwise configure takes the first `glslc`, `glslang` or `glslangValidator` on PATH. - SPIR-V is compiled at build time; n9.0.2 has no libshaderc or libglslang option. make runs glslc over the 58 `*.comp.glsl` shaders, and FFmpeg's own `bin2c` embeds the result in the libraries (`ffbuild/common.mak:113-133`). No shader compiler ships. - Shader compression follows zlib+gzip autodetection (configure:7275-7285). Which path the image takes is not measured yet; the build logs `config.mak: shader compression ON|OFF` right after the Vulkan gate. - **Build gate.** `Get-FfmpegVulkanConfigGap` fails the stage unless `ffbuild/config.mak` enables all 35 symbols (`CONFIG_VULKAN`, `CONFIG_VULKAN_1_4`, `HAVE_SPIRV_HEADERS_SPIRV_H`, 9 hwaccels, 5 encoders, 18 filters) and `GLSLC=` is the plan's glslc. `spirv_compiler` is in no configure list, so the components glslc builds stand in for it. - **No new pin.** The headers and glslc come from the SDK that `VULKAN_VERSION` already pins, so a `VULKAN_VERSION` bump now also recompiles the rocm FFmpeg. **Run time.** FFmpeg never links `vulkan-1.dll`; it dlopens it (`libavutil/hwcontext_vulkan.c` `load_libvulkan`), and its Windows `dlopen` searches only the application directory and System32. In the rocm image that finds the sdk layer's System32 copy (§ The ROCm layer); with no Vulkan ICD in a container, device creation then fails. On a bare host it loads the driver's loader. **Smoke check** (`rocm-checks/FFmpeg.ps1`, no GPU). It requires that `-encoders`, `-filters` and `-hwaccels` list the 5 encoders, the 18 filters and `vulkan`; that `ffmpeg -h decoder=X` lists `vulkan` under "Supported hardware devices" for each of the 9 hwaccel decoders; that the configuration line has `--enable-vulkan`; and that no DLL or EXE in the FFmpeg bin imports `vulkan-1.dll`, statically or delay-loaded. An `avutil-.dll` must be among the files read (hwcontext_vulkan lives there), so the import check cannot pass empty. It does NOT prove that a Vulkan device opens. **Unchanged by it.** A generic encoder lookup by codec id still resolves to `h264_amf`, `hevc_amf` and `av1_d3d12va`: the Vulkan encoders register after `*_d3d12va` and `*_amf` (allcodecs.c:861, 876, 891). `vulkan` is in no GPL, nonfree or version3 list (configure:2029-2070, 2192), so the licence stays `--enable-gpl --enable-version3`. The Vulkan-Headers are `Apache-2.0 OR MIT`, the SPIRV-Headers MIT, and glslc is a build-time tool only. **Not enabled: libplacebo.** FFmpeg's `libplacebo` filter needs libplacebo >= 7.351.0 through pkg-config (configure:7425). That would be a new meson source build with its own submodules to pin, plus a shader-compiler library shipped beside FFmpeg. `scale_vulkan` and the other Vulkan filters already cover GPU scaling. **Limits.** - The rocm chain built it with clang-cl on 2026-09-24, once `makedef` was fixed, and `FFmpeg.ps1` passed on its image on 2026-09-25 (`CHANGELOG.md`). The earlier host evidence was object files only: a real n9.0.2 configure with Strawberry mingw gcc 13.2 and the same SDK version enabled all 35 symbols, and all 103 Vulkan targets (45 C objects, 58 shaders) compiled. No DLL was linked, and that configure took gcc's path, not `--toolchain=msvc`. - No GPU-less check can show that a device opens or a shader runs. AMD's Windows Vulkan Video is reported buggy upstream (AMD-Gfx-Drivers#99; not re-checked). Verify on the bare host with the exported `C:\runtime\ffmpeg`, for example `ffmpeg -init_hw_device vulkan=vk:0 -filter_hw_device vk -f lavfi -i testsrc2 -vf format=nv12,hwupload,scale_vulkan=w=1280:h=720,hwdownload,format=nv12 -f null -`. Evidence (n9.0.2): `configure` 370, 405, 1114-1131, 3192-3677, 4149-4308, 7275-7287, 7800-7861; `libavutil/hwcontext_vulkan.c` 650-685; `compat/w32dlfcn.h`; `ffbuild/common.mak` 113-133; the `vulkan/Makefile`s under `libavcodec`, `libavfilter` and `libswscale`; `fftools/opt_common.c` 355-366. ## OpenCV on the rocm lane **What the rocm lane adds: nothing new at build time. The AMD path was already there.** OpenCV 5.0.0 has no HIP or ROCm code path. Its GPU paths that work on AMD hardware are all vendor-neutral, and every lane already compiles them: - **OpenCL T-API:** `WITH_OPENCL=ON`, `WITH_OPENCL_SVM=ON`. - **D3D11 interop:** `WITH_DIRECTX=ON`. - **Vulkan backend of the classic dnn engine:** `WITH_VULKAN=ON`. OpenCL builds against OpenCV's bundled `3rdparty/include/opencl/1.2` headers with no import library. At run time it loads `OpenCL.dll` by bare name through the standard DLL search and requires `clEnqueueReadBufferRect`. The rocm layer supplies a Khronos ICD loader, `C:\TheRock\build\bin\OpenCL.dll`, last on PATH, plus AMD's PAL-based `amdocl64.dll`, which `Install-Rocm.ps1` registers as an ICD (§ The ROCm layer). That is what gives the T-API a runtime on this lane. Until 2026-09-23 nothing registered the ICD, so the loader found no platform. `Build-OpencvFromSource.ps1` changes only on the rocm lane: - `Get-OpencvRocmCmakeArgs` adds `-DWITH_OPENCLAMDFFT=OFF -DWITH_OPENCLAMDBLAS=OFF`. clBLAS and clFFT default ON upstream and are probed through default paths. They are dormant libraries that neither TheRock nor the image ships. On cpu and nvidia the function returns nothing. - After configure, `Get-OpencvRocmConfigureFinding` fails the stage before the compile in two cases: - the summary lacks `OpenCL: YES`; - any line of `opencv-configure.log` names the ROCm root. This is the proof that TheRock's `lib/cmake` packages, zlib or OpenBLAS never reached OpenCV. - `rocm-checks/OpenCV.ps1`, run by `Test-RocmImage.ps1`, re-checks the shipped bytes and needs no GPU: - `cv2.getBuildInformation()` reports `OpenCL: YES` with `Link libraries: Dynamic load`; - no build-information line names the ROCm tree; - an `OpenCL.dll` with the 1.1 entry point loads through the standard DLL search. `cv2.ocl.haveOpenCL()` is reported warn-only, because the smoke container has no GPU. With the ICD registered it should print True, listing AMD's platform with no device; that is unverified in the image. **Not rocm-specific, but new on 2026-09-23:** on every Windows lane OpenCV stopped downloading its own ORT zip, links the chain's, and ships a working G-API DirectML EP. That change is owned by the ORT page ([`onnxruntime-single-source.md`](onnxruntime-single-source.md)). **Choosing the device.** The host exposes a dGPU and an iGPU, so select one with `OPENCV_OPENCL_DEVICE=::`, for example `:GPU:gfx1201`. `OPENCV_OPENCL_RUNTIME=` pins a specific loader. **Not available, and why:** - **HIP for `cv::cuda`/`cudev` and HIP-in-UMat.** Only unmerged upstream PRs exist, validated on Linux only: opencv#29285/#29527 with contrib #4147/#4178, and opencv#29372. - **MIGraphX dnn backend** (opencv#29726). It is `NOT UNIX`-guarded, and no Windows MIGraphX exists. - **RPP HAL** (opencv#29505). Closed unmerged. - **GPU inside the container.** Microsoft supports only DirectX in Windows containers, and on this host the 26100/26200 build skew blocks GPU passthrough. Measure the OpenCL T-API on the bare host or on a 26100 host. See [windows-build-resources.md](windows-build-resources.md). Upstream references, all at the 5.0.0 tag: `cmake/OpenCVDetectOpenCL.cmake`, `modules/core/src/opencl/runtime/opencl_core.cpp`, `CMakeLists.txt:333-344`, and TheRock's `win-tarball-list` (`bin/OpenCL.dll`, `bin/amdocl64.dll`; no `clAmd*`). ## IREE and TVM on the rocm lane The ROCm layer (`Dockerfile.rocm`) sits in the sdk slot, so `media-tvm` sees `GPU_TYPE=rocm` and TheRock at `C:\TheRock\build`. What that turns on: | Component | rocm lane | cpu / nvidia | | --- | --- | --- | | IREE runtime | HIP HAL driver (`-DIREE_HAL_DRIVER_HIP=ON`, `-DIREE_ROCM_TEST_TARGET_CHIP=` so the rocminfo probe is skipped) | unchanged | | IREE compiler | `rocm` target (`-DIREE_TARGET_BACKEND_ROCM=ON`: AMDGPU in IREE's in-tree LLVM, linked by `iree-lld`) | unchanged | | TVM runtime | `USE_OPENCL=ON` (TVM's lazy `OpenCL.dll` loader, no SDK) | `USE_OPENCL=OFF` | | TVM ROCm spike (`TVM_ROCM=1`; `-NoRocmSpikes` passes 0) | AMDGPU in a minimal LLVM that TVM builds for itself (`X86;AArch64;NVPTX;AMDGPU`; the toolchain LLVM on PATH has no AMDGPU), `USE_ROCM=C:/TheRock/build`, `tvm_runtime_rocm.dll` on HIP | off | **IREE needs nothing from TheRock at build time.** The HIP headers are vendored (`third_party/hip-build-deps`, HIP 6.1), so the configure keeps the rocm-lane package-prefix isolation. Two carried details: - **The DLL name.** IREE's Windows HIP driver dlopens only `amdhip64.dll`, but TheRock 10 and Adrenalin ship `amdhip64_7.dll` / `amdhip64_6.dll`. Without a fix the driver reports UNAVAILABLE on every AMD host. `Build-IreeFromSource.ps1` patches the name list to try `_7`, then `_6`, then the legacy name. TheRock's `amdhip64_7.dll` exports all 72 required symbols (checked 2026-09-23). Upstream candidate: iree-org/iree `runtime/src/iree/hal/drivers/hip/dynamic_symbols.c`. - **The device bitcode.** The rocm target downloads ocml/ockl at CONFIGURE time from shark-infra/amdgpu-device-libs v20231101, hash-pinned in IREE's own CMake. `IREE_ROCM_DEVICE_BC_SHA256` in versions.env mirrors that pin (`bump:hold`, slaved to `IREE_VERSION`). The build refuses to run when the two differ, or when upstream stops pinning. An IREE bump that moves the archive fails the rocm lane until someone re-derives the pin from `compiler/plugins/target/ROCM/CMakeLists.txt`. **TVM's `find_rocm` runs even with `USE_ROCM=OFF`.** It reads `$env:ROCM_PATH` and puts TheRock's `include\` (flatbuffers, nlohmann, CL, half, thrust, getopt.h) and `__HIP_PLATFORM_AMD__` on every translation unit. So without the spike, `Build-TvmFromSource.ps1` hides `ROCM_PATH` for the whole script, configure and scikit-build re-configure alike, then restores it. **The spike** carries two patches: - `rocm_device_api.cc` has no HSA on Windows: `kExist` asks `hipGetDeviceCount`. - `python/tvm/support/rocm.py` gets four Windows fixes: - `ld.lld` is found via PATHEXT, then `ROCM_PATH\lib\llvm\bin` by absolute path; - the bitcode comes from `HIP_DEVICE_LIB_PATH`; - the pre-2023 `oclc_daz_opt_*` and `oclc_correctly_rounded_sqrt_*` files are optional (TheRock does not ship them, ROCm/TheRock#5448); - a missing `rocminfo` falls back instead of raising. **Which LLVM TVM links.** Every amd64 lane puts the toolchain's patched LLVM (`C:\llvm-patched`) first on PATH. It is built with `LLVM_TARGETS_TO_BUILD=AArch64;X86` (`Build-LlvmFromSource.ps1`), so it has no AMDGPU. cpu, nvidia and the rocm lane without the spike link it unchanged, and never query it. On the spike, `Build-TvmFromSource.ps1` (`Get-TvmLlvmChoice`) asks the PATH `llvm-config --targets-built`. When AMDGPU is missing, that llvm-config counts as absent, and TVM builds its own minimal LLVM with `X86;AArch64;NVPTX;AMDGPU`. It uses the same pinned `LLVM_WINDOWS_VERSION` source tarball, goes through sccache, and installs to `C:\temp\llvm-dev\install`. That directory is not shipped: the merge copies only `C:\runtime\lib\tvm`. Adding AMDGPU to the toolchain LLVM instead would re-key every lane. After the choice, the build reads `--targets-built` back from the llvm-config TVM will link, and the spike throws when AMDGPU is missing. `ROCM-FEATURES.txt` records that read-back list as `LLVM_TARGETS`, not the requested one, so the rocm lane without the spike records `AArch64;X86`. Until 2026-09-23 the minimal-LLVM branch ran only when PATH had no llvm-config. On amd64 the spike therefore always found the patched LLVM, threw at the AMDGPU check before configure, and took the whole media-tvm RUN, IREE included, down with it. The rocm chain has since built the minimal LLVM with AMDGPU in the image: its media-tvm stage, IREE and that LLVM included, took 2:55:07, and `TVM.ps1` passed (`CHANGELOG.md`, 2026-09-25). The LLVM's own share of that time, and its memory use, are unmeasured. `llvm-config` must never resolve into the ROCm tree. **Compile only, no GPU.** A Windows container gets no HIP compute, so both gates compile and list: - The build gates check that `iree-compile` targets hip gfx1201, that `--list_drivers` shows `hip`, and that `import tvm` loads the OpenCL and ROCm sidecars. - The smoke checks `windows/scripts/build/rocm-checks/IREE.ps1` and `TVM.ps1` re-check the image. They assert a linked ELF64 AMDGPU code object with `EF_AMDGPU_MACH` 0x4E (gfx1201) from IREE (CLI and python) and from TVM's own AMDGPU codegen. They read `C:\runtime\lib\tvm\ROCM-FEATURES.txt` to know whether the spike ran. - `TVM.ps1` also requires AMDGPU in the marker's `LLVM_TARGETS` when `TVM_ROCM=1`. It compares `LLVM_TARGETS` both ways with what the runtime probe reads from `target.llvm_get_targets` (null when tvm_compiler has no LLVM). TVM names targets by `llvm::Triple::getArchTypeName`: X86 → `x86_64`, AArch64 → `aarch64`, NVPTX → `nvptx64`, AMDGPU → `amdgpu` at LLVM 23 (`amdgcn` before the rename, also accepted). A marker that claims a target the compiler lacks, or omits one it links, is a finding. - The unit suite runs `TVM.ps1` end to end against a fake python and the build's own marker writer, so dropping either LLVM check, or recording the requested list instead of the read-back one, fails a test. **Running on a host.** Kernels execute only on a host with AMD's driver and an RDNA3/4 GPU; IREE 3.11 rejects gfx1030/gfx1036. Python ≥3.8 does not search `PATH` for DLLs, so TVM's ROCm sidecar needs `amdhip64_7.dll` in System32 (the driver's copy) or a registered DLL directory. **Not enabled, on purpose:** | Feature | Why | | --- | --- | | IREE `amdgpu` HAL driver | needs HSA/ROCR, which Windows ROCm does not ship | | TVM hipBLAS | not now | | TVM rocThrust | forces hipcc/AMD clang as CXX | | TVM RCCL | not in TheRock 10 for Windows | Evidence: the research JSON (component "Apache TVM + IREE"). Pinned sources are IREE e4a3b04 (v3.11.0) and TVM 994e0216. ## LiteRT-LM GPU backend (rocm lane only) **What it enables.** On `-Variant rocm`, `C:\runtime\lib\litert-lm\bin` also carries LiteRT-LM's Windows GPU runtime, so `litert_lm_main.exe --backend=gpu` can run: - `libLiteRtWebGpuAccelerator.dll` (LiteRT's WebGPU accelerator; a closed-source prebuilt from the Apache-2.0 LiteRT-LM repo); - `libLiteRtTopKWebGpuSampler.dll`; - `webgpu_dawn.dll` (Dawn, which runs on Direct3D 12); - DXC's `dxcompiler.dll` and `dxil.dll`. DXC's licence texts go to `C:\runtime\lib\litert-lm\licenses\directx-shader-compiler\`. This feature does **not** use ROCm or HIP. For Windows, Google ships only the WebGPU accelerator: no HIP, Vulkan or OpenCL one (`litert/runtime/accelerators/gpu_registry.cc`). D3D12 is also the only GPU API that Windows containers accelerate. It is on the rocm lane by owner decision, but it works on any vendor's GPU. **How.** `windows/scripts/build/Build-LitertLmBazel.ps1` gates on `GPU_TYPE=rocm`, the same test as `Get-GpuEnvironment`'s `HasRocm`: 1. **[3/6]:** LiteRT-LM's `WORKSPACE` must pin the DXC zip at `LITERT_LM_DXC_ZIP_SHA256`. 2. **[4/6]:** the ROCm tree is hidden from Bazel before its server starts. Every env entry under `ROCM_PATH`/`HIP_PATH` is unset or filtered out of list variables (PATH, `HIP_PATH`, `ROCM_PATH`, `HIP_DEVICE_LIB_PATH`, `LLVM_PATH`, and any list var). `HIP_PLATFORM` is unset too. The script restores all of it in its `finally` block. Bazel does not go through `Invoke-CmakeConfigure`, so this is its equivalent of `CMAKE_IGNORE_PREFIX_PATH`. 3. **[5/6]:** one extra target, `@directx_shader_compiler//:dxc_dlls`. Bazel fetches it and checks it against LiteRT-LM's own `WORKSPACE` sha256. 4. **[6/6]:** the script copies these files into `C:\runtime\lib\litert-lm`: - the three prebuilt DLLs from `C:\llm\prebuilt\windows_x86_64` (git-LFS), each checked against its versions.env SHA256 first; - the DXC DLLs and licences from `C:\bzl\external\directx_shader_compiler`. On cpu and nvidia, the Bazel command stays `build //runtime/engine:litert_lm_main --config=windows --repo_env=ANDROID_NDK_VERSION=`, the env is not touched, and no file is added. **Pins.** Four keys in `linux/scripts/01-core/versions.env`, all tied to `LITERT_LM_VERSION`: - `LITERT_LM_WEBGPU_ACCELERATOR_SHA256`, `LITERT_LM_WEBGPU_SAMPLER_SHA256` and `LITERT_LM_WEBGPU_DAWN_SHA256`: the git-LFS oids of `prebuilt/windows_x86_64/*.dll` at the tag; - `LITERT_LM_DXC_ZIP_SHA256`: equal to the GitHub release digest of `dxc_2026_02_20.zip` in DXC v1.9.2602. `bump_versions.py` (`spec_litert_lm`, under `--write-all`) refreshes all four together with the version. They reach the branch through `Get-MediaBranchVersionArg media-litert` and the ARG+ENV in `media-litert-env`. They never reach the merge stage. **Smoke.** `windows/scripts/build/rocm-checks/LiteRtLm.ps1` runs from `Test-RocmImage.ps1` with `EXPECT_ROCM=1` and needs no GPU. It checks: - the files are present; - each DLL loads in a child pwsh whose PATH is only the bin dir plus System32, and its entry export resolves (`LiteRtAcceleratorImpl`, `LiteRtTopKWebGpuSampler_Create`, `wgpuCreateInstance`, `DxcCreateInstance`); - `--help` lists `--backend`. **Limits.** - **No GPU run has been proven.** GPU in a Windows container needs process isolation plus `--device class/5B45201D-F2F2-4F3B-85BB-30FF1F953599`. That passthrough is blocked on this host by the build skew (`docs/windows-build-resources.md`). Microsoft also lists Server Core base images as unsupported for GPU acceleration. - On the bare host, any D3D12 driver (AMD, NVIDIA or Intel) should work; upstream supports that path. - It is unverified whether Dawn accepts the WARP adapter a container can see. - Upstream `--backend` defaults to `gpu`. On the cpu and nvidia images, pass `--backend=cpu`. - Microsoft's DXC licence terms apply to redistributing `dxil.dll` and `dxcompiler.dll`. **Evidence** (all at the pinned versions): - LiteRT-LM v0.17.1: `README.md:141-144`; the `prebuilt/windows_x86_64` tree; `WORKSPACE:564-570` and `BUILD.directx_shader_compiler`; `runtime/components/sampler_factory.cc:469-482`; `runtime/engine/BUILD` (litert_lm_main exports `LiteRt*` via `windows_exported_symbols.def`); `runtime/engine/litert_lm_main.cc:55` (`--backend` default `gpu`). - LiteRT@9fe5be45 (LiteRT-LM's pin): `litert/runtime/accelerators/gpu_registry.cc`; `litert/c/litert_common.h` (`LITERT_HAS_WEBGPU_SUPPORT_DEFAULT 1`). - DXC: https://github.com/microsoft/DirectXShaderCompiler/releases/tag/v1.9.2602 (asset digest `sha256:a1e89031…`). - https://developers.google.com/edge/litert/next/gpu - https://learn.microsoft.com/en-us/virtualization/windowscontainers/deploy-containers/gpu-acceleration ## PyTorch on the rocm lane (torch stage) The rocm image's app venv (`C:\opt\OrchestrANT\.venv`) runs **PyTorch built from source in this repository** (owner decision 2026-09-29). It has torch 2.14.1+rocm10.1.0 and torchvision 0.29.1+rocm10.1.0 (the image interpreter's cpXY, win_amd64), built from the upstream `v2.14.1` / `v0.29.1` tags that `PYTORCH_VERSION` / `TORCHVISION_VERSION` name. They are compiled against this image's ROCm 10.1 SDK (`C:\TheRock\build`) for both GPUs of the pinned gfx120X-all family: gfx1201 (RX 9070 series) and gfx1200 (RX 9060 series). They load ROCm the way AMD's wheels do, from AMD's `rocm[libraries]` 10.1.0 runtime and a device wheel per GPU. The venv also needs `ai-edge-litert` (§ The LiteRT extra, below). The cpu and nvidia images are unchanged. Why here and not AMD's: AMD's Windows wheels for ROCm 10.0 stop at torch 2.13.0, and the app locks 2.14.1. torch 2.14 exists for ROCm 7.14 (pytorch.org, Linux only) and for ROCm 10.1 release candidates (`rc.repo.amd.com`), but not for 10.0. WebGPU for ONNX Runtime comes from the chain ORT itself ([§ ONNX Runtime WebGPU EP](#onnx-runtime-webgpu-ep-rocm-lane-spike)). The PyPI plugin `onnxruntime-ep-webgpu` that this venv carried for part of 2026-09-23 is gone, with its `TORCH_ROCM_WINDOWS_ORT_EP_WEBGPU_*` pins: it was ORT code not built by this chain. **How it is selected.** `windows/Dockerfile.torch` declares a global `ARG TORCH_ROCM=0` and ends with `FROM rocm-${TORCH_ROCM}`: - `app` is the old single stage, unchanged: `uv sync` with `PYTORCH_EXTRA=pytorch-cpu`, so the app lock resolves the same way on every lane. - `rocm-0` is `app` with no instruction of its own. This is what cpu and nvidia build, with the same cache key and layers as before. - `torch-rocm-wheels` builds the two wheels on the rocm `BASE_IMAGE`, **not on `app`**, so an `APP_REF` move never re-runs the compile. Only `rocm-1` mounts it, so cpu and nvidia never build it. It has two RUNs, torch and then torchvision. The torchvision ARGs sit below the torch RUN, so a torchvision failure, edit or bump never re-runs the torch compile. One RUN for both cost a finished 68-minute torch build on 2026-09-29, when only torchvision's import check failed. - `rocm-1` is built only when the driver passes `TORCH_ROCM=1`. That happens on the rocm lane only (`Get-BkRocmStageArg`, which also forwards the source build's versions and the sccache endpoint there). `rocm-1` runs `windows/scripts/build/Install-TorchRocm.ps1`, then `Build-TorchApp.ps1 -Mode verify`, so the app's own smoke suite runs on the ROCm torch. That verify runs the ORT census first: any ONNX Runtime distribution that is not a chain wheel fails the stage ([`failure-modes.md`](failure-modes.md#the-torch-stage-fails-with-ort-census-fail)). The installer itself also requires the venv's onnxruntime RECORD digest to be the same before and after it runs (`Assert-TorchRocmOrtUnchanged`). **The source build** (`windows/scripts/build/Build-TorchRocmFromSource.ps1`, then `Build-TorchvisionRocmFromSource.ps1`, following TheRock's `external-builds/pytorch/build_prod_wheels.py`): - **Sources.** `Save-GitCommitSource` fetches `TORCH_ROCM_WINDOWS_PYTORCH_COMMIT` and `_TORCHVISION_COMMIT`, the tags' commits, and git verifies every object. Submodules come at the superproject's gitlinks (`--depth 1`). Each commit's `version.txt` (`2.14.0a0` upstream) must carry the pinned x.y.z. - **Build venv** on the image's CPython (uv, no cache): torch's `requirements.txt` and `requirements-build.txt`, plus `build`. PyPI's `ninja` is removed so the image's 1.13.2 runs; 1.11.1 hangs and 1.13.0 breaks link.exe response files. The pinned `rocm` sdist, core and libraries go in too: `import torch` runs `_rocm_init.py`, and torchvision's setup imports torch. - The venv gets the base interpreter's `sitecustomize.py` and `python3.dll`, as the app venv does (`Build-TorchApp.ps1`). The image's CPython is built with clang, and without that shim `sysconfig.get_platform()` says `win32`. uv then resolves 32-bit wheels, and the torch wheel would be tagged `win32`. The first rocm build of this stage died on exactly that (2026-09-29): a 32-bit Cython failed to load while pyyaml built from its sdist. The builder now stops unless the venv reports `win-amd64`. - **Before the wheel:** HIPIFY (`tools/amd_build/build_amd.py`), then `torch/_rocm_init.py` (AMD's content, `check_version` = the release), force-included into the wheel because scikit-build-core drops files `.gitignore` names. - **OpenMP.** `torch_cpu.dll` imports MSVC's `libomp140.x86_64.dll` (`USE_OPENMP`, through MSVC's `libomp.lib`). TheRock copies it into `torch\lib`, but this wheel does not carry it. VS 18 (MSVC 14.51) ships the release DLL only under `debug_nonredist`, which Microsoft does not license for redistribution. The image's VS install puts it in `C:\Windows\System32`, where the loader finds it, and the builder stops before compiling if it is missing (`Assert-TorchRocmSystemLibomp`). So the wheel works only in this image, which is the only place it is installed. - **The environment** (`Get-TorchRocmCommonEnv`, `Get-TorchRocmTorchEnv`): - `ROCM_HOME`/`ROCM_PATH` are the SDK root and `CMAKE_PREFIX_PATH` is its `lib\cmake`. - TheRock's `clang-cl` is CC/CXX, plus `HIP_CLANG_PATH`. - `PYTORCH_ROCM_ARCH` comes from `ROCM_WINDOWS_GFX_FAMILY`, and OpenBLAS from the SDK's `lib\host-math`. - `PYTORCH_BUILD_VERSION=2.14.1+rocm` (`+rocm10.1.0` today) and `PYTORCH_EXTRA_INSTALL_REQUIREMENTS=rocm[libraries]==`. - `MAX_JOBS` allows 5 GB per job, and sccache wraps the host C/C++ when it is configured. - **Then** `python -m build --wheel --no-isolation`, an install into the build venv, and an `import torch` that prints the HIP version and the compiled arch list. The torch wheel is staged in `C:\torch-rocm-wheels`, and the torch tree is removed. The build venv, with torch installed, stays for the next RUN. - **The torchvision RUN** dot-sources the torch builder for its helpers and builds in that venv. It does not use `Start-MigraphxBuildSession`, which would reset the work dir. It adds pillow, because `import torchvision` imports PIL and torch's build requirements do not bring it. Then `setup.py bdist_wheel` (`FORCE_CUDA=1` on the GPU-less host) and an import that requires its C++ ops. `C:\torch-rocm-wheels` must then hold exactly the two wheels. Every step logs in full to the persistent `C:\sccache-logs` (`torch-rocm-*.log`, `torchvision-rocm-*.log`). - **Off in this first build, on purpose:** - **AOTriton** (flash and memory-efficient SDPA, `USE_FLASH_ATTENTION`/`USE_MEM_EFF_ATTENTION=OFF`). Its CMake clones dlfcn-win32, xz and aotriton and pip-installs into a venv of its own at build time, and nothing here pins any of it. SDPA falls back to the math kernel. - **`torch.distributed`** (gloo/libuv). - **Image IO in torchvision** (PNG, JPEG, WEBP). AMD's wheels lack it too. - **The SDK tree is AMD's tarball**, not the `rocm-sdk-devel` wheel TheRock's scripts expand. The tarball holds every devel path at the same place and has no symlinks, so the symlink-privilege expansion bug (TheRock#7807) does not apply. No public precedent for building against the tarball was found. It builds: see the next bullet. - **Measured 2026-09-29, on TheRock 10.0.0** (`-Variant rocm -Stages torch,final` on the 2026-09-28 rocm parent; the 10.1.0 build is unmeasured until the next rocm image): - The torch phase took 4051 s cold and 1719 s with the WebDAV sccache warm from the cold run. 2796 ninja steps, `MAX_JOBS=7`. - The torchvision RUN took 228 s. - The import printed `2.14.0+rocm10.0.0`, HIP `7.15.26333`, ROCm `10.0.0`, archs `gfx1200 gfx1201`. - In `rocm-1` the app's wheel smoke passed 13/15 with 0 failures (torch backward, torchvision `ops.nms`). - The smoke gate passed (215 assertions, 1 skipped) and `rocm-checks\Torch.ps1` found both GPUs covered. - The driver finished in 54 min. The GPU compute limit below still applies. **What the installer does.** - **Pins.** Five files are pinned by URL + SHA256 in `versions.env` (`TORCH_ROCM_WINDOWS__URL` / `_SHA256`, next to `ROCM_WINDOWS_*`): AMD's `rocm` sdist, `rocm-sdk-core`, `rocm-sdk-libraries` and one device wheel per GPU. They reach the stage as ARG defaults that `sync_versions.py` keeps in step. AMD publishes no hashes and its index carries no `#sha256=` fragments, so every SHA256 is self-measured (`curl -fL | sha256sum`). - **The built wheels** (`Get-TorchRocmBuiltWheel`): the mounted `C:\bkmnt\torch-rocm-wheels` must hold exactly one torch and one torchvision wheel, each win_amd64 and a `+rocm` build. Their SHA256 is taken on the spot, so the install stays `--require-hashes`. - **Coupling, checked at build time.** Every AMD file must be built for `ROCM_WINDOWS_RELEASE`. The venv must be the image interpreter's `cpXY` (`PYTHON_VERSION`). The app lock's torch/torchvision must be the release this build made, so a torch move in the app needs `PYTORCH_VERSION`/`TORCHVISION_VERSION` and their commits to follow. **Bump the whole `TORCH_ROCM_WINDOWS_*` block with `ROCM_WINDOWS_RELEASE`.** - **One device wheel per GPU.** `TORCH_ROCM_WINDOWS_SDK_DEVICE[_]` names the GPU (`Get-TorchRocmGpuTarget`). The unsuffixed one is gfx1201; gfx1200 uses `_GFX1200`. - **rocBLAS coverage.** Before any download, every GPU the image's rocBLAS ships kernels for (`%HIP_PATH%\bin\rocblas\library\TensileLibrary_lazy_.dat`, the files `rocm-checks\LlamaCpp.ps1` reads too) must have a device pin (`Assert-TorchRocmGpuCoverage`). Pinning more GPUs is allowed. A `ROCM_WINDOWS_GFX_FAMILY` change therefore fails the torch stage until the device pins follow. - **Offline install.** Pinned files are cached by hash on the torch stage's uv cache mount (`C:\uvcache\torch-rocm\\`) and re-hashed before reuse. Each is downloaded to its cache path directly, never as a `.part` renamed into place. On the mount that rename failed with `Could not find a part of the path` for the first wheel with a 68-character name, after two shorter ones had moved fine (2026-09-25). That is the create-then-rename class of [windows-build-lanes.md § Run-side wcifs symptoms](windows-build-lanes.md#run-side-wcifs-symptoms-process-isolation). A partial file is never used: a failed download deletes what it wrote, and every reuse re-hashes. The install is `uv pip install --force-reinstall --no-deps --no-index --no-build-isolation --require-hashes`. No index is contacted and nothing is resolved. The `rocm` sdist builds on the venv's setuptools (83.0.0 from the app lock). - **The no-GPU trap.** The `rocm` sdist's `setup.py` calls `offload-arch` to choose a GPU family. The build container has no GPU (TheRock's `offload-arch.exe` sits in `lib\llvm\bin`, which stays off PATH). The installer sets `ROCM_SDK_TARGET_FAMILY` to the first pinned GPU, gfx1201. That variable takes exactly one GPU, but it only fills the sdist's generic `device` extra, and `--no-deps` never reads it; the pinned per-GPU wheels stand in for AMD's per-GPU extras. `ROCM_BOOTSTRAP_DISABLE_DETECTION=1` is set too. The source build sets the same two for its build venv. - **Dependency confusion.** PyPI has an unrelated `rocm` 0.1.0. Pinning by URL + hash rules it out. - **The 10.1.0 index** adds `rocm-bootstrap` and `rocm-sdk-devel` packages. The 10.1.0 `rocm` sdist requires neither: against 10.0.0 it only renumbers its pins, drops the Linux `rocm_smi64` and adds the Linux `rpp` library entry (diffed 2026-10-07). TheRock's `therock-10.1` Windows preload list equals `Get-TorchRocmInitSource`'s, and `hiprtc0*.dll` matches the new `hiprtc0716.dll`. **Smoke.** `windows/scripts/build/rocm-checks/Torch.ps1` runs under `Test-RocmImage.ps1` (`EXPECT_ROCM=1`) and at the end of the install. It imports torch/torchvision/rocm_sdk without a GPU and asserts: - `+rocm` builds of torch and torchvision - a non-empty `torch.version.hip` - `torch.version.rocm` and `rocm_sdk` equal to the release - the runtime and device packages at the release - torch compiled for every GPU rocBLAS serves (`torch._C._cuda_getArchFlags()`, build-time data), and a `rocm-sdk-device-` for each (gfx1200 and gfx1201) - the chain onnxruntime still imports and lists `DmlExecutionProvider` - `ai_edge_litert.interpreter` imports, and `libLiteRtWebGpuAccelerator.dll` loads with `LiteRtAcceleratorImpl` resolved - every unconditional requirement of `ai-edge-litert` is installed; on the host, the import alone did not catch a missing `ml_dtypes` It never calls `torch.cuda`, and `torch.cuda.is_available()` is False in the container by design. ### The LiteRT extra `ai-edge-litert` is LiteRT's Python package with its prebuilt WebGPU accelerator `libLiteRtWebGpuAccelerator.dll` (Dawn statically linked), so it works on any vendor's D3D12 GPU; it is on the rocm lane by owner decision, and `rocm-checks\Torch.ps1` requires it. - **Until 2026-10-12 it was a PyPI pin**, 2.2.0's cp314 wheel, installed `--no-deps` in `rocm-1`. PyPI has no cp315 wheel (2.3.0, the newest, stops at cp314), so the Python 3.15 pin ended that. - **It is built from source for the pinned Python** (owner decision 2026-10-12, BACKLOG CON91) into the chain's wheel store, and `Build-TorchApp.ps1` installs it on every amd64 variant with the other store wheels. The accelerator stays the 2.3.0 wheel's closed-source DLL (ML Drift has no public source), pinned as `LITERT_WINDOWS_WEBGPU_ACCELERATOR_SHA256` and taken from `LITERT_WINDOWS_ACCELERATOR_WHEEL_*`. - **Until that wheel lands, the venv has none**: `Build-TorchApp.ps1` leaves the lock's ai-edge-litert out of `uv sync` with a warning when neither the store nor the lock has a wheel for the image's tag, and `rocm-1` fails on `Torch.ps1`'s LiteRT check. That red is on purpose; the check is not relaxed for the gap. - Its requirements are in the venv: the lock's own ai-edge-litert dependencies, plus `ml_dtypes`, which `Build-TorchApp.ps1` copies in from the base interpreter. - Its Dawn falls back to FXC when DXC is absent. Whether Dawn accepts the WARP adapter a container can see is unverified. **Limits.** - GPU compute is not proven by the image. The bare-host check (torch matmul on the GPU vs CPU, above) still applies. The upstream TheRock#8379, open, reports zeros from torch on RX 9070 XT with ROCm 7.14. - There is no Windows triton wheel, so GPU `torch.compile`/inductor is unavailable. - AMD's runtime wheels carry their own ROCm 10.1.0 (`_rocm_sdk_core`/`_rocm_sdk_libraries` in the venv, about 4 GB installed). It is the same release as `C:\TheRock\build`, which the source build compiled against, but it is a second copy. - No flash or memory-efficient SDPA (AOTriton is off), and no `torch.distributed`: § The source build says why. - The `torch-rocm-wheels` torch RUN is a one-hour compile (about 30 min with a warm sccache) whenever torch, the SDK or the torch builder moves. A torchvision move re-runs only its own RUN, and an app move neither. - The CPU torch installed by `uv sync` stays as dead bytes in the `app` layer, about 0.5 GB. That is the cost of leaving the cpu/nvidia stage untouched. - The rocm lane runs the app verify twice (CPU torch, then ROCm torch), about 1–2 minutes extra. - gfx1200 adds about 430 MB of downloads (the 10.1.0 rocm-sdk device wheel alone is 378,463,821 B) and about 0.7 GB installed. - `whl-next` is AMD's only Windows channel. A withdrawn file fails the download loudly; a cached copy keeps working. **Evidence for AMD's 2.13 wheels** (2026-09-23, Windows 11 host, CPython 3.14.7, uv 0.12.7; the runtime and device wheels are still these, torch and torchvision are now built here): - All ten original files downloaded and hashed; sizes match `Content-Length`. The torch wheel is 113,446,960 bytes; rocm-sdk-core is 758,050,981. - The gfx1200 wheels were hashed by streaming, sizes equal to `Content-Length` (51,106,952, 131,254 and 377,648,520 B). AMD's `rocm-10.0.0` sdist declares the per-GPU extras (`setup.py`, `src/rocm_sdk/_dist_info.py`). - The offline install replaced a CPU torch/torchvision in 59 s with the rocm-check clean. - `import torch` reports `2.13.0+rocm10.0.0 hip 7.15.26333 rocm 10.0.0`. - The app smoke's CPU ops pass with `HIP_VISIBLE_DEVICES=-1` (0 devices). - `amdhip64_7.dll` statically imports only system DLLs plus `amd_comgr`/`rocm_kpack`. The driver is loaded at run time, which is why `import torch` needs no GPU. Tests: `windows/scripts/tests/Torch.Rocm.Tests.ps1` and, for the source build's pure parts, `Torch.RocmSourceBuild.Tests.ps1`. ## llama.cpp HIP and Vulkan (`Dockerfile.rocm-llama`, rocm lane only) The stage installs two builds of the same llama.cpp tag, each in its own directory: the HIP build here, built against this image's ROCm, and upstream's prebuilt Vulkan build in [§ The Vulkan build](#the-vulkan-build-cruntimeoptllamacpp-vulkan). **What it adds.** The rocm image carries llama.cpp b11583 with a HIP backend built here, in `C:\runtime\opt\llama.cpp-hip`, which `LLAMA_CPP_HIP_HOME` names. It contains: - `ggml-hip.dll`, about 129 MB, built from the tag's source (`LLAMA_CPP_HIP_COMMIT`, archive `LLAMA_CPP_HIP_SOURCE_SHA256`) with TheRock's AMD clang, for gfx1200 and gfx1201; - `llama-server.exe`, `llama-cli.exe`, `llama.exe` and the other tools, and the CPU `ggml-cpu-*.dll` variants: the tag's official `llama-b-bin-win-cpu-x64.zip` (`LLAMA_CPP_CPU_SHA256`). That is how upstream composes its own `win-rocm` zip: release.yml's `windows-rocm` job builds only `--target ggml-hip`, and the release's merge step adds the `windows-cpu` zip's tools. Two things differ: - **No bundled HIP runtime.** Upstream ships `amdhip64_7.dll`, `rocm_kpack.dll` and `amd_comgr.dll` beside the exes (llama.cpp#26929). Here nothing of ROCm's sits in the directory: `ggml-hip.dll` loads the image's TheRock runtime, hipBLAS and rocBLAS from `C:\TheRock\build\bin`, last on PATH. - **Two GPUs, not twenty.** `GPU_TARGETS` comes from `ROCM_WINDOWS_GFX_FAMILY` (`gfx120X-all` = gfx1200;gfx1201), the GPUs the tarball's rocBLAS and hipBLASLt have kernels for. Upstream's DLL carries 20 targets and weighs 973 MB. **Why from source (owner decision 2026-10-07).** Upstream still builds against TheRock 10.0.0 (release.yml `windows-rocm`, `ROCM_VERSION: "10.0.0"`, at b11472 and still at b11583), and every build ships only a `win-rocm-10.0` zip. Its `ggml-hip.dll` and bundled runtime cannot serve a 10.1 image. **Where it sits.** base → sdk (rocm) → toolchain → media → [migraphx] → **llama** → torch → final. The stage is `Dockerfile.rocm-llama`, target `built`; `Build-Buildkit.ps1` forwards its `LLAMA_CPP_*` pins, `ROCM_WINDOWS_GFX_FAMILY` and the sccache endpoint. `-NoRocmSpikes` builds it directly on the merged media. **How it is built** (`windows/scripts/build/Build-LlamaCppHipFromSource.ps1`): - Rocm lane and amd64 only (`Initialize-MigraphxBuild`), and the pins must be well-formed before anything is fetched: a build number and a 40-hex lower-case commit. - The archive of `LLAMA_CPP_HIP_COMMIT` is verified against `LLAMA_CPP_HIP_SOURCE_SHA256` (`Resolve-PinnedSource`/`Save-PinnedSource`, the MIGraphX helpers). - Upstream's recipe: `-DGGML_HIP=ON -DGGML_BACKEND_DL=ON -DGGML_NATIVE=OFF -DGGML_CPU=OFF`, C and C++ by TheRock's `clang.exe`/`clang++.exe` passed by absolute path. On Windows ggml compiles HIP as C++ through `hip::device`, so the whole tree uses AMD's clang; clang-cl has no AMDGPU backend. TheRock 10.1's HIP headers compile against MSVC 14.51 without the retired `hip-msvc-cmath` overlay ([§ HIP compiles against MSVC 14.51](#hip-compiles-against-msvc-1451)). - Nothing downloads at build time: the tools, the server's prebuilt UI (fetched unpinned from a Hugging Face bucket) and BoringSSL are switched off, and `FETCHCONTENT_FULLY_DISCONNECTED` is set. `LLAMA_BUILD_NUMBER`/`_COMMIT` come from the pin, because the tarball has no git history. - Only the `ggml-hip` target is built. Its PE machine must be amd64, and the pins it was built from go to `llama-cpp-hip-build.json` next to it. **How it is installed** (`windows/scripts/build/Install-LlamaCpp.ps1 -Backend hip -BuiltDir `). One installer serves both builds; a backend table (`Get-LlamaCppBackendSpec`) holds what differs. For HIP it fails closed: - It runs only when `Get-GpuEnvironment` reports `HasRocm`, and only when ROCm's bin has `amdhip64_7.dll`, `hipblas.dll` and `rocblas.dll`. - The build record must name `LLAMA_CPP_HIP_BUILD` and the image's `ROCM_WINDOWS_RELEASE`. - The CPU zip is verified against `LLAMA_CPP_CPU_SHA256`. It must be flat, hold the load-bearing files (`llama-server.exe`, `llama-cli.exe`, `ggml-base.dll`, `ggml.dll`, `llama.dll`), and carry neither a `ggml-hip.dll` nor any DLL name ROCm's bin has. - The tag's LICENSE is verified against `LLAMA_CPP_HIP_LICENSE_SHA256`. It writes `llama-cpp-hip-manifest.json` last: size and SHA256 of every file, and the source build's record. The same RUN then runs `rocm-checks\LlamaCpp.ps1 -Backend hip`, so a bad pin or a broken build fails this stage. **What the smoke check proves** (`windows/scripts/build/rocm-checks/LlamaCpp.ps1`). Run with no arguments, as the smoke gate runs it, it grades both builds. For HIP it runs with no GPU: - the shipped bytes equal the manifest's; - the directory is not on PATH; - nothing of ROCm's sits next to llama-server; - every static import of `ggml-hip.dll` resolves, transitively, a ROCm one to ROCm's bin, and exports what is imported; - a child pwsh loads `ggml-hip.dll` and resolves `ggml_backend_init`, with `ggml-base.dll` from the llama directory and every ROCm DLL it imports from ROCm's bin; - `llama-cli --list-devices` loads ggml-hip and asks the HIP runtime for devices. In a container that answers `ggml_cuda_init: failed to initialize ROCm: no ROCm-capable device is detected` and exits 0; on a GPU host it lists the devices. Any other outcome, or no such line at all, is a finding; - ggml-hip has device code for every GPU that ROCm's rocBLAS has kernels for; - `llama-server --version` reports the pinned build. Measured 2026-10-07 at b11472 in a `:winamd64` container with TheRock 10.1.0 installed by `Install-Rocm.ps1`, the hub clone mounted: - `Build-LlamaCppHipFromSource.ps1` built `ggml-hip.dll` (128,678,400 B) cold in 3 min 27 s, 177 s of it ninja at `-j14`, beside a MIGraphX build. The log is 19 MB; without `-Wno-ignored-attributes` it was 160 MB. - `Install-LlamaCpp.ps1 -Backend hip` installed it, and `LlamaCpp.ps1` (each backend, then both as the smoke gate runs it) reported no finding. - `ggml-hip.dll` imports `ggml-base.dll`, `hipblas.dll` and `amdhip64_7.dll`; its device code is gfx1200 and gfx1201. - `llama-server --version`: `version: 0.6.0-dev (build 11472, commit d0b490f25)`, `built with Clang 20.1.8`, which is the CPU zip's frontend. - `llama-cli --list-devices` exits 0 in 0.2 s with `ggml_cuda_init: failed to initialize ROCm: no ROCm-capable device is detected` and `Available devices: (none)`. **Limits.** - **Nothing proves a kernel runs.** Microsoft accelerates only DirectX inside Windows containers, and GPU passthrough on this host is blocked by the 26100/26200 build skew (see [windows-build-resources.md](windows-build-resources.md)). - **RDNA4 only.** `ggml-hip.dll` carries device code for gfx1200 and gfx1201 alone, the GPUs the gfx120X-all tarball's rocBLAS/hipBLASLt serve. Any other AMD GPU (for example a gfx1036 iGPU) takes the Vulkan build. - **On a bare host the driver's runtime wins.** With no bundled copy, the loader takes Adrenalin's System32 `amdhip64_7.dll` before TheRock's on PATH (llama.cpp#26929). The image has no System32 copy. Whether a given driver's runtime serves TheRock 10.1's hipBLAS is unproven. - **Driver floor unknown.** The minimum Adrenalin driver for TheRock 10.1.0's HIP 7.16 runtime is not known. - **The frontend is upstream's binary.** The tools and CPU backends are the pinned CPU zip, built by upstream with VS's clang; only `ggml-hip.dll` is compiled here. A full source build would lose the server's web UI and HTTPS, which upstream fetches unpinned at build time. - **Size.** 178 MB in 54 files and +1 layer, down from about 1.24 GB with upstream's zip. - **Pinning.** Upstream publishes several builds a day, all flagged prerelease. Renovate reports new builds (approval-gated). `python docs/scripts/bump_versions.py --write-all` moves the build to the newest one that publishes both a win-cpu and a win-vulkan zip, with the tag's commit, the source archive's SHA256, both zips' digests and the LICENSE. **Evidence.** - https://github.com/ggml-org/llama.cpp/releases/tag/b11583 - https://raw.githubusercontent.com/ggml-org/llama.cpp/b11583/.github/workflows/release.yml (`windows-cpu` job, lines 872-939; `windows-rocm` job, lines 940-1096) - https://raw.githubusercontent.com/ggml-org/llama.cpp/b11583/ggml/src/ggml-hip/CMakeLists.txt (`CXX_IS_HIPCC` on WIN32) - https://raw.githubusercontent.com/ggml-org/llama.cpp/b11583/tools/ui/CMakeLists.txt (the prebuilt UI download) - https://raw.githubusercontent.com/ggml-org/llama.cpp/b11115/common/arg.cpp (`--version` handler) and tools/server/server.cpp (`llama_server` start-up order) - https://github.com/ggml-org/llama.cpp/issues/26929 ### The Vulkan build (`C:\runtime\opt\llama.cpp-vulkan`) **What it adds.** llama.cpp's official Windows Vulkan zip of the same build, b11583: asset `llama-b11583-bin-win-vulkan-x64.zip`, 32,112,442 B. It lives in its own directory, which `LLAMA_CPP_VULKAN_HOME` names, and it is never on PATH. It is the vendor-neutral path to an AMD GPU, and the only one for GPUs the gfx120X-all rocBLAS has no kernels for, such as the gfx1036 iGPU. - Upstream builds only `ggml-vulkan.dll` (`-DGGML_VULKAN=ON -DGGML_CPU=OFF -DGGML_BACKEND_DL=ON`, Vulkan SDK 1.4.357.0) and then adds the windows-cpu zip's tools. - Measured 2026-10-07 at b11472, and again at b11476 and b11583 (2026-10-13): the Vulkan zip is the CPU zip's 51 files, byte-identical (CRC32 and size), plus `ggml-vulkan.dll`. The HIP directory ships the same 51 beside its own `ggml-hip.dll`. **Pins.** Only `LLAMA_CPP_VULKAN_SHA256` is Vulkan's own. The build stays one pin, `LLAMA_CPP_HIP_BUILD`, the asset name follows from it, and the LICENSE pin is the HIP build's, because it is the same tag's LICENSE. The ARG is declared after the HIP RUN, so a Vulkan-only change keeps the HIP layer and its ggml-hip compile cached. **How it is installed:** `Install-LlamaCpp.ps1 -Backend vulkan`, in its own RUN, which then runs `rocm-checks\LlamaCpp.ps1 -Backend vulkan`. It fails closed: rocm lane only; the zip verified against its SHA256 and flat; `ggml-vulkan.dll`, `ggml-base.dll`, `ggml.dll`, `llama.dll` and `llama-server.exe` present; no bundled `vulkan-1.dll` (the loader comes from the image) and no ROCm DLL; the LICENSE fetched at the tag; `llama-cpp-vulkan-manifest.json` written last. **The Vulkan loader.** `ggml-vulkan.dll` statically imports `vulkan-1.dll` (for `vkGetInstanceProcAddr`, `vkGetDeviceProcAddr`, `vkGetPhysicalDeviceFeatures2` and `vkCmdCopyBuffer`). In the image that is the sdk layer's System32 copy (§ The ROCm layer); on a bare host it is Adrenalin's. The check accepts a loader from System32 or from a PATH entry, and refuses one in the llama directory or the Windows directory. **What the smoke check proves,** with no GPU: - the shipped bytes equal the manifest; a file that cannot be read is a finding; - the directory is not on PATH; - a child pwsh loads `ggml-vulkan.dll` and resolves `ggml_backend_init`, with `ggml-base.dll` from the llama directory and `vulkan-1.dll` from where the loader order says; - `vkEnumerateInstanceVersion`, which the loader answers without an instance, driver or GPU, reports API 1.2 or later. `ggml_vk_instance_init` refuses anything lower (ggml-vulkan.cpp@b11115:4969-4973); - `llama-server --version` reports the pinned build. It never calls `ggml_backend_init`, because that creates a VkInstance. `--list-devices` is no probe: release builds load backends silently, and `ggml_backend_vk_reg` returns NULL on any Vulkan error, so a missing loader and a missing GPU both print "(none)". **Why a separate directory.** Each directory is verified file by file against its own manifest. `ggml-vulkan.dll` beside `ggml-hip.dll` would load, but `ggml_backend_load_all` loads HIP first and llama.cpp then skips a second device with the same `device_id`, so on the RX 9070 XT Vulkan would only add the iGPU. ggml also searches the current directory for backends, so do not run one build with the other's directory as CWD. **Limits.** - No GPU in the container: `ggml-vulkan` registers no device there. On the bare host it uses Adrenalin's Vulkan ICD. Nothing proves a kernel runs. - Size: about +95 MB and +1 layer. - **Windows Defender** flagged `llama-gguf-split.exe` (byte-identical in every zip of a build) as `Trojan:Win32/Wacatac.B!ml`, an ML heuristic, on this host on 2026-09-23, and quarantined it. The check then reports it missing. The build stores are excluded by `Sync-DefenderExclusions.ps1`; a bare-host export of either llama directory meets the same detection. Whether it is a false positive is unverified. - ggml-vulkan compiles in Khronos Vulkan-Headers code, and the zip ships no Khronos licence text. Whether that creates a notice obligation is unverified. **Evidence** (at b11115): the release's asset digest; `release.yml` 1137-1224 and 1876-1884; `ggml/src/ggml-backend-reg.cpp` 480-600; `ggml/src/ggml-vulkan/ggml-vulkan.cpp` 4959-4973 and 15853-15873; `src/llama.cpp` 222-256. ## MIGraphX and the ORT plugin EP (rocm lane, spike) `-Variant rocm` builds `windows/Dockerfile.rocm-migraphx` (target `built`) on the merged rocm media, before the llama stage. `-NoRocmSpikes` skips it. cpu and nvidia never solve it, and no other Dockerfile or media build-arg map carries its pins. What it produces: | Path | Contents | |---|---| | `C:\runtime\lib\migraphx` (`MIGRAPHX_ROOT`) | AMD MIGraphX 2.18.0 (the head of `release/rocm-rel-10.1`, pinned by commit; no `rocm-10.1` tag yet) built from source: `bin\migraphx*.dll`, `migraphx-driver.exe`, `migraphx-hiprtc-driver.exe`, `lib\cmake\migraphx`, headers | | `C:\runtime\lib\onnxruntime-ep-amdgpu` (`ORT_AMDGPU_EP_ROOT`) | AMD's out-of-tree ONNX Runtime plugin EP `migraphx-ep.dll` ([onnxruntime/onnxruntime-ep-amdgpu](https://github.com/onnxruntime/onnxruntime-ep-amdgpu), commit on `gpuep-releases/gpuep-rel-2611`), built against the image's ORT 1.31.0, with its MIGraphX closure and TheRock's `amdhip64_7`/`amd_comgr`/`hiprtc*` beside it | The MIGraphX stage leaves ORT and GenAI untouched. On this lane they are built with the cpu lane's feature flags, DirectML included, on the rocm sdk layer and with the rocm-lane `CMAKE_IGNORE_PREFIX_PATH`. ORT also carries the WebGPU spike, which is on whenever this stage builds ([§ ONNX Runtime WebGPU EP](#onnx-runtime-webgpu-ep-rocm-lane-spike)). Nobody has compared their bytes with the cpu image's. The EP is built against that chain ORT and is loaded at run time with `onnxruntime.register_execution_provider_library(, r"C:\runtime\lib\onnxruntime-ep-amdgpu\migraphx-ep.dll")`. How it is built (`Build-MigraphxFromSource.ps1`, `Build-OrtAmdgpuEpFromSource.ps1`, helpers in `WindowsMigraphx.Common.psm1`): - **Two compilers, on purpose.** MIGraphX has HIP device code, and the hub's clang-cl has no AMDGPU backend. MIGraphX therefore compiles with TheRock's `lib\llvm\bin\clang++.exe`, passed by absolute path; `lib\llvm\bin` never goes on PATH. Its host-only deps (abseil 20250512.0, protobuf v30.0, msgpack-c cpp-3.3.0, SQLite 3.50.4) build with clang-cl, as upstream's Windows CI does. The EP is host-only C++ and builds with clang-cl. - **Targets are explicit.** `GPU_TARGETS` comes from `ROCM_WINDOWS_GFX_FAMILY` (`gfx120X-all` = `gfx1200;gfx1201`), never from a host probe. - **Configuration no upstream CI builds.** rocMLIR is not in the TheRock tarball, so `MIGRAPHX_ENABLE_MLIR=OFF`. That costs the rocMLIR fusions (performance, not correctness). Also used: `BUILD_DEV=OFF` (hiprtc JIT), composable_kernel off, CMake 4.4.3, TheRock 10.1. Upstream's only Windows CI is build-only: gfx942, `BUILD_DEV=On`, ROCm 7.13. - **CRT.** MIGraphX and its deps are `/MD`. The EP is `/MT`, as upstream forces it; it talks to ORT and MIGraphX only through C APIs. - **Supply chain.** Every archive is SHA256-pinned in `versions.env` (`MIGRAPHX_WINDOWS_*`, `ORT_AMDGPU_EP_*`). The EP declares 8 FetchContent URLs (11 with DirectML) with no `URL_HASH`. Each one is pre-seeded through `FETCHCONTENT_SOURCE_DIR_*` from a verified archive, and so is the abseil that protobuf v34.1 fetches. The build refuses when the pinned commit declares a URL no pin rebuilds byte-for-byte, and `FETCHCONTENT_FULLY_DISCONNECTED` with CMP0170 turns any other fetch into a configure error. rocm-cmake is the one dependency with no archive pin. MIGraphX `rocm-10.0` calls `rocm_add_version_resource` (rocm-cmake 33541cd51f), and TheRock's rocm-cmake predates it by four commits: `Unknown CMake command`, 2026-09-25. So phase 2 reads the commit that MIGraphX's own `requirements.txt` pins, from the SHA256-pinned MIGraphX tree (`Get-MigraphxRocmCmakeCommit`, 40-hex only). It fetches exactly that commit with git, which verifies every object against the id (`Save-GitCommitSource`), and installs it into the deps prefix, which precedes TheRock on `CMAKE_PREFIX_PATH`. A MIGraphX bump carries its own rocm-cmake, with nothing to re-derive in `versions.env`. nlohmann_json stays TheRock's, through a two-file package shim in the deps prefix (`Write-NlohmannJsonConfigShim`). TheRock's copy was installed by an MSVC-style build, so its exported target lists `/nlohmann_json.natvis` as an interface source, and TheRock's dist does not ship that file (`Cannot find source file`, 2026-09-25). The shim includes TheRock's config and clears that property. Until 2026-10-11 HIP sources compiled with a two-header overlay ([§ HIP compiles against MSVC 14.51](#hip-compiles-against-msvc-1451)), at first passed by MIGraphX's own `-isystem` (`Write-HipMsvcCmathOverlay`). Under clang, MSVC 14.51's `` defines `isgreater`, `isgreaterequal`, `isless`, `islessequal`, `islessgreater` and `isunordered` as `constexpr` builtin wrappers, which HIP makes `__host__ __device__`. clang's `__clang_cuda_math_forward_declares.h` and `__clang_hip_cmath.h` then cannot declare their `__device__` versions (114 errors, 2026-09-25). Each overlay header renamed those six names around an `#include_next` of the untouched original, so device code used MSVC's versions; TheRock 10.1 needs none of it. `-Xclang -fno-cuda-host-device-constexpr` fixes these errors but breaks device code, because `std::integral_constant`'s conversion becomes host-only. A VS 2022 toolset was declined in favour of the image's own toolchain. MIGraphX builds unpatched since 2.18.0: the pin carries upstream 5a80dc91ba. At `rocm-10.0` an MLIR-off build left `is_module_fusible`, `adjust_param_shapes`, `dump_mlir_to_file` and `dump_mlir_to_mxr` undefined, and `migraphx_gpu.dll` failed to link (2026-09-25); `patches\migraphx\001-mlir-off-stubs.patch` backported the fix until 2026-10-07. Every pinned archive is extracted by `Expand-PinnedArchive`, which skips the symlinks that 7-Zip 25+ refuses as dangerous: any link whose target climbs with `..`, even one that stays inside the tree. The flatbuffers seed carries nine (Java test dirs, `ts/package.json`), and no CMake file reads them. Any other 7-Zip error still fails the stage (2026-09-25). It also deletes the intermediate `.tar` of a `.tar.gz`. The EP's own source leaves `onnxruntime-ep-amdgpu.tar`, which is an ONNX Runtime archive by G2's name rule, so the EP's verify phase refused it. - **Isolation.** The deps configure with the rocm-lane `CMAKE_IGNORE_PREFIX_PATH`. MIGraphX and the EP pass `-AllowRocmPrefix`, because they need hip, MIOpen, rocBLAS, hipBLASLt and hiprtc from TheRock. - **Chain ORT only.** The EP build ends in the ORT build gate (`Assert-ChainOrtOnly -Consumer amdgpu-ep`, phase 5), which reads its configure log, `CMakeCache.txt` and `build.ninja` and writes the stamp the image census requires ([`onnxruntime-single-source.md`](onnxruntime-single-source.md)). - **Measured 2026-10-07, 2.18.0 on TheRock 10.1.0**, in a `:winamd64` container with the tarball installed by `Install-Rocm.ps1`: the build ran in 58 min (host deps 128.5 s, configure 76 s, build and install 3260 s), and `migraphx_gpu.dll` linked unpatched with MLIR off. `src/targets/gpu/prefuse_ops.cpp`, a host-only file unchanged since `rocm-10.0`, took about 50 min of that on one core at `-O3` with AMD clang 24.0.0git, so it is the stage's critical path. At `-O0` it still ran past 11 min, so the cost is in the front end, not the optimizer. The EP then built against it in 7 min 19 s, and `rocm-checks\MIGraphX.ps1` found nothing: `migraphx-ep.dll` loads GPU-less and exports `CreateEpFactories`. Smoke (`windows/scripts/build/rocm-checks/MIGraphX.ps1`, GPU-less) checks: - both trees are complete; - every static import of a MIGraphX-built PE resolves beside it, in `HIP_PATH\bin` or in System32; - the TheRock HIP runtime set sits beside the EP, byte-identical; - `amdhip64` is only delay-imported; - `migraphx-ep.dll` loads (`LoadLibraryEx`) and exports `CreateEpFactories` without starting HIP. It does **not** call `RegisterExecutionProviderLibrary`. The EP's `GetSupportedDevices` calls `hipGetDeviceCount()` first, so registration needs a HIP device ([mgx_factory.cc](https://github.com/onnxruntime/onnxruntime-ep-amdgpu/blob/99ab5cb43caa421e0b19870fa4ce9323117e5e56/src/migraphx/mgx_factory.cc)), and Windows containers get DirectX only. Limits: - **Run it on the bare host.** Put `%HIP_PATH%\bin` on PATH, or copy the TheRock tree. The EP directory holds the load closure; MIOpen, rocBLAS, hipBLASLt and `rocm_kpack.dll` with their kernel payloads stay in TheRock's `bin`. - **Prove it against CPU output.** Compare against ORT's CPU EP (TheRock#8379: zeros on gfx1201). - **Redistribution is open.** The bundled `amdhip64_7.dll` carries the same open redistribution question as the rest of the TheRock tree (see [Redistribution](#redistribution)). - **GenAI is not wired.** onnxruntime-genai's AMDGPU EP (v0.16.0+) needs the `amdgpu-ep.dll` umbrella (`-DUSE_AMDGPU=ON`, which renames the MIGraphX backend to `migraphx-backend.dll`). It is not built here. Evidence: - [AMDMIGraphX CMakeLists/requirements/windows.yaml at the pinned 2.18.0 commit](https://github.com/ROCm/AMDMIGraphX/tree/95672916ed289d4be9d230a9be732e1bbad97e8c); its `requirements.txt` differs from `rocm-10.0`'s only in the rocMLIR line, which an MLIR-off build never reads - [onnxruntime-ep-amdgpu src/CMakeLists.txt@99ab5cb](https://github.com/onnxruntime/onnxruntime-ep-amdgpu/blob/99ab5cb43caa421e0b19870fa4ce9323117e5e56/src/CMakeLists.txt) - [src/migraphx/CMakeLists.txt (/DELAYLOAD, POST_BUILD closure)](https://github.com/onnxruntime/onnxruntime-ep-amdgpu/blob/99ab5cb43caa421e0b19870fa4ce9323117e5e56/src/migraphx/CMakeLists.txt) - [src/shared/CMakeLists.txt (HIP DLL list without amd_comgr.dll)](https://github.com/onnxruntime/onnxruntime-ep-amdgpu/blob/99ab5cb43caa421e0b19870fa4ce9323117e5e56/src/shared/CMakeLists.txt) - [ORT v1.31.0 CMake package install](https://github.com/microsoft/onnxruntime/blob/v1.31.0/cmake/CMakeLists.txt) - [TheRock ml-libs MIOPEN_USE_MLIR=OFF](https://github.com/ROCm/TheRock/blob/therock-10.1/ml-libs/CMakeLists.txt) - [AMDMIGraphX#4252](https://github.com/ROCm/AMDMIGraphX/issues/4252) ## Redistribution The TheRock tree ships licence notices for its libraries under `share\doc`: MIT (rocBLAS, hipBLASLt, MIOpen, rocFFT, rocRAND, rocSPARSE, hipcc, ...), BSD (rocSOLVER, hipCUB) and Apache-2.0 (rocThrust; LLVM with the LLVM exception). The HIP runtime (`amdhip64_7.dll`, `hiprtc*`) and the OpenCL runtime (`OpenCL.dll`, `amdocl64.dll`) ship **without** a licence file, and the Windows HIP runtime links AMD's prebuilt PAL. The PyTorch wheels carry copies of the same runtime; neither llama.cpp directory carries one (since b11472 the HIP build loads the image's). **The owner cleared that question: `:winamd64-rocm` was published on 2026-09-28** (image built at ad08bc30), with the runtime shipped as TheRock delivers it. `docs/deps/deps.json` keeps `LicenseRef-Proprietary-EULA` for the runtime: the id describes the licence, not the permission. The Vulkan loader is LunarG's signed build of the Khronos Vulkan-Loader, Apache-2.0 with MIT parts; its `VulkanRT-License.txt` ships in `C:\vulkan-loader`. The DXC `dxcompiler.dll` beside the chain ORT carries Microsoft's licence terms (`LICENSE-MS.txt`) next to NCSA, the same review question as the LiteRT-LM copy. `ai-edge-litert`'s wheel ships no licence or NOTICE text of its own. ## Open points - **HIP inside a Windows container** is unproven and cannot be measured on this host (build skew). GStreamer's `gsthip` cannot load in any Server Core container on any lane at all: it links `gstgl`, which imports `OPENGL32.dll`. - **GPU correctness:** upstream TheRock#8379 reports torch+ROCm returning zeros on the RX 9070 XT. Compare every GPU path against its CPU result on the bare host. - **The spikes build, but none has run.** By 2026-09-25 all three built in the image and passed their GPU-less smoke checks: MIGraphX with `MLIR=OFF`, `BUILD_DEV=OFF`, gfx120X and TheRock 10 (a configuration no upstream CI builds; one backported patch until 2.18.0), TVM's ROCm runtime on its minimal LLVM with AMDGPU, and Dawn and Tint under the image's clang-cl. No GPU has executed any of them. - **The 2026-09-23 additions have run in a container.** By 2026-09-25 the OpenCL ICD registration, the Vulkan loader, FFmpeg's Vulkan build, the llama.cpp Vulkan zip, gfx1200 and `ai-edge-litert` were in the rocm image, and their GPU-less checks passed (`GpuLoaders.ps1`, `FFmpeg.ps1`, `LlamaCpp.ps1`, `Torch.ps1`). Each section says what its check cannot see. - **Found on every lane, not fixed here (owner decision):** LiteRT-LM is cloned by tag, not by commit. OpenCV's configure-time ORT download, listed here until 2026-09-23, was a Windows-lane defect (Linux pre-set `HAVE_ONNXRUNTIME` all along) and is fixed on every Windows lane ([`onnxruntime-single-source.md`](onnxruntime-single-source.md)).