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). 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). 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). 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=<ROCm tree> 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). 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-<family>-<release>.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/<v>/windows/<zip>.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 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).

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 <atlbase.h> 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.

  • <ort>\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.

  • <ort>\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 <ort>\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<MM><mm>.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<MMmm>.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 <plugin> 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.

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-<AMF_HEADERS_VERSION>.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<src>/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 <AMF/core/Factory.h>; 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<VULKAN_SDK>/Include --glslc=<VULKAN_SDK>/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-<major>.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/Makefiles 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).

Choosing the device. The host exposes a dGPU and an iGPU, so select one with OPENCV_OPENCL_DEVICE=<platform>:<type>:<name>, for example :GPU:gfx1201. OPENCV_OPENCL_RUNTIME=<path> 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.

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). 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). 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<release> (+rocm10.1.0 today) and PYTORCH_EXTRA_INSTALL_REQUIREMENTS=rocm[libraries]==<release>.

    • 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_<NAME>_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 <url> | 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<release> 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[_<GFX>] 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_<gfx>.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\<sha256>\<file>) 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. 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<release> 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-<gfx> 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.

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<build>-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).

  • 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 <the build's output>). 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).

  • 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, 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). 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(<name>, 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 <prefix>/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), at first passed by MIGraphX’s own -isystem (Write-HipMsvcCmathOverlay). Under clang, MSVC 14.51’s <cmath> 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).

  • 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), 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).

  • 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:

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).