hipdnn-superbuild

# hipDNN Superbuild Use this skill when the user asks to configure or build hipDNN through the rocm-libraries repository superbuild. It builds only; use `hipdnn-superbuild-test` for tests after a successful build. ## Inputs Infer options from the user request: - **Preset**: default `hipdnn-providers` - **Clean rebuild**: remove the build directory before configuring only when the user asks for a clean build and the active host policy permits deletion - **ROCm path**: optional `ROCM_PATH=<path>` override; Linux defaults to `/opt/rocm`. On Windows, when omitted it is derived from the wheel venv (`<venv>/Lib/site-packages/_rocm_sdk_devel`, venv default `D:/develop/latest_wheels`) - **Clang path**: optional Windows `CLANG_PATH=<path>` override; default `D:/develop/dist/clang/bin`. Clang is a prerequisite and is not provisioned; install it via `projects/hipdnn/scripts/windows/windows_build_setup.ps1` (repo-relative) if missing - **GPU targets**: optional `GPU_TARGETS=<arch>` override; Windows wheel setup defaults to `gfx1151` - **Wheel SHA**: optional Windows `SHA=<commit>` to install pinned S3 staging wheels instead of nightlies - **Provision mode**: Windows `--provision auto|always|never`; default `auto` provisions (creates the venv and pip-installs the ROCm SDK wheels) only when the SDK is missing, `always` forces a fresh wheel pull, `never` validates existing paths only - **Jobs**: optional explicit parallelism only when the user requests it and active workspace instructions permit it; otherwise let Ninja auto-detect ## Presets Read `CMakePresets.json` from the repository root if exact preset contents matter. Common hipDNN presets: | Preset | Components | |--------|------------| | `hipdnn` | hipDNN only | | `hipdnn-integration-tests` | hipDNN plus integration tests | | `hipdnn-providers` | hipDNN, miopen-provider, hipblaslt-provider, integration tests | | `hipdnn-providers-all` | All providers, including unsupported providers | | `miopen-provider` | hipDNN, miopen-provider, integration tests | | `hipblaslt-provider` | hipDNN, hipblaslt-provider, integration tests | | `hip-kernel-provider` | hipDNN, hip-kernel-provider, integration tests | | `hipdnn-samples` | hipDNN, supported providers, integration tests, samples | ## Workflow Placeholder used in the command blocks below: `<rocm-bin>` — the ROCm `bin` directory, on Windows the wheel venv's `_rocm_sdk_devel/bin`. 1. Determine the repository root: ```bash git rev-parse --show-toplevel ``` 2. Choose the build and log locations: - First honor any active workspace or repository instructions for artifact directories and build output safety. - If no such instructions exist, use `BUILD_DIR=<repo-root>/build`. - Keep full configure/build output in a log file and show only a short tail on failure. 3. Locate this skill's helper directory. Skills are host-level, not tied to a repo checkout — **default to the scripts bundled with the skill you were invoked from** (`<skill-directory>/scripts`), even when you are working inside a repo or worktree. Do NOT run the `<repo-root>/projects/hipdnn/tools/ai/skills/hipdnn-superbuild/scripts` copy just because a checkout is present: that copy can be a stale stub (on `develop`) or an unmerged in-progress version (on a feature branch). Use the source-checkout copy only when you are actively developing this skill itself and intend to exercise your in-progress edits, or when the invoked skill has no bundled `scripts/` directory. 4. Resolve ROCm and Clang paths (Windows also provisions the ROCm SDK wheels when missing): ```bash python3 <scripts>/windows_rocm_setup.py --repo-root <repo-root> [--venv-path <path>] [--rocm-path <path>] [--clang-path <path>] [--gpu-targets <arch>] [--sha <commit>] [--provision auto|always|never] ``` On Linux this echoes only provided overrides. On Windows it validates the wheel-based ROCm install and, when the SDK is absent (or `--provision always`), creates the venv and pip-installs the ROCm SDK wheels before printing `KEY=VALUE` lines on stdout. Progress goes to stderr, so stdout carries only the `ROCM_PATH=`/`CLANG_PATH=`/`GPU_TARGETS=` lines. Clang is a prerequisite and is not provisioned; a missing clang is reported as an error. 5. If a clean rebuild was requested, remove the selected build directory using the active host's normal approval/safety flow. 6. Configure from the repository root. Always bind the preset configure to the selected build directory so configure and build operate on the same tree: ```bash cmake --preset <preset> -B <build-dir> [extra -D options] ``` Add `-DROCM_PATH=<path>` when a ROCm path is resolved or provided. On Windows also add `-DCMAKE_PROGRAM_PATH=<clang-path>` and `-DGPU_TARGETS=<arch>`. **Generic-kernel-ingestor / rocKE builds** need flags no preset sets: | Flag | Default | Needed when | |---|---|---| | `HIPDNN_ENABLE_KERNEL_INGESTOR` | OFF | Any descriptor-backed engine. Also gates `hipdnn_validate_descriptors`, which is why that binary is usually absent. | | `HIPDNN_ENABLE_SDPA` | OFF | Any attention graph. This is the **frontend**: with it off the SDPA API is `#ifdef`-compiled out and plans silently DECLINE. Must be ON for both the SDK and the provider. | | `ENABLE_ASM_SDPA_ENGINE` | ON | Controls the incumbent ASM engine; disabling it is not proof that the intended new engine serves a graph. | | `HIPKERNELPROVIDER_ENABLE_ROCKE` | OFF | Packing any rocKE bundle. With `HIPDNN_ENABLE_KERNEL_INGESTOR` ON and this OFF, the build still configures and packs with the hip producer alone and needs no rocKE wheel, pip or comgr; the `rocKE/` family folder is excluded from every root and any other `rocke` UKD is pruned like an arch-pruned one. With it ON, rocKE is resolved for every descriptor root, so an unresolvable comgr is fatal at configure. | | `HIPKERNELPROVIDER_PRODUCTION_SOURCE_ROOT` | the in-tree `.../kernel_ingestor_engine/descriptors` | `CACHE PATH` naming the authored tree production packaging packages from. Every root, production or test, that is empty or has nothing to pack for the selected architectures under the build's filters is dormant: skipped at pack, its stale output tree removed, not an error. The in-tree root carries the rocKE `gfx950_attention_dense` bundle. Set but not a directory is fatal. | | `HIPKERNELPROVIDER_KPACK_PYTHON_DIR` | unset | Directory **containing** `rocm_kpack/`; this locates a package, not a compiler interpreter. | | `Python3_EXECUTABLE` | system | Explicit environment for packaging dependencies such as `msgpack` and `zstandard`; production compilation retains its selected hermetic wheel interpreter. | These flags only do anything on a preset that actually builds hip-kernel-provider. The default `hipdnn-providers` preset does **not** include it; the presets that do are `hipdnn-providers-all`, `hip-kernel-provider`, `hipdnn-dev-all` and `miopen-hipdnn-dev-all`. **There is no per-producer production switch.** Producer selection is per-UKD on `kernel_source.kind`, so one source root feeds every producer and the descriptors under the root decide what gets built. With `HIPKERNELPROVIDER_ENABLE_ROCKE=ON`, rocKE is resolved once for *every* root, test roots included, so an unresolvable comgr is fatal at configure even in a hip-only build; `HIPKERNELPROVIDER_ROCKE_COMGR_LIB` names an explicit `libamd_comgr` where a System32 copy would otherwise shadow the ROCm one. With it OFF, no root resolves comgr or runs the rocKE producer. For an ingestor create/extend task, [the ingestor RUNBOOK](../hipdnn-ingestor-engine/RUNBOOK.md) owns the full sequence. Early device/workspace feasibility has no installation requirement; installed probing follows build and installation. Build production packaging as well as provider, validator and applicable tests; a plugin build alone does not show that current descriptors were packed. Declarations travel in UKD `provenance.specialization_contract`. Only the producing compiler writes `provenance.effective_spec`, distinct from authored `provenance.spec`; generic generation is toolchain-free. No packaging `--profile`, CMake `PROFILES` or external root manifest is part of this interface. Read the packaging reference at `dnn-providers/hip-kernel-provider/descriptor-packaging/README.md`, resolved against the `<repo-root>` from step 1 rather than this skill's own directory — an installed skill is copied without the tree above it. A build is not compiler-agreement, native-registration or numerical evidence by itself; the RUNBOOK requires those observations against the final installation. 7. Build with output redirected to a log: ```bash cmake --build <build-dir> > <log> 2>&1 ``` If explicit jobs are allowed and requested, pass them through to CMake/Ninja. On failure, report the log path and tail the last relevant lines. 8. If the build fails with a stale CMake cache error such as `does not match the source`, clean the selected build directory once, reconfigure with the same `-B <build-dir>` command, and retry once. Do not loop. 9. On Windows, always stage the wheel's System32-shadowed DLLs (`amd_comgr.dll` and the HIP runtime `amdhip64_<N>.dll`) app-local into `<build-dir>/bin` after a successful build: ```bash python3 <scripts>/stage_shadowed_dlls.py --rocm-bin <rocm-bin> --build-dir <build-dir> --verbose ``` The AMD driver leaves old copies of `amd_comgr.dll` and `amdhip64_<N>.dll` in `C:\Windows\System32` that outrank the wheel's copies on PATH. Stale comgr makes MIOpen fail to JIT-build kernels at runtime (GCN-assembly Winograd solvers are the common example, but the mismatch is not limited to them); a stale HIP runtime makes the wheel's rocBLAS fault with an access violation (`SEH exception with code 0xc0000005`) in MIOpen's GEMM conv solvers, after which the test process can hang. Do this on every Windows build rather than only when a specific kernel path is expected. The Win32 loader checks the executable's own directory before System32, so an app-local copy in `<build-dir>/bin` wins; PATH manipulation alone cannot. The helper compares each wheel DLL's PE version against any already-staged copy and **skips the copy when the versions match** (content-hash fallback when version metadata is absent), so it is cheap to re-run. This step is a no-op on Linux. The build's `stage_shadowed_rocm_dlls` target and the test runner (`cmake_run.py`) stage the same DLLs on their own as well, so this build step is belt-and-suspenders that makes the app-local copies present immediately after build. ## Report Summarize: - Preset used and components expected from that preset - Build result - Build directory and log path - Windows ROCm, Clang, and GPU target values when applicable - Next step: run `hipdnn-superbuild-test` if tests are needed ## Notes - `scripts/windows_rocm_setup.py` and `scripts/stage_shadowed_dlls.py` are bundled in this skill so linked and copied installs work independently. `windows_rocm_setup.py`'s Windows wheel-provisioning logic is a Python port of `projects/hipdnn/scripts/windows/wheel_build_setup.ps1`; that PowerShell script is available for interactive users. - `stage_shadowed_dlls.py` only does work on Windows; it stages the wheel's `amd_comgr.dll` and `amdhip64_<N>.dll` app-local and emits a diagnostic for each one also present in `C:\Windows\System32` (those copies shadow PATH and are why the app-local copies are needed). - The build's `stage_shadowed_rocm_dlls` target (`projects/hipdnn/cmake/WindowsDllStaging.cmake` and `dnn-providers/cmake/WindowsDllStaging.cmake`) is the primary mechanism for app-local staging. `stage_shadowed_dlls.py` is kept on purpose rather than as a leftover: it covers build trees configured before that target existed, and a newly discovered System32-shadowed DLL can be added to the script right away, ahead of the matching CMake change. When you add a DLL to one, add it to the other. - Missing provider dependencies such as MIOpen or hipBLASLt still need to be installed or available through the selected ROCm environment. - Product test execution is intentionally out of scope for this skill.