diff --git a/.github/skip-ci-config.toml b/.github/skip-ci-config.toml index 68b2edf827b6..52ba0aa89364 100644 --- a/.github/skip-ci-config.toml +++ b/.github/skip-ci-config.toml @@ -110,6 +110,15 @@ common = [ # Changes to experimental code do not run standard build/test workflows. "experimental/*", + + # HRR's own CI and the HRR paths TheRock never reads. CLR compiles against + # projects/hrr/include, and TheRock #8965 adds hrr-playback (CMakeLists.txt, + # include/ and playback/, with HRR_BUILD_TESTS=OFF). Remove + # projects/hrr/tests/* once TheRock builds the HRR test suites. + ".github/workflows/hrr-*.yml", + "projects/hrr/tests/*", + "projects/hrr/tools/*", + "projects/hrr/skills/*", ] # Patterns that skip CI only on Linux (e.g., Windows-only projects) diff --git a/.github/workflows/hrr-ci.yml b/.github/workflows/hrr-ci.yml new file mode 100644 index 000000000000..4099414f4242 --- /dev/null +++ b/.github/workflows/hrr-ci.yml @@ -0,0 +1,444 @@ +# HRR CI: API-ID and workload checks, HRR script tests and the host unit suite, +# plus the GPU capture/replay suite (hrr-gpu.yml). +# projects/hrr/tests/config/ci.yaml and hrr_ci_configure.py decide what runs. +# +# To run the capture/GPU suite on a pull request, add the `ci:hrr-capture` label. +# +# Interim: once TheRock CI builds and tests HRR (TheRock PR #8965 adds the +# core-hrr artifact), the build and test jobs here are expected to go. This +# workflow would keep only the checks TheRock does not run (API-ID and workload +# checks, helper-script tests) and the rocJitsu simulator jobs until TheRock has +# a capture variant. Not a committed plan. + +name: HRR CI + +# The paths filters are the union of paths.cpu, paths.capture and paths.suite in +# ci.yaml, less paths.excluded; test_hrr_ci_configure.py fails if they differ. +# `labeled` stays in the types so the label can start a run; the summary job +# reports the run that tested the commit, so an irrelevant label cannot mask a +# failure (see `pass_through_ok`). +on: + pull_request: + types: [opened, synchronize, reopened, labeled] + paths: + - ".github/workflows/hrr-ci.yml" + - ".github/workflows/hrr-gpu.yml" + - "projects/clr/hipamd/include/hip/amd_detail/hip_api_trace.hpp" + - "projects/clr/hipamd/src/amdhip.def.in" + - "projects/clr/hipamd/src/hip_code_object.cpp" + - "projects/clr/hipamd/src/hip_code_object.hpp" + - "projects/clr/hipamd/src/hip_context.cpp" + - "projects/clr/hipamd/src/hrr/**" + - "projects/hip/include/hip/hip_runtime_api.h" + - "projects/hrr/**" + - "!projects/hrr/README.md" + - "!projects/hrr/skills/**" + push: + branches: [develop] + paths: + - ".github/workflows/hrr-ci.yml" + - ".github/workflows/hrr-gpu.yml" + - "projects/clr/hipamd/include/hip/amd_detail/hip_api_trace.hpp" + - "projects/clr/hipamd/src/amdhip.def.in" + - "projects/clr/hipamd/src/hip_code_object.cpp" + - "projects/clr/hipamd/src/hip_code_object.hpp" + - "projects/clr/hipamd/src/hip_context.cpp" + - "projects/clr/hipamd/src/hrr/**" + - "projects/hip/include/hip/hip_runtime_api.h" + - "projects/hrr/**" + - "!projects/hrr/README.md" + - "!projects/hrr/skills/**" + schedule: + # Nightly: capture-enabled build and the rocJitsu simulator suite on + # simulator.default_families. Real GPUs run only on workflow_dispatch. + - cron: "0 3 * * *" + workflow_dispatch: + inputs: + families: + description: "Comma-separated GPU families (TheRock keys, e.g. gfx94x,gfx110x). Empty: the families ci.yaml selects." + type: string + default: "" + backend: + description: "simulator runs rocJitsu on ubuntu-24.04. gpu runs the real runners." + type: choice + options: [simulator, gpu] + default: simulator + ignore_test_config: + description: "Run the cases the suite config skips" + type: boolean + default: false + +permissions: + contents: read + +# Adding a label that cannot change what runs (label_matters() in +# hrr_ci_configure.py makes the same test) gets a group of its own, so it does +# not cancel the run in progress; configure then runs nothing for it. +concurrency: + group: >- + ${{ github.workflow }}-${{ github.event_name }}-${{ github.event.pull_request.number || (github.event_name == 'push' && github.sha) || github.ref }}${{ + github.event.action == 'labeled' + && github.event.label.name != 'ci:hrr-capture' + && format('-label-{0}', github.event.label.name) || '' }} + cancel-in-progress: ${{ github.event_name == 'pull_request' }} + +env: + # The unit suite only needs HIP headers and the hip CMake package; ci.yaml + # pins the version. + ROCM_INDEX_URL: "https://rocm.nightlies.amd.com/whl-multi-arch/" + +jobs: + configure: + name: Configure + runs-on: ubuntu-24.04 + timeout-minutes: 10 + permissions: + contents: read + pull-requests: read + outputs: + run_cpu: ${{ steps.decide.outputs.run_cpu }} + run_gpu: ${{ steps.decide.outputs.run_gpu }} + families: ${{ steps.decide.outputs.families }} + simulator_families: ${{ steps.decide.outputs.simulator_families }} + gpu_families: ${{ steps.decide.outputs.gpu_families }} + skipped_families: ${{ steps.decide.outputs.skipped_families }} + reason: ${{ steps.decide.outputs.reason }} + rocm_sdk_version: ${{ steps.decide.outputs.rocm_sdk_version }} + steps: + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + sparse-checkout: | + projects/hrr/tests/scripts/hrr_ci_configure.py + projects/hrr/tests/config/ci.yaml + projects/hrr/tests/config/integration.yaml + projects/hrr/requirements.txt + sparse-checkout-cone-mode: false + + - name: Checkout TheRock family matrix + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + repository: ROCm/TheRock + # Keep in sync with the TheRock ref in therock-multi-arch-ci.yml. + ref: 488e66e7c2220df926c2762e9b97a50becfb4239 # 2026-10-21 + path: therock + sparse-checkout: build_tools/github_actions + + # Current runner labels, as TheRock's setup_multi_arch.yml loads them. + - name: Checkout TheRock CI runner config + id: ci_config + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + continue-on-error: true + with: + persist-credentials: false + repository: ROCm/therock-ci-config + ref: main + path: ci-config + + - name: Warn when runner config is unavailable + if: ${{ steps.ci_config.outcome == 'failure' }} + run: echo "::warning::could not check out ROCm/therock-ci-config; runner labels may be out of date" + + - uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0 + with: + python-version: "3.12" + + - name: Install PyYAML + run: python -m pip install --no-cache-dir -r projects/hrr/requirements.txt + + - name: Decide + id: decide + env: + GH_TOKEN: ${{ github.token }} + CI_CONFIG_PATH: ci-config + EVENT: ${{ github.event_name }} + REPO: ${{ github.repository }} + PR_NUMBER: ${{ github.event.pull_request.number }} + BEFORE: ${{ github.event.before }} + AFTER: ${{ github.sha }} + LABELS: ${{ toJSON(github.event.pull_request.labels.*.name) }} + ACTION: ${{ github.event.action }} + LABEL: ${{ github.event.label.name }} + # Empty when the head repository was deleted, which counts as a fork. + HEAD_REPO: ${{ github.event.pull_request.head.repo.full_name }} + FAMILIES: ${{ inputs.families }} + BACKEND: ${{ inputs.backend }} + run: >- + python projects/hrr/tests/scripts/hrr_ci_configure.py + --event "${EVENT}" --repo "${REPO}" --pr-number "${PR_NUMBER}" + --before "${BEFORE}" --after "${AFTER}" --labels "${LABELS}" + --action "${ACTION}" --label "${LABEL}" --head-repo "${HEAD_REPO}" --families "${FAMILIES}" + --backend "${BACKEND}" + --therock-dir therock + + api-checks: + name: API IDs and workloads + needs: configure + if: ${{ needs.configure.outputs.run_cpu == 'true' }} + runs-on: ubuntu-24.04 + timeout-minutes: 15 + steps: + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + # The API-ID check reads hrr_api_args.h at the base commit. + fetch-depth: 0 + filter: blob:none + sparse-checkout: | + projects/hrr + projects/hip/include + projects/clr/hipamd/include + projects/clr/hipamd/src/hrr + + - uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0 + with: + python-version: "3.12" + + - name: Install PyYAML + run: python -m pip install --no-cache-dir -r projects/hrr/requirements.txt + + - name: Check API IDs and the API matrix workloads + env: + BASE_SHA: ${{ github.event.pull_request.base.sha || github.event.before }} + run: python projects/hrr/tools/hrr_check_sources.py all --base-sha "${BASE_SHA}" + + scripts: + name: HRR script tests + needs: configure + if: ${{ needs.configure.outputs.run_cpu == 'true' }} + runs-on: ubuntu-24.04 + timeout-minutes: 15 + steps: + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + # Cone mode also checks out projects/clr/hipamd/src/amdhip.def.in, + # which gen_hrr_api_args.py reads on import. A test reads hrr-ci.yml. + # The generator-rule tests in tools/tests regenerate the handlers from + # the HIP headers. A configure test checks the rocJitsu config files exist. + sparse-checkout: | + projects/hrr + projects/hip/include + projects/clr/hipamd/include + projects/clr/hipamd/src/hrr + .github/workflows + emulation/rocjitsu/configs + + - uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0 + with: + python-version: "3.12" + + - name: Install test requirements + run: python -m pip install --no-cache-dir -r projects/hrr/requirements-test.txt + + - name: Lint the suite configs + run: >- + python projects/hrr/tests/scripts/hrr_test_config.py lint + projects/hrr/tests/config/unit.yaml + projects/hrr/tests/config/handler.yaml + projects/hrr/tests/config/integration.yaml + + - name: Run tests + run: python -m pytest projects/hrr/tests/scripts projects/hrr/tools/tests + + unit: + name: Unit (${{ matrix.os }}) + needs: configure + if: ${{ needs.configure.outputs.run_cpu == 'true' }} + runs-on: ${{ matrix.os }} + timeout-minutes: 30 + strategy: + fail-fast: false + matrix: + include: + - os: ubuntu-24.04 + exe: hrr-unit-tests + linker_flags: "" + - os: windows-2022 + exe: hrr-unit-tests.exe + # The MinGW runtime DLLs are not on PATH in later steps. + linker_flags: "-static" + defaults: + run: + shell: bash + steps: + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + sparse-checkout: | + projects/hrr + projects/hip/include + projects/clr/hipamd/include + + - uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0 + with: + python-version: "3.12" + + - name: Install the ROCm SDK + env: + ROCM_SDK_VERSION: ${{ needs.configure.outputs.rocm_sdk_version }} + run: | + python -m pip install --no-cache-dir -r projects/hrr/requirements.txt + python -m pip install --no-cache-dir --index-url "${ROCM_INDEX_URL}" "rocm[devel]==${ROCM_SDK_VERSION}" + rocm-sdk init + echo "ROCM_SDK=$(rocm-sdk path --root)" >> "${GITHUB_ENV}" + rocm-sdk path --bin >> "${GITHUB_PATH}" + + # This tree's HIP headers come first: an API added here is not in the + # pinned SDK's headers yet. The SDK supplies the hip CMake package. + # On Windows, cygpath -m gives the drive-letter form a native cmake accepts. + - name: Configure + env: + LINKER_FLAGS: ${{ matrix.linker_flags }} + run: | + tree="$(cygpath -m "${PWD}" 2>/dev/null || echo "${PWD}")" + cmake -S projects/hrr -B build -GNinja -Wno-dev \ + -DCMAKE_BUILD_TYPE=Release \ + "-DCMAKE_PREFIX_PATH=${ROCM_SDK}" \ + "-DROCM_PATH=${ROCM_SDK}" \ + "-DHRR_HIP_INCLUDE_DIRS=${tree}/projects/hip/include;${tree}/projects/clr/hipamd/include" \ + "-DCMAKE_EXE_LINKER_FLAGS=${LINKER_FLAGS}" \ + -DHRR_BUILD_PLAYBACK=ON \ + -DHRR_BUILD_TESTS=ON \ + -DHRR_BUILD_INTEGRATION_TESTS=OFF + + - name: Build + run: cmake --build build -j + + # A case compiled out by an #if leaves no failed or skipped result, so + # compare the compiled case list with the suite config first. + - name: Check the compiled case list + env: + EXE: ${{ matrix.exe }} + run: >- + python projects/hrr/tests/scripts/hrr_test_config.py check + --binary "build/tests/unit/${EXE}" + --config projects/hrr/tests/config/unit.yaml + + - name: Run unit suite + env: + EXE: ${{ matrix.exe }} + OS: ${{ matrix.os }} + HRR_TEST_IGNORE_CONFIG: ${{ inputs.ignore_test_config && '1' || '' }} + run: >- + python projects/hrr/tests/scripts/run_catch2.py + --xml "unit-${OS}.xml" + --config projects/hrr/tests/config/unit.yaml + --arch none + -- "build/tests/unit/${EXE}" + + # hrr-handler-tests is POSIX only (tests/unit/CMakeLists.txt), and an + # AddressSanitizer build that reaches no GPU. Its results land beside the + # unit results as an "Other" row in the summary. + - name: Check the compiled handler case list + if: ${{ runner.os == 'Linux' }} + run: >- + python projects/hrr/tests/scripts/hrr_test_config.py check + --binary build/tests/unit/hrr-handler-tests + --config projects/hrr/tests/config/handler.yaml + + - name: Run handler suite + if: ${{ runner.os == 'Linux' }} + env: + OS: ${{ matrix.os }} + HRR_TEST_IGNORE_CONFIG: ${{ inputs.ignore_test_config && '1' || '' }} + run: >- + python projects/hrr/tests/scripts/run_catch2.py + --xml "handler-${OS}.xml" + --config projects/hrr/tests/config/handler.yaml + --arch none + -- build/tests/unit/hrr-handler-tests + + - name: Upload results + if: ${{ !cancelled() }} + uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7.0.1 + with: + name: results-unit-${{ matrix.os }} + path: | + unit-${{ matrix.os }}.xml + handler-${{ matrix.os }}.xml + if-no-files-found: ignore + retention-days: 7 + + gpu: + name: GPU + needs: configure + if: ${{ needs.configure.outputs.run_gpu == 'true' }} + uses: ./.github/workflows/hrr-gpu.yml + # Note: not using 'secrets: inherit' here; no secrets are needed. + with: + simulator_families: ${{ needs.configure.outputs.simulator_families }} + gpu_families: ${{ needs.configure.outputs.gpu_families }} + ignore_test_config: ${{ inputs.ignore_test_config || false }} + + summary: + name: HRR CI Summary + needs: [configure, api-checks, scripts, unit, gpu] + # Also when configure failed: then nothing else ran, and this job fails. + if: >- + ${{ always() && (needs.configure.result != 'success' + || needs.configure.outputs.run_cpu == 'true' + || needs.configure.outputs.run_gpu == 'true' + || (github.event_name == 'pull_request' && github.event.action == 'labeled')) }} + runs-on: ubuntu-24.04 + timeout-minutes: 10 + permissions: + actions: read + contents: read + steps: + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + sparse-checkout: | + projects/hrr/tests/scripts/hrr_ci_report.py + projects/hrr/tests/scripts/summarize_junit.py + sparse-checkout-cone-mode: false + + - name: Download results + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 + with: + pattern: results-* + merge-multiple: true + path: results + + - name: Download base build record + if: ${{ needs.configure.outputs.run_gpu == 'true' }} + continue-on-error: true + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 + with: + name: hrr-base-${{ github.run_id }} + path: base + + - name: Summarize + env: + GH_TOKEN: ${{ github.token }} + CONFIGURE_RESULT: ${{ needs.configure.result }} + REASON: ${{ needs.configure.outputs.reason }} + RUN_GPU: ${{ needs.configure.outputs.run_gpu }} + FAMILIES: ${{ needs.configure.outputs.families }} + SKIPPED: ${{ needs.configure.outputs.skipped_families }} + # A labeled run that configured nothing has no HRR job of its own. + # GitHub keeps the most recent check run of a name on a commit, so + # its summary reports the run that tested this commit instead of + # skipping, which would let the label turn a failing HRR CI green. + PASS_THROUGH: >- + ${{ github.event_name == 'pull_request' && github.event.action == 'labeled' + && needs.configure.result == 'success' + && needs.configure.outputs.run_cpu == 'false' + && needs.configure.outputs.run_gpu == 'false' }} + HEAD_SHA: ${{ github.event.pull_request.head.sha || github.sha }} + NEEDS_JSON: ${{ toJSON(needs) }} + run: >- + python3 projects/hrr/tests/scripts/hrr_ci_report.py + --configure-result "${CONFIGURE_RESULT}" + --reason "${REASON}" --run-gpu "${RUN_GPU}" --families "${FAMILIES}" + --skipped-families "${SKIPPED}" + --results-dir results --base-json base/base.json + --needs-json "${NEEDS_JSON}" + --pass-through "${PASS_THROUGH}" --head-sha "${HEAD_SHA}" diff --git a/.github/workflows/hrr-gpu.yml b/.github/workflows/hrr-gpu.yml new file mode 100644 index 000000000000..116f3d047a02 --- /dev/null +++ b/.github/workflows/hrr-gpu.yml @@ -0,0 +1,458 @@ +# HRR GPU suite: build an amdhip64 with HRR capture compiled in, then run the +# HRR capture/replay integration suite on each requested GPU family. +# +# Called by hrr-ci.yml, which decides when this runs and on which families. +# +# Interim: the build job and projects/hrr/tests/scripts/hrr_capture_build.py +# rebuild CLR with capture enabled on top of TheRock release artifacts. They +# stay until TheRock can build and run the HRR integration suite. Defining the +# hrr-capture variant is not enough: nothing selects it, and it does not +# install the test binaries. + +name: HRR GPU + +on: + workflow_call: + inputs: + simulator_families: + description: JSON list of {family, runner, backend, arch, config} rows from hrr_ci_configure.py for the rocJitsu jobs. [] skips them. + type: string + default: "[]" + gpu_families: + description: Real-GPU rows in the same form. [] skips the GPU jobs. + type: string + default: "[]" + ignore_test_config: + description: Run the cases the suite config skips + type: boolean + default: false + +permissions: + contents: read + +jobs: + build: + name: Build capture-enabled HIP and HRR + runs-on: ubuntu-24.04 + timeout-minutes: 90 + container: + image: ghcr.io/rocm/therock_build_manylinux_x86_64@sha256:cf4f6d9909056906e4a1c0bc4c245658a0af4c6a6174082a86adfa2d518b7aeb # 2026-08-18 + steps: + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + sparse-checkout: | + projects/clr + projects/hip + projects/hrr + + - name: Checkout TheRock artifact tools + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + repository: ROCm/TheRock + # Keep in sync with the TheRock ref in therock-multi-arch-ci.yml. + ref: 488e66e7c2220df926c2762e9b97a50becfb4239 # 2026-10-21 + path: therock + sparse-checkout: | + build_tools + cmake/therock_amdgpu_targets.cmake + requirements.txt + sparse-checkout-cone-mode: false + + - name: Install TheRock Python requirements + run: python3 -m pip install --no-cache-dir -r therock/requirements.txt + + - name: Install base ROCm artifacts + env: + GITHUB_TOKEN: ${{ github.token }} + run: python3 projects/hrr/tests/scripts/hrr_capture_build.py fetch-base + + # No cache restore: a cache hit left librocm_kpack.so.0 out of the tree and + # Verify failed. Off until that failure is understood, so the build always + # runs. + - name: Build + run: python3 projects/hrr/tests/scripts/hrr_capture_build.py build + + - name: Verify + run: python3 projects/hrr/tests/scripts/hrr_capture_build.py verify + + - name: Upload base build record + if: ${{ !cancelled() }} + uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7.0.1 + with: + name: hrr-base-${{ github.run_id }} + path: build/base.json + if-no-files-found: error + retention-days: 7 + + - name: Upload build + uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7.0.1 + with: + name: hrr-capture-${{ github.run_id }} + path: | + build/base.json + build/capture-lib/ + build/hrr/playback/hrr-playback + build/hrr/tests/integration/hrr-integration-tests + build/hrr/tests/integration/*.so + if-no-files-found: error + retention-days: 7 + + # rocJitsu has no family or architecture argument: every simulator job would + # build the same binary, so build it once here. The simulator jobs download + # it from the artifact. The launcher is tools/rocjitsu/rocjitsu and finds + # librocjitsu.so at ../../ from its own path, so the tarball keeps the build/rocjitsu layout, and tar keeps the + # exec bit that upload-artifact would drop. + rocjitsu-build: + name: Build rocJitsu + if: inputs.simulator_families != '[]' + runs-on: ubuntu-24.04 + timeout-minutes: 60 + steps: + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + sparse-checkout: | + emulation/rocjitsu + runtimes/api-headers + shared/machine-readable-isa + + - name: Install rocJitsu build packages + shell: bash + run: | + set -euo pipefail + sudo -n apt-get update + sudo -n DEBIAN_FRONTEND=noninteractive apt-get install -y --no-install-recommends \ + build-essential ca-certificates ccache clang libdrm-dev ninja-build + + - uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0 + with: + python-version: "3.12" + + - name: Install Python requirements + shell: bash + run: | + set -euo pipefail + python -m venv "${RUNNER_TEMP}/hrr-venv" + "${RUNNER_TEMP}/hrr-venv/bin/python" -m pip install --no-cache-dir 'cmake==3.31.10' + echo "${RUNNER_TEMP}/hrr-venv/bin" >> "${GITHUB_PATH}" + + # ccache holds only the rocJitsu compile objects, in its own directory + # under RUNNER_TEMP. It never covers the ROCm install in the build job, + # which stays uncached (see the note above that job's Build step). + # ccache keys each object on the preprocessed source and the compiler, + # so a stale hit is not possible; the key below only decides when a new + # cache is saved. Pull requests restore the cache that develop pushes and + # the schedule saved, and save their own, which only that PR reads. + - name: Configure ccache + id: ccache + shell: bash + run: | + set -euo pipefail + { + echo "CCACHE_DIR=${RUNNER_TEMP}/ccache-rocjitsu" + echo "CCACHE_BASEDIR=${GITHUB_WORKSPACE}" + echo "CCACHE_COMPILERCHECK=content" + echo "CCACHE_MAXSIZE=500M" + } >> "${GITHUB_ENV}" + echo "compiler=$( (c++ --version | head -n1; cc --version | head -n1) | sha256sum | cut -c1-12)" >> "${GITHUB_OUTPUT}" + + - name: Restore ccache + uses: actions/cache@0057852bfaa89a56745cba8c7296529d2fc39830 # v4.3.0 + with: + path: ${{ runner.temp }}/ccache-rocjitsu + key: ccache-rocjitsu-${{ runner.os }}-${{ runner.arch }}-${{ steps.ccache.outputs.compiler }}-${{ hashFiles('emulation/rocjitsu/**', 'runtimes/api-headers/**', 'shared/machine-readable-isa/**') }} + restore-keys: | + ccache-rocjitsu-${{ runner.os }}-${{ runner.arch }}-${{ steps.ccache.outputs.compiler }}- + + - name: Build rocJitsu + shell: bash + run: | + set -euo pipefail + ccache --zero-stats + cmake -S emulation/rocjitsu -B build/rocjitsu -G Ninja -DCMAKE_BUILD_TYPE=Release \ + -DCMAKE_C_COMPILER_LAUNCHER=ccache -DCMAKE_CXX_COMPILER_LAUNCHER=ccache + cmake --build build/rocjitsu --target rocjitsu_bin rocjitsu_shared rocjitsu_memory -j + + - name: ccache stats + if: ${{ !cancelled() }} + shell: bash + run: ccache --show-stats --verbose || true + + - name: Package rocJitsu + shell: bash + run: | + set -euo pipefail + # The glob must expand inside build/rocjitsu, so cd there. tar keeps + # symlinks, so a versioned librocjitsu.so chain survives. + # + # librocjitsu.so execs the private rocjitsu-memory keeper. The path + # baked in for a build tree is this runner's, which the simulator + # containers do not have, so the keeper goes where the installed + # layout looks for it: ../libexec/rocjitsu/ beside the library's + # directory (build/libexec/rocjitsu/ next to build/rocjitsu/). + ls -l build/rocjitsu/tools/rocjitsu/rocjitsu build/rocjitsu/librocjitsu.so* \ + build/rocjitsu/tools/rocjitsu-memory + mkdir -p build/libexec/rocjitsu + cp build/rocjitsu/tools/rocjitsu-memory build/libexec/rocjitsu/rocjitsu-memory + (cd build && tar -cf "${GITHUB_WORKSPACE}/rocjitsu.tar" \ + rocjitsu/tools/rocjitsu/rocjitsu rocjitsu/librocjitsu.so* \ + libexec/rocjitsu/rocjitsu-memory) + tar -tvf rocjitsu.tar + + - name: Upload rocJitsu + uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7.0.1 + with: + name: rocjitsu-${{ github.run_id }} + path: rocjitsu.tar + if-no-files-found: error + retention-days: 7 + + simulator: + name: rocJitsu (${{ matrix.family }}) + needs: [build, rocjitsu-build] + if: inputs.simulator_families != '[]' + runs-on: ${{ matrix.runner }} + timeout-minutes: 45 + strategy: + fail-fast: false + matrix: + include: ${{ fromJSON(inputs.simulator_families) }} + # The same image as the real-GPU job, without devices. The build job's + # workspace path is this container's workspace, so the paths compiled into + # the test binaries resolve without a symlink on the runner. + container: + # Digest pinned 2026-10-13. + image: ghcr.io/rocm/no_rocm_image_ubuntu24_04@sha256:4150afe4759d14822f0e3f8930e1124f26e11f68b5c7b91ec9a02b20b1ebbb98 + options: --user 0:0 + steps: + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + sparse-checkout: | + emulation/rocjitsu/configs + projects/hrr/tests + + - name: Checkout TheRock artifact tools + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + repository: ROCm/TheRock + ref: 488e66e7c2220df926c2762e9b97a50becfb4239 # 2026-10-21 + path: therock + sparse-checkout: | + build_tools + cmake/therock_amdgpu_targets.cmake + requirements.txt + sparse-checkout-cone-mode: false + + # The prebuilt rocJitsu and the capture-enabled amdhip64 load libdrm at run + # time. The container runs as root, so this does not use sudo. + - name: Install libdrm + shell: bash + run: | + set -euo pipefail + apt-get update + DEBIAN_FRONTEND=noninteractive apt-get install -y --no-install-recommends \ + libdrm2 libdrm-amdgpu1 + + - uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0 + with: + python-version: "3.12" + + - name: Install Python requirements + shell: bash + run: | + set -euo pipefail + python -m venv "${RUNNER_TEMP}/hrr-venv" + "${RUNNER_TEMP}/hrr-venv/bin/python" -m pip install --no-cache-dir -r therock/requirements.txt + echo "${RUNNER_TEMP}/hrr-venv/bin" >> "${GITHUB_PATH}" + + - name: Download rocJitsu + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 + with: + name: rocjitsu-${{ github.run_id }} + path: rocjitsu-artifact + + - name: Unpack rocJitsu + shell: bash + run: | + set -euo pipefail + mkdir -p build + tar -C build -xf rocjitsu-artifact/rocjitsu.tar + ls -l build/rocjitsu/tools/rocjitsu build/rocjitsu/librocjitsu.so* \ + build/libexec/rocjitsu/rocjitsu-memory + echo "${GITHUB_WORKSPACE}/build/rocjitsu/tools/rocjitsu" >> "${GITHUB_PATH}" + + - name: Download build + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 + with: + name: hrr-capture-${{ github.run_id }} + path: build + + - name: Install base ROCm artifacts and the capture-enabled amdhip64 + env: + GITHUB_TOKEN: ${{ github.token }} + run: python projects/hrr/tests/scripts/hrr_capture_build.py install-base + + - name: Run integration suite + env: + FAMILY: ${{ matrix.family }} + ARCH: ${{ matrix.arch }} + # The config path stays data. The shell does not assemble the command. + # Local mode: the simulator runs inside each process. Daemon mode + # (rocjitsu --daemon) fails hipMemCreate and hangs gfx90a replay. + LAUNCHER: ${{ format('rocjitsu --config {0} --', matrix.config) }} + HRR_TEST_IGNORE_CONFIG: ${{ inputs.ignore_test_config && '1' || '' }} + shell: bash + run: >- + python projects/hrr/tests/scripts/run_catch2.py + --xml "integration-${FAMILY}.xml" + --timeout 1200 + --config projects/hrr/tests/config/integration.yaml + --os linux --arch "${ARCH}" + --launcher "${LAUNCHER}" + -- build/hrr/tests/integration/hrr-integration-tests + + - name: Upload results + if: ${{ !cancelled() }} + uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7.0.1 + with: + name: results-integration-${{ matrix.family }} + path: integration-${{ matrix.family }}.xml + if-no-files-found: ignore + retention-days: 7 + + test: + name: Integration (${{ matrix.family }}) + needs: build + if: inputs.gpu_families != '[]' + runs-on: ${{ matrix.runner }} + timeout-minutes: 45 + strategy: + fail-fast: false + matrix: + include: ${{ fromJSON(inputs.gpu_families) }} + # The image and options TheRock's test_component.yml uses for GPU tests. + # Running in a container also gives this job the build job's workspace + # path, which the test binaries have compiled in. + container: + image: ghcr.io/rocm/no_rocm_image_ubuntu24_04@sha256:4150afe4759d14822f0e3f8930e1124f26e11f68b5c7b91ec9a02b20b1ebbb98 + options: >- + --ipc host + --user 0:0 + --ulimit memlock=-1:-1 + --ulimit nofile=1048576:1048576 + --security-opt seccomp=unconfined + --group-add video + --device /dev/kfd + --device /dev/dri + --group-add 993 + --group-add 992 + --group-add 110 + --env-file /etc/podinfo/gha-gpu-isolation-settings + -e ROCR_VISIBLE_DEVICES + -e KUBE_CPU_REQUEST + steps: + # The pre-job cleanup in TheRock's test_component.yml: before the + # checkout below cleans the workspace, stop the GPU processes that an + # earlier job on this runner left running from its build directory. + - name: Fetch the GPU process cleanup script + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + repository: ROCm/TheRock + ref: 488e66e7c2220df926c2762e9b97a50becfb4239 # 2026-10-21 + path: prejob + sparse-checkout: build_tools/github_actions/cleanup_processes.sh + sparse-checkout-cone-mode: false + + - name: Pre-job cleanup GPU processes + timeout-minutes: 5 + run: bash prejob/build_tools/github_actions/cleanup_processes.sh + + - name: Checkout + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + sparse-checkout: | + projects/hrr/tests/config + projects/hrr/tests/scripts + + - name: Checkout TheRock artifact tools + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 + with: + persist-credentials: false + repository: ROCm/TheRock + ref: 488e66e7c2220df926c2762e9b97a50becfb4239 # 2026-10-21 + path: therock + sparse-checkout: | + build_tools + cmake/therock_amdgpu_targets.cmake + requirements.txt + sparse-checkout-cone-mode: false + + - uses: actions/setup-python@a309ff8b426b58ec0e2a45f0f869d46889d02405 # v6.2.0 + with: + python-version: "3.12" + + # Install into a virtualenv under RUNNER_TEMP, which is removed when the + # job ends. A bare "python -m pip install" writes into setup-python's + # interpreter under _work/_tool on the runner's host, which the next job + # on the same machine reuses. Later steps pick the venv up from PATH. + - name: Install Python requirements + run: | + python -m venv "${RUNNER_TEMP}/hrr-venv" + "${RUNNER_TEMP}/hrr-venv/bin/python" -m pip install --no-cache-dir -r therock/requirements.txt + echo "${RUNNER_TEMP}/hrr-venv/bin" >> "${GITHUB_PATH}" + + - name: Download build + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 + with: + name: hrr-capture-${{ github.run_id }} + path: build + + - name: Install base ROCm artifacts and the capture-enabled amdhip64 + env: + GITHUB_TOKEN: ${{ github.token }} + run: python projects/hrr/tests/scripts/hrr_capture_build.py install-base + + - name: Run integration suite + env: + FAMILY: ${{ matrix.family }} + HRR_TEST_IGNORE_CONFIG: ${{ inputs.ignore_test_config && '1' || '' }} + # A regions case faults the GPU on purpose; without this ROCr writes a + # gpucore file of about 160 MB into the workspace. + HSA_DISABLE_COREDUMP_ON_EXCEPTION: "1" + # --arch auto finds this machine's GPU with the installed rocminfo. + # github.workspace is the runner's host path, which a container need + # not have; GITHUB_WORKSPACE is the path inside the container. + run: >- + ROCM_SDK="${GITHUB_WORKSPACE}/dist" + python projects/hrr/tests/scripts/run_catch2.py + --xml "integration-${FAMILY}.xml" + --timeout 1200 + --config projects/hrr/tests/config/integration.yaml + --os linux --arch auto + -- build/hrr/tests/integration/hrr-integration-tests + + - name: Upload results + if: ${{ !cancelled() }} + uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7.0.1 + with: + name: results-integration-${{ matrix.family }} + path: integration-${{ matrix.family }}.xml + if-no-files-found: ignore + retention-days: 7 + + # The job container's init never reaps a killed process, so the script + # can report a failure after a passing suite; that must not fail the job. + - name: Post-job cleanup GPU processes + if: ${{ always() }} + continue-on-error: true + timeout-minutes: 5 + run: bash therock/build_tools/github_actions/cleanup_processes.sh diff --git a/projects/hrr/requirements-test.txt b/projects/hrr/requirements-test.txt new file mode 100644 index 000000000000..c704a979f4d7 --- /dev/null +++ b/projects/hrr/requirements-test.txt @@ -0,0 +1,3 @@ +# Adds what the HRR script tests (projects/hrr/tests/scripts, projects/hrr/tools/tests) need. +-r requirements.txt +pytest==8.4.2 diff --git a/projects/hrr/requirements.txt b/projects/hrr/requirements.txt new file mode 100644 index 000000000000..6f1a3b3dfab6 --- /dev/null +++ b/projects/hrr/requirements.txt @@ -0,0 +1,4 @@ +# Python requirements of the HRR helper scripts, the api-matrix tools and the +# CPU jobs of .github/workflows/hrr-ci.yml. hrr-gpu.yml takes PyYAML from +# TheRock's requirements.txt instead; keep both pins equal. +PyYAML==6.0.2 diff --git a/projects/hrr/tests/README.md b/projects/hrr/tests/README.md index bc2ee7fa036f..36ec1c6dfdd1 100644 --- a/projects/hrr/tests/README.md +++ b/projects/hrr/tests/README.md @@ -10,7 +10,8 @@ are the same ones it uses. tests/ unit/ C++ (Catch2). hrr-unit-tests: CPU only, no GPU, no capture runtime. integration/ C++ (Catch2). hrr-integration-tests: real capture, then replay, on a GPU. - config/ One YAML file per suite: every case, and where it runs. + config/ One YAML file per suite: every case, and where it runs. ci.yaml: + which CI jobs run and on which GPU families. include/ Headers both suites include (hrr_test_common.hh, hrr_test_process.hh). scripts/ CI helpers that run a suite and report on it, plus the unittest suites for those helpers and for ../tools. Not built. @@ -133,19 +134,63 @@ The host unit binary runs once in either command. ## Testing the Python helpers `scripts/` holds both the helpers and their tests. The tests need only Python and -PyYAML (used by the api-matrix tools), not ROCm: +PyYAML (used by the api-matrix tools), not ROCm. Install them with `python3 -m pip install -r projects/hrr/requirements-test.txt` +(`requirements.txt` has only the runtime pins): ```bash python3 -m unittest discover -s projects/hrr/tests/scripts -p 'test_*.py' -v ``` -They cover `run_catch2.py`, `summarize_junit.py`, and `hrr_test_config.py`, plus -the tools in `../tools`: the HRR API ID compatibility check, the playback-handler default -in the generator, and the [api-matrix](../tools/api-matrix/README.md) scripts. +They cover `run_catch2.py`, `summarize_junit.py`, `hrr_test_config.py`, and the +HRR CI scripts (`hrr_ci_configure.py`, `hrr_capture_build.py`, `hrr_ci_report.py`), +plus the tools in `../tools`: the HRR API ID compatibility check, the +API-ID and workload checks (`hrr_check_sources.py`), the playback-handler default in +the generator, and the [api-matrix](../tools/api-matrix/README.md) scripts. + +## Choosing what CI runs + +[`config/ci.yaml`](config/ci.yaml) is the policy the configure job of `hrr-ci.yml` +applies (`scripts/hrr_ci_configure.py`): + +- `paths.cpu`: a change under one of these runs the API check, script and unit + jobs. `paths.capture` also runs the capture-enabled build and the GPU suite. + `paths.excluded` counts for neither. The `paths` filter of `hrr-ci.yml` is the same + lists, so a pull request that touches none of them starts no HRR CI run; + `test_hrr_ci_configure.py` fails when the filter and `ci.yaml` disagree. +- `capture_label` (`ci:hrr-capture`) runs the GPU suite on a pull request that touches + an HRR path but no capture path. Adding it restarts the run; adding any other label, + including TheRock's `ci:gfx...` labels, leaves the run in progress alone. +- `backend` and `simulator`: every event except a `workflow_dispatch` that asks for + `backend: gpu` runs on the rocJitsu simulator, on the families in + `simulator.default_families`, each with its rocJitsu config from `simulator.configs`. + No pull request, push or nightly run queues on a real GPU runner. +- `families`: the GPU families as TheRock family keys. They matter only for a + `workflow_dispatch` with `backend: gpu`, which runs the `enabled` ones on the runner + TheRock's family matrix names, or the ones the `families` input names, disabled or + not. Only gfx94x is enabled today; the others are listed with `enabled: false`. + The input also replaces the default families of a simulator run. `ci:gfx...` labels + are TheRock's and do not select an HRR family. + `targets.arch` in `config/integration.yaml` stays wide on purpose: configure fails + unless it covers every family listed here, and a requested family needs its + architecture compiled into the binary. +- `rocm_sdk_version`: the ROCm SDK the unit job installs. + +An unknown key or family fails the configure job, and so does a family whose +GPU targets in TheRock's family matrix are not all in `targets.arch` of +`config/integration.yaml`. To take a pool out of a real-GPU run, set `enabled: false`. + +TheRock never reads the HRR workflows or `projects/hrr/tests`, `tools` and `skills`, +so `.github/skip-ci-config.toml` lists them: a pull request that changes only those +skips TheRock CI. CLR compiles against `projects/hrr/include`. Any other change to +`projects/hrr` builds every TheRock stage until the TheRock commit pinned by +`.github/workflows/therock-multi-arch-ci.yml` includes +[TheRock #8965](https://github.com/ROCm/TheRock/pull/8965), which adds +`hrr-playback` to TheRock's `compiler-runtime` stage on Linux. ## Configuring where a case runs -The build is one fat binary for every architecture in `HRR_ARCHS`. Nothing about a +The build is one fat binary for every architecture in `targets.arch` of +`config/integration.yaml`. Nothing about a GPU is decided at build time. CI decides at run time, from [`config/integration.yaml`](config/integration.yaml) (and [`config/unit.yaml`](config/unit.yaml) for the CPU suite). @@ -179,8 +224,9 @@ skip: ``` Direct workloads (`[hrr-direct]`, `[direct]`) are spawned by a driver and must not -carry rules. `targets.arch` is the same list as `HRR_ARCHS` in -`.github/workflows/hrr-ci.yml`; `hrr_test_config.py lint --workflow` checks that. +carry rules. The GPU build compiles the integration binary for exactly the +`targets.arch` of `config/integration.yaml`, so adding an architecture there is +what builds it. CI compares each binary's registered cases against its YAML, in both directions, because a case behind an `#if` or a missing define disappears without any failure. diff --git a/projects/hrr/tests/config/ci.yaml b/projects/hrr/tests/config/ci.yaml new file mode 100644 index 000000000000..53403328af0f --- /dev/null +++ b/projects/hrr/tests/config/ci.yaml @@ -0,0 +1,137 @@ +# What HRR CI runs and where, read by +# projects/hrr/tests/scripts/hrr_ci_configure.py (hrr-ci.yml's configure job). +# Unknown keys, triggers or families are errors. +version: 1 + +# The ROCm SDK the host unit job installs. Pinned so a run does not depend on +# the day's nightly. +rocm_sdk_version: "10.1.0a20260822" + +# Path prefixes, relative to the repository root. A path ending in "/" matches +# everything under it; any other path matches only itself. +paths: + # A change under any of these runs the API check, script and unit jobs. + # Changes under emulation/rocjitsu, runtimes/api-headers and + # shared/machine-readable-isa are not path triggers; the nightly schedule and + # workflow_dispatch cover them. + cpu: + - projects/hrr/ + - projects/clr/hipamd/src/hrr/ + - projects/clr/hipamd/include/hip/amd_detail/hip_api_trace.hpp + - projects/clr/hipamd/src/amdhip.def.in + - projects/hip/include/hip/hip_runtime_api.h + - .github/workflows/hrr-ci.yml + - .github/workflows/hrr-gpu.yml + # A change under any of these also runs the capture-enabled build and the GPU + # suite. hip_runtime_api.h is deliberately absent: almost every HIP API change + # touches it, and the script tests job (test_generated_files_current.py) + # already regenerates HRR from it and compares. + # amdhip.def.in is absent for the same reason (cpu only). + capture: + - projects/clr/hipamd/src/hrr/ + - projects/clr/hipamd/src/hip_context.cpp + - projects/clr/hipamd/src/hip_code_object.cpp + - projects/clr/hipamd/src/hip_code_object.hpp + - projects/clr/hipamd/include/hip/amd_detail/hip_api_trace.hpp + - projects/hrr/CMakeLists.txt + - projects/hrr/include/ + - projects/hrr/playback/ + - projects/hrr/tests/CMakeLists.txt + - projects/hrr/tests/include/ + - projects/hrr/tests/integration/ + - projects/hrr/tests/config/integration.yaml + - projects/hrr/tests/scripts/run_catch2.py + - projects/hrr/tests/scripts/hrr_test_config.py + - projects/hrr/tests/scripts/summarize_junit.py + - projects/hrr/tests/scripts/hrr_capture_build.py + - projects/hrr/tests/scripts/hrr_ci_configure.py + - projects/hrr/tests/scripts/hrr_ci_report.py + - .github/workflows/hrr-ci.yml + - .github/workflows/hrr-gpu.yml + # Runs the suite without changing the capture binary. + suite: + - projects/hrr/tests/config/ci.yaml + # Ignored by both lists; hrr-skills.yml covers these. + excluded: + - projects/hrr/skills/ + - projects/hrr/README.md + +# A pull request with this label runs the GPU suite even when it touches no +# capture path. It must still touch a cpu, capture or suite path: hrr-ci.yml's +# paths filter, which is those lists, starts no run otherwise. +capture_label: ci:hrr-capture + +# simulator runs rocJitsu on ubuntu-24.04. gpu runs TheRock's real runners. +# workflow_dispatch passes its backend input. Any other event uses default, so +# a pull request, a push and the nightly schedule run on the simulator and +# never queue on a real GPU runner; a person asks for one with +# workflow_dispatch (backend gpu). +backend: + default: simulator + +# Default simulator jobs. These do not consult families.*.enabled. Each row is +# a rocJitsu KMD config that exposes the family's GPU to HIP, and the whole +# integration suite passes on it, less the cases integration.yaml skips with +# "emulator: true" (each names its reason). Checked 2026-10-19 on a +# two-GPU gfx950 config and the single-GPU configs of the others. +# +# Not listed: gfx103x, gfx1150 and gfx1153, which rocJitsu has no config for. +# A workflow_dispatch families input that names one fails the configure job. +simulator: + runner: ubuntu-24.04 + default_families: [gfx90a, gfx94x, gfx950, gfx110x, gfx1151, gfx120x, gfx125x] + configs: + gfx94x: emulation/rocjitsu/configs/gfx942_cdna3_kmd.json + gfx950: emulation/rocjitsu/configs/gfx950_mi355x_kmd_2gpu.json + gfx90a: emulation/rocjitsu/configs/gfx90a_mi210_kmd.json + gfx110x: emulation/rocjitsu/configs/gfx1100_w7900.json + gfx1151: emulation/rocjitsu/configs/gfx1151.json + gfx120x: emulation/rocjitsu/configs/gfx1201_r9700.json + gfx125x: emulation/rocjitsu/configs/gfx1250_mi455x_kmd_4gpu.json + +# The GPU families the suite runs on, as TheRock family keys +# (build_tools/github_actions/amdgpu_family_matrix.py). The integration binary +# is built for targets.arch in integration.yaml; the configure job fails unless +# that covers every GPU target TheRock lists for each family here. +# +# Per family, all optional: +# enabled false: run only when the workflow_dispatch families input names +# it. Default true. ci:gfx labels belong to TheRock and are not +# read here. +# +# A family without a runner label is skipped. Every job queues on its own +# runner pool and the summary waits for all of them, so to take a pool out, +# set enabled: false. +# +# Real GPUs run only when a person starts workflow_dispatch with backend gpu. +# Then only gfx94x is enabled; the other families stay listed and disabled, so +# the families input can still request one. Delete an `enabled: false` line to +# make that family eligible for the GPU backend; it does not add a simulator +# job. Simulator membership is simulator.default_families. +# targets.arch in integration.yaml stays wide on purpose: configure fails +# unless it covers every family listed here, and a requested family needs its +# architecture compiled into the binary. +families: + gfx90a: + enabled: false + gfx94x: {} + gfx950: + enabled: false + gfx103x: + enabled: false + gfx110x: + enabled: false + gfx1150: + enabled: false + gfx1151: + enabled: false + gfx1153: + enabled: false + gfx120x: + enabled: false + # Off until understood: in two runs the MI455 runner picked the job up, but + # --arch auto found no GPU, so no case ran. TheRock also reserves this small + # pool for designated workflows. The workflow_dispatch families input still + # runs it. + gfx125x: + enabled: false diff --git a/projects/hrr/tests/config/handler.yaml b/projects/hrr/tests/config/handler.yaml new file mode 100644 index 000000000000..b33c54443abe --- /dev/null +++ b/projects/hrr/tests/config/handler.yaml @@ -0,0 +1,86 @@ +# HRR test configuration. Every case the binary registers has an entry here. +# An empty entry runs on every OS and GPU architecture. +# +# A skip rule matches when every key under "when" matches (a list is OR, a +# missing key matches anything, and a value may be a glob such as gfx11*). +# "emulator: true" matches only a run under a GPU emulator (run_catch2.py +# --launcher), "emulator: false" only a run on a real device. +# kind: disabled is a temporary regression and requires a public issue URL. +# kind: unsupported is permanent and needs only a reason. +# +# hrr-handler-tests is built only on POSIX (tests/unit/CMakeLists.txt), so this +# suite targets Linux only. +# +# Regenerate the case names from a built binary with: +# hrr_test_config.py check --binary --config --update +# Listing needs no GPU. Say so in the commit message when the set changes. +version: 1 +suite: unit +binary: hrr-handler-tests +select: "*" +targets: + os: [linux] + arch: [gfx90a, gfx942, gfx950, gfx1030, gfx1100, gfx1101, gfx1102, gfx1103, gfx1150, gfx1151, gfx1153, gfx1200, gfx1201, gfx1250] +cases: + Unit_HRR_AccessPolicyWindow_KernelNodeAttributeTranslatesBasePtr: {} + Unit_HRR_AccessPolicyWindow_StreamAttributeTranslatesBasePtr: {} + Unit_HRR_Blob_AbsentBlobDoesNotEndTheReplay: {} + Unit_HRR_Blob_CapRefusalEndsTheReplay: {} + Unit_HRR_CaptureState_EndCaptureClearsFlag: {} + Unit_HRR_GraphDeps_InlineCapacity: {} + Unit_HRR_GraphDeps_MemAllocNodeRefusesOversizeCount: {} + Unit_HRR_GraphDeps_SymbolCopyNodesRefuseOversizeCount: {} + Unit_HRR_GraphDeps_SymbolCopyNodesRefuseTruncatedList: {} + Unit_HRR_GraphDeps_TruncatedDependencyListIsRefused: {} + Unit_HRR_HostBudget_BlobCacheChargesOnce: {} + Unit_HRR_HostBudget_BlobCacheRefusesOverBudgetAndMissing: {} + Unit_HRR_HostBudget_CodeObjectCacheIsChargedSeparately: {} + Unit_HRR_HostBudget_ConcurrentLoadsOfOneBlobChargeOnce: {} + Unit_HRR_HostBudget_DeviceMemCreateNotCharged: {} + Unit_HRR_HostBudget_DispatchFreesReplacedHostAllocation: {} + Unit_HRR_HostBudget_DispatchTableUsesTheChargingHandlers: {} + Unit_HRR_HostBudget_HostAllocationHandlersRefuseOverBudget: {} + Unit_HRR_HostBudget_HostBlocksStartOnCacheLine: {} + Unit_HRR_HostBudget_HostMemCreateRefusedOverBudget: {} + Unit_HRR_HostBudget_HostRegisterAllocationFailureRefuses: {} + Unit_HRR_HostBudget_LandingBufferChargesWholeBlock: {} + Unit_HRR_HostBudget_LandingBufferPointerSurvivesGrowth: {} + Unit_HRR_HostBudget_ManagedAllocationRefusedOverBudget: {} + Unit_HRR_HostBudget_ManagedAllocationReturnsChargeOnFree: {} + Unit_HRR_HostBudget_MemCreateChargeHeldWhileMapped: {} + Unit_HRR_HostBudget_RegionReleaseKeepsReplacingAllocation: {} + Unit_HRR_HostUnregister_FailedCallKeepsBuffer: {} + Unit_HRR_HostUnregister_FailedReplaceTakesNewBlock: {} + Unit_HRR_HostUnregister_GraphNodeKeepsBuffer: {} + Unit_HRR_KernelArgsSignature_ArgumentsCarryReadPadding: {} + Unit_HRR_KernelArgsSignature_GraphKernelNode: {} + Unit_HRR_KernelArgsSignature_GraphSetParams: {} + Unit_HRR_KernelArgsSignature_NarrowArgument: {} + Unit_HRR_KernelArgsSignature_SetParamsWithoutParameters: {} + Unit_HRR_KernelArgsSignature_TooFewArguments: {} + Unit_HRR_Memcpy3DArray_FreedHandleIsForgotten: {} + Unit_HRR_Memcpy3DArray_HostCopiesAreNotIssued: {} + Unit_HRR_Memcpy3DArray_TranslatesRecordedHandles: {} + Unit_HRR_Memcpy3DArray_UnknownHandleIsRefused: {} + Unit_HRR_Memcpy3DBatchArray_BadOperandIsRefused: {} + Unit_HRR_Query_CompletingQueryReturnsSuccess: {} + Unit_HRR_Query_DefaultHasNoAttemptBound: {} + Unit_HRR_Query_NeverCompletingQueryEndsTheReplay: {} + Unit_HRR_Query_OtherErrorIsReturnedAtOnce: {} + Unit_HRR_Query_StopsWhenReplayIsFatal: {} + Unit_HRR_Query_ZeroBoundMakesNoCall: {} + Unit_HRR_Refusals_GeneratedHandlersRefuseUntranslatableSlots: {} + Unit_HRR_Refusals_LinkAddFileIsRefused: {} + Unit_HRR_Refusals_LinkCreateRefusesPointerCarryingOptions: {} + Unit_HRR_Refusals_Memcpy2DNoOpCopiesAreNotRefused: {} + Unit_HRR_Refusals_Memcpy2DWithUnmappedEndIsRefused: {} + Unit_HRR_SnapshotOrder_FailedUnregisterKeepsBuffer: {} + Unit_HRR_SnapshotOrder_GuardReleasesTheHold: {} + Unit_HRR_SnapshotOrder_HostFreeWaitsForLaunch: {} + Unit_HRR_SnapshotOrder_LaunchHandlersTakeTheHold: {} + Unit_HRR_SnapshotOrder_UnregisterWaitsForLaunch: {} + Unit_HRR_SymbolCopyBounds_InRangeCopyReachesHip: {} + Unit_HRR_SymbolCopyBounds_OutOfRangeIsRefusedBeforeHip: {} + Unit_HRR_SymbolCopyBounds_WrappingSumIsRefused: {} + Unit_HRR_UnmappedNote_IpcCloseOfSkippedImport: {} + Unit_HRR_UnmappedNote_MemMap: {} diff --git a/projects/hrr/tests/config/integration.yaml b/projects/hrr/tests/config/integration.yaml index db5232e51081..539cac34681e 100644 --- a/projects/hrr/tests/config/integration.yaml +++ b/projects/hrr/tests/config/integration.yaml @@ -3,6 +3,8 @@ # # A skip rule matches when every key under "when" matches (a list is OR, a # missing key matches anything, and a value may be a glob such as gfx11*). +# "emulator: true" matches only a run under a GPU emulator (run_catch2.py +# --launcher), "emulator: false" only a run on a real device. # kind: disabled is a temporary regression and requires a public issue URL. # kind: unsupported is permanent and needs only a reason. # @@ -15,9 +17,16 @@ binary: hrr-integration-tests select: "[hrr]~[direct]" targets: os: [linux, windows] - arch: [gfx90a, gfx942, gfx950, gfx1030, gfx1100, gfx1101, gfx1102, gfx1150, gfx1151, gfx1153, gfx1200, gfx1201, gfx1250] + arch: [gfx90a, gfx942, gfx950, gfx1030, gfx1100, gfx1101, gfx1102, gfx1103, gfx1150, gfx1151, gfx1153, gfx1200, gfx1201, gfx1250] cases: Unit_HRR_AllApisRoundtrip: {} + Unit_HRR_AllApisArraysRoundtrip: + skip: + - when: {arch: ["gfx90a", "gfx11*", "gfx120*"], emulator: true} + kind: disabled + reason: "rocJitsu does not execute image_store, which hipMemcpy3D to and from a hipArray needs on image-capable archs" + issue: "https://github.com/ROCm/rocm-systems/issues/14612" + Unit_HRR_AllApisArrays_Direct: {} Unit_HRR_AllApis_Direct: {} Unit_HRR_ApiMatrix_Breadth2_Direct: {} Unit_HRR_ApiMatrix_Breadth_Direct: {} @@ -39,10 +48,24 @@ cases: Unit_HRR_ApiMatrix_SptCoopLaunch_Direct: {} Unit_HRR_ApiMatrix_SptKernelLaunch_Direct: {} Unit_HRR_ApiMatrix_T0_Roundtrip: {} - Unit_HRR_ApiMatrix_T1_Roundtrip: {} + Unit_HRR_ApiMatrix_T1_Roundtrip: + skip: + - when: {emulator: true} + kind: disabled + reason: "hipIpcGetMemHandle fails under rocJitsu since the shared runtime API headers change (hsa_amd_ipc_memory_create returns HSA_STATUS_ERROR)" + issue: "https://github.com/ROCm/rocm-systems/issues/13803" Unit_HRR_ApiMatrix_T2_Roundtrip: {} Unit_HRR_ApiMatrix_T3_Roundtrip: {} - Unit_HRR_ApiMatrix_T4_Roundtrip: {} + Unit_HRR_ApiMatrix_T4_Roundtrip: + skip: + - when: {arch: ["gfx11*", "gfx120*"], emulator: true} + kind: disabled + reason: "hipMemcpyToArray and hipMemcpyFromArray copy no data under the RDNA3 and RDNA4 rocJitsu configs" + issue: "https://github.com/ROCm/rocm-systems/issues/13724" + - when: {arch: gfx90a, emulator: true} + kind: disabled + reason: "rocJitsu does not execute image_store, which hipMemcpy3D to and from a hipArray needs on image-capable archs" + issue: "https://github.com/ROCm/rocm-systems/issues/14612" Unit_HRR_ApiMatrix_T5_Roundtrip: {} Unit_HRR_ApiMatrix_UC1Dense_Direct: {} Unit_HRR_BlankCaptureOutputCapturesNothing: {} @@ -50,7 +73,12 @@ cases: Unit_HRR_CaptureActiveMarker: {} Unit_HRR_CaptureActiveMarker_Direct: {} Unit_HRR_CaptureArchiveIsPrivate: {} - Unit_HRR_CaptureCapabilities: {} + Unit_HRR_CaptureCapabilities: + skip: + - when: {emulator: true} + kind: disabled + reason: "hipIpcGetMemHandle fails under rocJitsu since the shared runtime API headers change (hsa_amd_ipc_memory_create returns HSA_STATUS_ERROR)" + issue: "https://github.com/ROCm/rocm-systems/issues/13803" Unit_HRR_CaptureCapabilitiesImport_Direct: {} Unit_HRR_CaptureCapabilities_Direct: {} Unit_HRR_CaptureClosesArchiveDirs: {} @@ -129,7 +157,11 @@ cases: Unit_HRR_DiskSpace_StopsBeforeReserve_Direct: {} Unit_HRR_DispatchSwap_ChildKeepsInheritedPreload: {} Unit_HRR_DispatchSwap_Direct: {} - Unit_HRR_DispatchSwap_NoUnreplayableAfterTrailer: {} + Unit_HRR_DispatchSwap_NoUnreplayableAfterTrailer: + skip: + - when: {emulator: true} + kind: unsupported + reason: "The case replaces LD_PRELOAD in its child, which drops the rocJitsu interposer and leaves the child with no device" Unit_HRR_DispatchSwap_Preload_Direct: {} Unit_HRR_DispatchSwap_Unreplayable_Direct: {} Unit_HRR_DispatchSwap_WrapperOutlivesCapture: {} @@ -296,7 +328,11 @@ cases: Unit_HRR_ReplaceKernelBadSpec: {} Unit_HRR_ReplaceKernelMissingCO: {} Unit_HRR_ReplaceKernelRoundtrip: {} - Unit_HRR_SecureExecIgnoresCaptureOutput: {} + Unit_HRR_SecureExecIgnoresCaptureOutput: + skip: + - when: {emulator: true} + kind: unsupported + reason: "The loader ignores LD_PRELOAD in the set-group-ID copy, which drops the rocJitsu interposer and leaves the copy with no device" Unit_HRR_SecureExecProbe_Direct: {} Unit_HRR_ShutdownWhileChildOpensArchive: {} Unit_HRR_ShutdownWhileChildOpensArchive_Direct: {} @@ -325,6 +361,11 @@ cases: Unit_HRR_VMMRoundtrip: {} Unit_HRR_VMM_Direct: {} Unit_HRR_VaPlacement_Direct: {} - Unit_HRR_VaPlacement_StoredPointer: {} + Unit_HRR_VaPlacement_StoredPointer: + skip: + - when: {arch: ["gfx95*", "gfx125*"], emulator: true} + kind: disabled + reason: "The multi-GPU rocJitsu configs reject overlapping GEM_VA ranges, so replay cannot place allocations at their capture addresses" + issue: "https://github.com/ROCm/rocm-systems/issues/13723" Unit_HRR_ZeroInitRead_Direct: {} Unit_HRR_ZeroInitRoundtrip: {} diff --git a/projects/hrr/tests/config/unit.yaml b/projects/hrr/tests/config/unit.yaml index cb0b6107d36a..95365a712dd4 100644 --- a/projects/hrr/tests/config/unit.yaml +++ b/projects/hrr/tests/config/unit.yaml @@ -3,6 +3,8 @@ # # A skip rule matches when every key under "when" matches (a list is OR, a # missing key matches anything, and a value may be a glob such as gfx11*). +# "emulator: true" matches only a run under a GPU emulator (run_catch2.py +# --launcher), "emulator: false" only a run on a real device. # kind: disabled is a temporary regression and requires a public issue URL. # kind: unsupported is permanent and needs only a reason. # @@ -15,7 +17,7 @@ binary: hrr-unit-tests select: "*" targets: os: [linux, windows] - arch: [gfx90a, gfx942, gfx950, gfx1030, gfx1100, gfx1101, gfx1102, gfx1150, gfx1151, gfx1153, gfx1200, gfx1201, gfx1250] + arch: [gfx90a, gfx942, gfx950, gfx1030, gfx1100, gfx1101, gfx1102, gfx1103, gfx1150, gfx1151, gfx1153, gfx1200, gfx1201, gfx1250] cases: Unit_HRR_AllocKind_FreeDispatch: {} Unit_HRR_Format_BadMagic: {} @@ -29,6 +31,40 @@ cases: Unit_HRR_HostRect_Overflow: {} Unit_HRR_HostRect_PitchedPadded3D: {} Unit_HRR_HostRect_RuntimeDefaults: {} + Unit_HRR_Limits_ByteBudget: {} + Unit_HRR_Limits_ByteBudgetConcurrent: {} + Unit_HRR_Limits_ContiguousSequenceReplays: {} + Unit_HRR_Limits_DuplicateIsReportedImmediately: {} + Unit_HRR_Limits_DuplicateSequenceEndsReplay: {} + Unit_HRR_Limits_EventCap: {} + Unit_HRR_Limits_FatalFlagEndsWait: {} + Unit_HRR_Limits_FindDuplicateSeq: {} + Unit_HRR_Limits_FindSeqGap: {} + Unit_HRR_Limits_LoadByteBudget: {} + Unit_HRR_Limits_LoadChargesBlobAndCodeObjectMaps: {} + Unit_HRR_Limits_LoadChargesEventVectorAndSort: {} + Unit_HRR_Limits_LoadChargesParsedKernelExpansion: {} + Unit_HRR_Limits_LoadChargesPerAllocationOverhead: {} + Unit_HRR_Limits_NoStallWhileACallIsInFlight: {} + Unit_HRR_Limits_ParseU64: {} + Unit_HRR_Limits_PlaybackContextHostBudget: {} + Unit_HRR_Limits_PlaybackInfoRefusesOverByteBudget: {} + Unit_HRR_Limits_PlaybackRefusesTooManyEvents: {} + Unit_HRR_Limits_PlaybackRefusesTooManyThreads: {} + Unit_HRR_Limits_PlaybackRefusesWallSecondsAboveMaximum: {} + Unit_HRR_Limits_QueryRetryIsBounded: {} + Unit_HRR_Limits_ReadFileMissing: {} + Unit_HRR_Limits_ReadFileRefusesOversize: {} + Unit_HRR_Limits_RecordChargedBeforeAllocation: {} + Unit_HRR_Limits_ScopedCharge: {} + Unit_HRR_Limits_SequenceGapEndsReplay: {} + Unit_HRR_Limits_ThreadCap: {} + Unit_HRR_Limits_WallClockEndsBlockedLoad: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_limits_test.cc compiles this case out under _WIN32 (it blocks the load on a FIFO) + Unit_HRR_Limits_WallClockWatchdog: {} Unit_HRR_Playback_PlacementHeldRanges: {} Unit_HRR_Playback_PlacementRangeChecked: {} Unit_HRR_Playback_PlacementPlan: {} @@ -41,6 +77,7 @@ cases: Unit_HRR_Recovery_NoTrailer: {} Unit_HRR_Recovery_OversizeRecordBeforeTrailer: {} Unit_HRR_Recovery_RecordAfterTrailer: {} + Unit_HRR_Recovery_RepairRootHonoursMaxHostBytes: {} Unit_HRR_Recovery_RepairRootRewritesMiscountedTrailer: {} Unit_HRR_Recovery_RepairWritesProcessManifest: {} Unit_HRR_Recovery_RootInfoChecksTrailerCount: {} @@ -49,6 +86,94 @@ cases: Unit_HRR_Recovery_TornPayload: {} Unit_HRR_Recovery_TrailerCountMismatch: {} Unit_HRR_Recovery_TrailerOvercounts: {} + Unit_HRR_Repair_DropsSetuidAndSetgid: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_EventsLinkIsRefused: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_KeepsFilePermissions: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_LinkAtOldTempNameIsLeftAlone: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_LinkedPidDirectoryIsNotIndexed: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_ManifestLinkIsRefused: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_PidDirectoryLinkIsRefused: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_RecoversTornEventsWithoutLeftovers: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_RootManifestLinkIsRefusedBeforeFanOut: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Repair_UserNamedLinkIsResolved: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_SafeFs_DotDotAfterLinkIsResolvedByTheKernel: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_SafeFs_FdOwnsItsDescriptor: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_SafeFs_OpenDirRefusesLinks: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_SafeFs_RandomHexAndErrnoText: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_SafeFs_TempFileModeAndName: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_SafeFs_TempIsExclusiveAndReplaceDoesNotFollow: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_SafeFs_UnlinkAtRemovesTheNameNotTheTarget: + skip: + - when: {os: windows} + kind: unsupported + reason: hrr_repair_test.cc is compiled out under _WIN32 (POSIX file-descriptor APIs) + Unit_HRR_Slots_BoundedCstr: {} + Unit_HRR_Slots_JitOptionArray: {} + Unit_HRR_Slots_JitOptionClassification: {} Unit_HRR_Scrub_RefusesOutOfBoundsTables: {} Unit_HRR_Scrub_SkipsUnterminatedName: {} Unit_HRR_Scrub_ZeroesAndRetypesBuildInfo: {} diff --git a/projects/hrr/tests/integration/hrr_api_matrix_expectations.h b/projects/hrr/tests/integration/hrr_api_matrix_expectations.h index 4014428d5378..ee254beb91f8 100644 --- a/projects/hrr/tests/integration/hrr_api_matrix_expectations.h +++ b/projects/hrr/tests/integration/hrr_api_matrix_expectations.h @@ -675,6 +675,7 @@ inline constexpr const char* const kHrrWorkloadsT4[] = { "Unit_HRR_StreamCaptureQuerySpt_Direct", "Unit_HRR_ApiMatrix_MemPoolShare_Direct", "Unit_HRR_AllApis_Direct", + "Unit_HRR_AllApisArrays_Direct", "Unit_HRR_StressApis_Direct", "Unit_HRR_DeviceInfo_Direct", "Unit_HRR_DeviceExtra_Direct", diff --git a/projects/hrr/tests/integration/hrr_host_alloc_budget_test.cc b/projects/hrr/tests/integration/hrr_host_alloc_budget_test.cc index 70bc3f6857b0..5fd9c1029ead 100644 --- a/projects/hrr/tests/integration/hrr_host_alloc_budget_test.cc +++ b/projects/hrr/tests/integration/hrr_host_alloc_budget_test.cc @@ -13,8 +13,8 @@ * refuses a record that would pass the cap, and gives the charge back on hipFree * and hipMemRelease. Each workload allocates 2 MiB three times in a row, * freeing the buffer in between, so a replay whose cap holds one buffer (3 MiB) - * only gets through if the free returns the charge, and a cap below one buffer - * (1 MiB) refuses the first. + * only gets through if the free returns the charge, and a cap with less than one + * buffer (1 MiB) to spare refuses the first. Caps are relative to the archive's cached blobs. */ #include "hrr_test_common.hh" @@ -22,17 +22,41 @@ #include namespace { +// Replay charges every blob and code object it loads to --max-host-bytes and +// never evicts them. The registered fat binary is among them, and its size +// follows the arch list the test binary was built for (over 1 MiB for the CI +// build of every arch), so a cap is the archive's cached bytes plus the budget +// the test means to give the allocations. +uint64_t hrr_archive_cache_bytes(const fs::path& capture_root) { + // The capture root holds one pid-/ archive; the blobs sit inside it. + const fs::path archive = hrr_single_process_archive(capture_root); + uint64_t total = 0; + for (const char* sub : {"blobs", "code_objects"}) { + std::error_code ec; + if (!fs::is_directory(archive / sub, ec)) continue; + for (fs::recursive_directory_iterator it(archive / sub, ec), end; !ec && it != end; + it.increment(ec)) { + std::error_code fec; + if (it->is_regular_file(fec)) total += fs::file_size(it->path(), fec); + } + } + return total; +} +std::string hrr_host_cap(const fs::path& archive, uint64_t alloc_bytes) { + return std::to_string(hrr_archive_cache_bytes(archive) + alloc_bytes); +} + constexpr size_t kBudgetBuf = 2u << 20; constexpr int kBudgetRounds = 3; // Capture is not capped; only replay is. -void hrr_budget_replay(const char* direct_case, const char* dir, const char* cap_bytes, - const char* api) { +void hrr_budget_replay(const char* direct_case, const char* dir, const char* api) { ScopedDir cap{hrr_temp_dir(dir)}; hrr_capture_direct(direct_case, cap.path, 5); for (const bool over : {false, true}) { const std::string cap_arg = - std::string("--max-host-bytes ") + (over ? "1048576" : cap_bytes); + std::string("--max-host-bytes ") + + hrr_host_cap(cap.path, over ? (1u << 20) : (3u << 20)); auto [ret, out] = hrr_playback_merged(cap.path, cap_arg); INFO(cap_arg << ": playback exit " << ret << ", output:\n" << out); if (over) { @@ -85,12 +109,12 @@ TEST_CASE("Unit_HRR_HostBudgetMemCreate_Direct", "[.][hrr-direct]") { // Within a one-buffer cap the free returns the charge; under the buffer, the // first allocation is refused. HRR_TEST_CASE(Unit_HRR_HostBudgetManagedRoundtrip) { - hrr_budget_replay("Unit_HRR_HostBudgetManaged_Direct", "hrr_budget_managed", "3145728", + hrr_budget_replay("Unit_HRR_HostBudgetManaged_Direct", "hrr_budget_managed", "hipMallocManaged"); } HRR_TEST_CASE(Unit_HRR_HostBudgetMemCreateRoundtrip) { - hrr_budget_replay("Unit_HRR_HostBudgetMemCreate_Direct", "hrr_budget_memcreate", "3145728", + hrr_budget_replay("Unit_HRR_HostBudgetMemCreate_Direct", "hrr_budget_memcreate", "hipMemCreate (host location)"); } @@ -98,7 +122,7 @@ HRR_TEST_CASE(Unit_HRR_HostBudgetMemCreateRoundtrip) { // hipMemUnmap. The workload maps h1, releases it, then creates h2 before the // unmap: with the hold the 3 MiB replay is refused at h2 (h1's 2 MiB is still // charged); without it the release would have returned the charge and h2 fits. -// A 5 MiB cap takes h1 + h2 plus the replay's cached blobs, so the replay +// A cap of 5 MiB over the replay's cached blobs takes h1 + h2, so the replay // completes. TEST_CASE("Unit_HRR_HostBudgetMemMapHold_Direct", "[.][hrr-direct]") { HRR_HIP_CHECK(hipSetDevice(0)); @@ -138,7 +162,8 @@ HRR_TEST_CASE(Unit_HRR_HostBudgetMemMapHoldRoundtrip) { hrr_capture_direct("Unit_HRR_HostBudgetMemMapHold_Direct", cap.path, 5); for (const bool refuse : {false, true}) { const std::string cap_arg = - std::string("--max-host-bytes ") + (refuse ? "3145728" : "5242880"); + std::string("--max-host-bytes ") + + hrr_host_cap(cap.path, refuse ? (3u << 20) : (5u << 20)); auto [ret, out] = hrr_playback_merged(cap.path, cap_arg); INFO(cap_arg << ": playback exit " << ret << ", output:\n" << out); if (refuse) { @@ -165,7 +190,8 @@ HRR_TEST_CASE(Unit_HRR_HostBudgetManagedLeakRoundtrip) { ScopedDir cap{hrr_temp_dir("hrr_budget_managed_leak")}; hrr_capture_direct("Unit_HRR_HostBudgetManagedLeak_Direct", cap.path, 5); auto [ret, out] = hrr_playback_merged( - cap.path, "--kernel-filter hrr_no_such_kernel --max-host-bytes 3145728"); + cap.path, "--kernel-filter hrr_no_such_kernel --max-host-bytes " + + hrr_host_cap(cap.path, 3u << 20)); INFO("playback exit " << ret << ", output:\n" << out); CHECK(ret == 0); CHECK(out.find("[HRR] Fatal") == std::string::npos); diff --git a/projects/hrr/tests/integration/hrr_roundtrip_test.cc b/projects/hrr/tests/integration/hrr_roundtrip_test.cc index 150ca91938bd..c8cb79f8e85d 100644 --- a/projects/hrr/tests/integration/hrr_roundtrip_test.cc +++ b/projects/hrr/tests/integration/hrr_roundtrip_test.cc @@ -371,21 +371,44 @@ HRR_TEST_CASE(Unit_HRR_AllApisRoundtrip) { REQUIRE(arc.events.size() >= 40); // ~55 distinct APIs exercised } - // The workload's array copies (block 19) need image support, which gfx950 - // (MI350X) reports as absent. Where it has them, a capture without the copies - // would replay perfectly well, so ask for them. - int images = 0; - HRR_HIP_CHECK(hipDeviceGetAttribute(&images, hipDeviceAttributeImageSupport, 0)); - if (images) - hrr_require_recorded_apis(cap.path, {HRR_API_HIPARRAY3DCREATE, HRR_API_HIPARRAYCREATE, - HRR_API_HIPMEMCPY3D, HRR_API_HIPMEMCPY3DASYNC}); - // ------------------------------------------------------------------------- // Step 3: playback + D2H validation (d2[i] == 94) // ------------------------------------------------------------------------- hrr_run_playback(cap.path); } +/** + * Test Description + * ---------------- + * - Spawns HrrTest Unit_HRR_AllApisArrays_Direct as a subprocess with + * HIP_HRR_CAPTURE_OUTPUT set to a temp directory: hipArrayCreate, + * hipArray3DCreate and hipMemcpy3D/hipMemcpy3DAsync with an array on either + * end. + * - Requires the array creation and copy calls in the archive, then replays it + * and validates the D2H blobs of the copied-out buffers. + * - Skips where the device reports no image support (gfx950 / MI350X), since + * the workload then records no array calls. + */ +HRR_TEST_CASE(Unit_HRR_AllApisArraysRoundtrip) { + int images = 0; + HRR_HIP_CHECK(hipDeviceGetAttribute(&images, hipDeviceAttributeImageSupport, 0)); + if (!images) HRR_SKIP("device reports no image support"); + + ScopedDir cap{fs::temp_directory_path() / "hrr_roundtrip_allapis_arrays"}; + { + hrr::test::SpawnProc proc(hrr_self_exe()); + proc.setEnv("HIP_HRR_CAPTURE_OUTPUT", cap.path.string()); + set_proc_search_path(proc); + int ret = proc.run("\"Unit_HRR_AllApisArrays_Direct\""); + INFO("AllApisArrays capture subprocess exit code: " << ret); + REQUIRE(ret == 0); + } + + hrr_require_recorded_apis(cap.path, {HRR_API_HIPARRAY3DCREATE, HRR_API_HIPARRAYCREATE, + HRR_API_HIPMEMCPY3D, HRR_API_HIPMEMCPY3DASYNC}); + hrr_run_playback(cap.path, "", /*require_d2h=*/true); +} + /** * Test Description * ---------------- diff --git a/projects/hrr/tests/integration/hrr_workload_test.cc b/projects/hrr/tests/integration/hrr_workload_test.cc index 0a4e16258a8a..5d9f28cd46f5 100644 --- a/projects/hrr/tests/integration/hrr_workload_test.cc +++ b/projects/hrr/tests/integration/hrr_workload_test.cc @@ -50,6 +50,7 @@ #include #include #include +#include #include #include #include @@ -331,6 +332,157 @@ TEST_CASE("Unit_HRR_HostMemWorkload_Direct", "[.][hrr-direct]") { HRR_HIP_CHECK(hipFree(buf)); } +// --------------------------------------------------------------------------- +// Array handle-map and hipMemcpy3D coverage — hidden ([.]). +// +// Split from Unit_HRR_AllApis_Direct so that a GPU emulator which cannot execute +// the array copies skips only these, not the rest of the API coverage. +// The D2H blobs captured are the output buffers of the array copies. +// --------------------------------------------------------------------------- +TEST_CASE("Unit_HRR_AllApisArrays_Direct", "[.][hrr-direct]") { + HRR_HIP_CHECK(hipSetDevice(0)); + + // ========================================================================= + // 19. hipArrayCreate + hipArray3DCreate — handle-map coverage + // + // Array creation requires image support. Query device capability first and + // skip gracefully on hardware that reports no texture support. + // ========================================================================= + { + int supportsImages = 0; + HRR_HIP_CHECK(hipDeviceGetAttribute(&supportsImages, hipDeviceAttributeImageSupport, 0)); + if (supportsImages) { + HIP_ARRAY_DESCRIPTOR desc1d{}; + desc1d.Width = static_cast(N); + desc1d.Height = 0; // 1-D + desc1d.Format = HIP_AD_FORMAT_SIGNED_INT32; + desc1d.NumChannels = 1; + hipArray_t arr1d = nullptr; + HRR_HIP_CHECK(hipArrayCreate(&arr1d, &desc1d)); + HRR_HIP_CHECK(hipFreeArray(arr1d)); + + HIP_ARRAY3D_DESCRIPTOR desc3d{}; + desc3d.Width = 4; + desc3d.Height = 4; + desc3d.Depth = 4; + desc3d.Format = HIP_AD_FORMAT_SIGNED_INT32; + desc3d.NumChannels = 1; + desc3d.Flags = 0; + hipArray_t arr3d = nullptr; + HRR_HIP_CHECK(hipArray3DCreate(&arr3d, &desc3d)); + HRR_HIP_CHECK(hipFreeArray(arr3d)); + + // hipMemcpy3D and hipMemcpy3DAsync with an array on either end: device + // buffers into a 3-D and a 2-D array, array to array, and back into device + // buffers, whose hipMemcpy to the host is the D2H blob replay validates. A + // replay that handed the recorded array address to HIP instead of the array + // it created for it would crash or move other bytes. No copy has a host end, + // which replay does not issue. + constexpr size_t W = 16, H = 4, D = 2; // elements; one int each + constexpr size_t ROW = W * sizeof(int); + constexpr size_t BYTES3 = ROW * H * D; + constexpr size_t BYTES2 = ROW * H; + + std::vector h3(W * H * D), h2(W * H); + for (size_t i = 0; i < h3.size(); ++i) h3[i] = static_cast(i * 2654435761u) ^ 0x3d3d3d3d; + for (size_t i = 0; i < h2.size(); ++i) h2[i] = static_cast(i * 2654435761u) ^ 0x2d2d2d2d; + + HIP_ARRAY3D_DESCRIPTOR d3{}; + d3.Width = W; + d3.Height = H; + d3.Depth = D; + d3.Format = HIP_AD_FORMAT_SIGNED_INT32; + d3.NumChannels = 1; + hipArray_t a3 = nullptr, a3_copy = nullptr; + HRR_HIP_CHECK(hipArray3DCreate(&a3, &d3)); + HRR_HIP_CHECK(hipArray3DCreate(&a3_copy, &d3)); + + HIP_ARRAY_DESCRIPTOR d2{}; + d2.Width = W; + d2.Height = H; + d2.Format = HIP_AD_FORMAT_SIGNED_INT32; + d2.NumChannels = 1; + hipArray_t a2 = nullptr; + HRR_HIP_CHECK(hipArrayCreate(&a2, &d2)); + + // Device buffers the arrays are filled from (in*) and read back into (out*). + void *in3 = nullptr, *in2 = nullptr, *out3 = nullptr, *out3_copy = nullptr, *out2 = nullptr; + HRR_HIP_CHECK(hipMalloc(&in3, BYTES3)); + HRR_HIP_CHECK(hipMalloc(&in2, BYTES2)); + HRR_HIP_CHECK(hipMalloc(&out3, BYTES3)); + HRR_HIP_CHECK(hipMalloc(&out3_copy, BYTES3)); + HRR_HIP_CHECK(hipMalloc(&out2, BYTES2)); + HRR_HIP_CHECK(hipMemcpy(in3, h3.data(), BYTES3, hipMemcpyHostToDevice)); + HRR_HIP_CHECK(hipMemcpy(in2, h2.data(), BYTES2, hipMemcpyHostToDevice)); + HRR_HIP_CHECK(hipMemset(out3, 0, BYTES3)); + HRR_HIP_CHECK(hipMemset(out3_copy, 0, BYTES3)); + HRR_HIP_CHECK(hipMemset(out2, 0, BYTES2)); + + // device -> array, plain and async + { + hipMemcpy3DParms p{}; + p.srcPtr = make_hipPitchedPtr(in3, ROW, W, H); + p.dstArray = a3; + p.extent = make_hipExtent(W, H, D); + p.kind = hipMemcpyDeviceToDevice; + HRR_HIP_CHECK(hipMemcpy3D(&p)); + } + { + hipMemcpy3DParms p{}; + p.srcPtr = make_hipPitchedPtr(in2, ROW, W, H); + p.dstArray = a2; + p.extent = make_hipExtent(W, H, 1); + p.kind = hipMemcpyDeviceToDevice; + HRR_HIP_CHECK(hipMemcpy3DAsync(&p, nullptr)); + HRR_HIP_CHECK(hipDeviceSynchronize()); + } + // array -> array + { + hipMemcpy3DParms p{}; + p.srcArray = a3; + p.dstArray = a3_copy; + p.extent = make_hipExtent(W, H, D); + p.kind = hipMemcpyDeviceToDevice; + HRR_HIP_CHECK(hipMemcpy3D(&p)); + } + // array -> device + auto array_to_device = [&](hipArray_t src, void* dst, size_t depth) { + hipMemcpy3DParms p{}; + p.srcArray = src; + p.dstPtr = make_hipPitchedPtr(dst, ROW, W, H); + p.extent = make_hipExtent(W, H, depth); + p.kind = hipMemcpyDeviceToDevice; + HRR_HIP_CHECK(hipMemcpy3D(&p)); + }; + array_to_device(a3, out3, D); + array_to_device(a3_copy, out3_copy, D); + array_to_device(a2, out2, 1); + HRR_HIP_CHECK(hipDeviceSynchronize()); + + // The D2H blobs of the output buffers are what replay validates. + std::vector r3(W * H * D), r3c(W * H * D), r2(W * H); + HRR_HIP_CHECK(hipMemcpy(r3.data(), out3, BYTES3, hipMemcpyDeviceToHost)); + HRR_HIP_CHECK(hipMemcpy(r3c.data(), out3_copy, BYTES3, hipMemcpyDeviceToHost)); + HRR_HIP_CHECK(hipMemcpy(r2.data(), out2, BYTES2, hipMemcpyDeviceToHost)); + + // Each array copy must have moved the bytes: if the capture's copies moved + // nothing, the blobs would be zeros and the roundtrip would still agree. + REQUIRE(r3 == h3); + REQUIRE(r3c == h3); + REQUIRE(r2 == h2); + + HRR_HIP_CHECK(hipFree(in3)); + HRR_HIP_CHECK(hipFree(in2)); + HRR_HIP_CHECK(hipFree(out3)); + HRR_HIP_CHECK(hipFree(out3_copy)); + HRR_HIP_CHECK(hipFree(out2)); + HRR_HIP_CHECK(hipFreeArray(a3)); + HRR_HIP_CHECK(hipFreeArray(a3_copy)); + HRR_HIP_CHECK(hipFreeArray(a2)); + } + } +} + // --------------------------------------------------------------------------- // Comprehensive HIP API coverage workload — hidden ([.]). // @@ -640,146 +792,6 @@ TEST_CASE("Unit_HRR_AllApis_Direct", "[.][hrr-direct]") { HRR_HIP_CHECK(hipMemPoolSetAccess(pool, &access_desc, 1)); } - // ========================================================================= - // 19. hipArrayCreate + hipArray3DCreate — handle-map coverage - // - // Array creation requires image support. Query device capability first and - // skip gracefully on hardware that reports no texture support. - // ========================================================================= - { - int supportsImages = 0; - HRR_HIP_CHECK(hipDeviceGetAttribute(&supportsImages, hipDeviceAttributeImageSupport, 0)); - if (supportsImages) { - HIP_ARRAY_DESCRIPTOR desc1d{}; - desc1d.Width = static_cast(N); - desc1d.Height = 0; // 1-D - desc1d.Format = HIP_AD_FORMAT_SIGNED_INT32; - desc1d.NumChannels = 1; - hipArray_t arr1d = nullptr; - HRR_HIP_CHECK(hipArrayCreate(&arr1d, &desc1d)); - HRR_HIP_CHECK(hipFreeArray(arr1d)); - - HIP_ARRAY3D_DESCRIPTOR desc3d{}; - desc3d.Width = 4; - desc3d.Height = 4; - desc3d.Depth = 4; - desc3d.Format = HIP_AD_FORMAT_SIGNED_INT32; - desc3d.NumChannels = 1; - desc3d.Flags = 0; - hipArray_t arr3d = nullptr; - HRR_HIP_CHECK(hipArray3DCreate(&arr3d, &desc3d)); - HRR_HIP_CHECK(hipFreeArray(arr3d)); - - // hipMemcpy3D and hipMemcpy3DAsync with an array on either end: device - // buffers into a 3-D and a 2-D array, array to array, and back into device - // buffers, whose hipMemcpy to the host is the D2H blob replay validates. A - // replay that handed the recorded array address to HIP instead of the array - // it created for it would crash or move other bytes. No copy has a host end, - // which replay does not issue. - constexpr size_t W = 16, H = 4, D = 2; // elements; one int each - constexpr size_t ROW = W * sizeof(int); - constexpr size_t BYTES3 = ROW * H * D; - constexpr size_t BYTES2 = ROW * H; - - std::vector h3(W * H * D), h2(W * H); - for (size_t i = 0; i < h3.size(); ++i) h3[i] = static_cast(i * 2654435761u) ^ 0x3d3d3d3d; - for (size_t i = 0; i < h2.size(); ++i) h2[i] = static_cast(i * 2654435761u) ^ 0x2d2d2d2d; - - HIP_ARRAY3D_DESCRIPTOR d3{}; - d3.Width = W; - d3.Height = H; - d3.Depth = D; - d3.Format = HIP_AD_FORMAT_SIGNED_INT32; - d3.NumChannels = 1; - hipArray_t a3 = nullptr, a3_copy = nullptr; - HRR_HIP_CHECK(hipArray3DCreate(&a3, &d3)); - HRR_HIP_CHECK(hipArray3DCreate(&a3_copy, &d3)); - - HIP_ARRAY_DESCRIPTOR d2{}; - d2.Width = W; - d2.Height = H; - d2.Format = HIP_AD_FORMAT_SIGNED_INT32; - d2.NumChannels = 1; - hipArray_t a2 = nullptr; - HRR_HIP_CHECK(hipArrayCreate(&a2, &d2)); - - // Device buffers the arrays are filled from (in*) and read back into (out*). - void *in3 = nullptr, *in2 = nullptr, *out3 = nullptr, *out3_copy = nullptr, *out2 = nullptr; - HRR_HIP_CHECK(hipMalloc(&in3, BYTES3)); - HRR_HIP_CHECK(hipMalloc(&in2, BYTES2)); - HRR_HIP_CHECK(hipMalloc(&out3, BYTES3)); - HRR_HIP_CHECK(hipMalloc(&out3_copy, BYTES3)); - HRR_HIP_CHECK(hipMalloc(&out2, BYTES2)); - HRR_HIP_CHECK(hipMemcpy(in3, h3.data(), BYTES3, hipMemcpyHostToDevice)); - HRR_HIP_CHECK(hipMemcpy(in2, h2.data(), BYTES2, hipMemcpyHostToDevice)); - HRR_HIP_CHECK(hipMemset(out3, 0, BYTES3)); - HRR_HIP_CHECK(hipMemset(out3_copy, 0, BYTES3)); - HRR_HIP_CHECK(hipMemset(out2, 0, BYTES2)); - - // device -> array, plain and async - { - hipMemcpy3DParms p{}; - p.srcPtr = make_hipPitchedPtr(in3, ROW, W, H); - p.dstArray = a3; - p.extent = make_hipExtent(W, H, D); - p.kind = hipMemcpyDeviceToDevice; - HRR_HIP_CHECK(hipMemcpy3D(&p)); - } - { - hipMemcpy3DParms p{}; - p.srcPtr = make_hipPitchedPtr(in2, ROW, W, H); - p.dstArray = a2; - p.extent = make_hipExtent(W, H, 1); - p.kind = hipMemcpyDeviceToDevice; - HRR_HIP_CHECK(hipMemcpy3DAsync(&p, nullptr)); - HRR_HIP_CHECK(hipDeviceSynchronize()); - } - // array -> array - { - hipMemcpy3DParms p{}; - p.srcArray = a3; - p.dstArray = a3_copy; - p.extent = make_hipExtent(W, H, D); - p.kind = hipMemcpyDeviceToDevice; - HRR_HIP_CHECK(hipMemcpy3D(&p)); - } - // array -> device - auto array_to_device = [&](hipArray_t src, void* dst, size_t depth) { - hipMemcpy3DParms p{}; - p.srcArray = src; - p.dstPtr = make_hipPitchedPtr(dst, ROW, W, H); - p.extent = make_hipExtent(W, H, depth); - p.kind = hipMemcpyDeviceToDevice; - HRR_HIP_CHECK(hipMemcpy3D(&p)); - }; - array_to_device(a3, out3, D); - array_to_device(a3_copy, out3_copy, D); - array_to_device(a2, out2, 1); - HRR_HIP_CHECK(hipDeviceSynchronize()); - - // The D2H blobs of the output buffers are what replay validates. - std::vector r3(W * H * D), r3c(W * H * D), r2(W * H); - HRR_HIP_CHECK(hipMemcpy(r3.data(), out3, BYTES3, hipMemcpyDeviceToHost)); - HRR_HIP_CHECK(hipMemcpy(r3c.data(), out3_copy, BYTES3, hipMemcpyDeviceToHost)); - HRR_HIP_CHECK(hipMemcpy(r2.data(), out2, BYTES2, hipMemcpyDeviceToHost)); - - // Each array copy must have moved the bytes: if the capture's copies moved - // nothing, the blobs would be zeros and the roundtrip would still agree. - REQUIRE(r3 == h3); - REQUIRE(r3c == h3); - REQUIRE(r2 == h2); - - HRR_HIP_CHECK(hipFree(in3)); - HRR_HIP_CHECK(hipFree(in2)); - HRR_HIP_CHECK(hipFree(out3)); - HRR_HIP_CHECK(hipFree(out3_copy)); - HRR_HIP_CHECK(hipFree(out2)); - HRR_HIP_CHECK(hipFreeArray(a3)); - HRR_HIP_CHECK(hipFreeArray(a3_copy)); - HRR_HIP_CHECK(hipFreeArray(a2)); - } - } - // ========================================================================= // Cleanup (reverse allocation order) // ========================================================================= @@ -4117,6 +4129,8 @@ TEST_CASE("Unit_HRR_CaptureCrashDuringFork_Direct", "[.][hrr-direct]") { // The crashes are the point; core files are not. struct rlimit no_core{0, 0}; (void)setrlimit(RLIMIT_CORE, &no_core); + // Non-dumpable too: the parent's core dump outlasts the CI watchdog (#15217). + (void)prctl(PR_SET_DUMPABLE, 0); HRR_HIP_CHECK(hipSetDevice(0)); int* d = nullptr; diff --git a/projects/hrr/tests/scripts/hrr_capture_build.py b/projects/hrr/tests/scripts/hrr_capture_build.py new file mode 100644 index 000000000000..d9fe4c5a251d --- /dev/null +++ b/projects/hrr/tests/scripts/hrr_capture_build.py @@ -0,0 +1,450 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +"""Build an amdhip64 with HRR capture compiled in, for the HRR GPU suite. + +Interim: this rebuilds CLR with capture enabled on top of TheRock release +artifacts. It stays until TheRock can build and run the HRR integration +suite. The hrr-capture variant alone does not: no workflow selects it, and +it does not install the test binaries. This script and +test_hrr_capture_build.py go away once that suite runs from TheRock artifacts. + +Release builds compile capture out (HIP_HRR_CAPTURE_ENABLED undefined). This +script installs the ROCm artifacts that rocm-systems' own TheRock CI built for +a recent develop commit, rebuilds only CLR from this checkout with capture +enabled, installs it over those artifacts, and builds projects/hrr on top. HRR +lives in projects/hrr and projects/clr only; ROCr and everything else stay +prebuilt. The GPU jobs then install the same artifacts plus this amdhip64, so +build and test use the same libraries, and the binaries' RUNPATH needs no +LD_LIBRARY_PATH. + + fetch-base Pick the base run and install the build-time artifacts. + build Build CLR with capture, install it into the base, build HRR. + verify Check the suite's case list, the embedded GPU targets and + which amdhip64 and ROCr the suite loads. + install-base On a GPU machine: install the run-time artifacts of the + recorded base run and the capture-enabled amdhip64. +""" + +from __future__ import annotations + +import argparse +import json +import os +import re +import shutil +import subprocess +import sys +from pathlib import Path +from typing import Callable, Sequence + +import hrr_test_config + +Runner = Callable[[Sequence[str]], None] + +BASE_REPO = "ROCm/rocm-systems" +BASE_WORKFLOW = "therock-multi-arch-ci.yml" +BASE_BRANCH = "develop" +# The generic (target-neutral) artifacts are the same in every family group. +BASE_ARTIFACT_GROUP = "gfx94X-dcgpu" + +# Artifacts needed to build CLR and projects/hrr. +BUILD_ARTIFACTS = ( + "amd-llvm_run", + "amd-llvm_lib", + "amd-llvm_dev", + "base_run", + "base_lib", + "base_dev", + "sysdeps_lib", + "sysdeps_dev", + "core-runtime_run", + "core-runtime_lib", + "core-runtime_dev", + "core-hip_run", + "core-hip_lib", + "core-hip_dev", +) + +# Artifacts the GPU suite loads at run time. +RUNTIME_ARTIFACTS = ( + "amd-llvm_lib", + "base_run", + "base_lib", + "sysdeps_lib", + "core-runtime_run", + "core-runtime_lib", + "core-hip_lib", +) + +# What CLR's install writes into lib/: amdhip64 plus hiprtc and its builtins. +# All of it comes from this checkout, so all of it is carried to the test job. +CAPTURE_LIB_GLOBS = ("libamdhip64.so*", "libhiprtc*.so*") + +# The suite must load exactly one copy of each, from the installed tree: the +# capture-enabled amdhip64 built here and the base's ROCr. +LOADED_LIBS = ("libamdhip64.so", "libhsa-runtime64.so") + +# Its targets.arch is the architecture list the GPU suite is built for. +INTEGRATION_CONFIG = Path("projects/hrr/tests/config/integration.yaml") + +# What the GPU suite runs. The host unit suite is built and run by hrr-ci.yml. +HRR_TARGETS = ("hrr-playback", "hrr-integration-tests", "hrr-test-dispatch-tool") + +# Uploaded by hrr-gpu.yml's build job, relative to --build. +EXECUTABLES = ( + "hrr/playback/hrr-playback", + "hrr/tests/integration/hrr-integration-tests", +) + + +def run(cmd: Sequence[str]) -> None: + print("+ " + " ".join(str(c) for c in cmd), flush=True) + subprocess.run([str(c) for c in cmd], check=True) + + +def pick_base_run(finder_json: dict) -> dict: + """The base run from find_latest_artifacts.py --json output.""" + if finder_json.get("status") != "found" or not finder_json.get("results"): + raise SystemExit( + f"no {BASE_BRANCH} run of {BASE_WORKFLOW} has the needed artifacts" + ) + result = finder_json["results"][0] + return { + "repo": result["github_repository_name"], + "run_id": str(result["workflow_run_id"]), + "commit": result["git_commit_sha"], + "run_url": result["workflow_run_html_url"], + "artifact_group": result["artifact_group"], + } + + +def find_base_run(therock: Path) -> dict: + cmd = [ + sys.executable, + therock / "build_tools" / "find_latest_artifacts.py", + "--repo", + BASE_REPO, + "--workflow", + BASE_WORKFLOW, + "--branch", + BASE_BRANCH, + "--artifact-group", + BASE_ARTIFACT_GROUP, + "--json", + ] + for artifact in BUILD_ARTIFACTS: + cmd += ["--require-artifact", f"{artifact}_generic"] + print("+ " + " ".join(str(c) for c in cmd), flush=True) + # Only stdout is captured: it carries the JSON, stderr the finder's log. + out = subprocess.run( + [str(c) for c in cmd], check=True, stdout=subprocess.PIPE, text=True + ) + return pick_base_run(json.loads(out.stdout)) + + +def fetch_command( + therock: Path, base: dict, dist: Path, artifacts: Sequence[str] +) -> list: + return [ + sys.executable, + therock / "build_tools" / "fetch_artifacts.py", + "--run-id", + base["run_id"], + "--run-github-repo", + base["repo"], + "--artifact-group", + base["artifact_group"], + "--output-dir", + dist, + "--flatten", + *artifacts, + ] + + +def clr_configure_command(src: Path, build: Path, dist: Path) -> list: + # HIP_HRR_CAPTURE_ENABLED is what compiles capture into amdhip64; CLR has + # no CMake option for it. + return [ + "cmake", + "-S", + src / "projects" / "clr", + "-B", + build / "clr", + "-GNinja", + "-Wno-dev", + "-DCMAKE_BUILD_TYPE=Release", + "-DCMAKE_CXX_FLAGS=-DHIP_HRR_CAPTURE_ENABLED", + f"-DCMAKE_PREFIX_PATH={dist}", + f"-DCMAKE_INSTALL_PREFIX={dist}", + # The artifacts use lib/; GNUInstallDirs would pick lib64 on this image. + "-DCMAKE_INSTALL_LIBDIR=lib", + # libhiprtc sets no RUNPATH of its own; comgr and ROCr sit next to it. + "-DCMAKE_INSTALL_RPATH=$ORIGIN", + f"-DROCM_PATH={dist}", + f"-DHIP_COMMON_DIR={src / 'projects' / 'hip'}", + "-DHIP_PLATFORM=amd", + "-DCLR_BUILD_HIP=ON", + "-DCLR_BUILD_OCL=OFF", + "-DHIPCC_BIN_DIR=", + ] + + +BUILD_RPATH = "$ORIGIN/../../../../dist/lib;$ORIGIN/../../../dist/lib" + + +def hrr_configure_command(src: Path, build: Path, dist: Path, archs: str) -> list: + llvm = dist / "lib" / "llvm" + # One compiler for C, C++ and HIP, as TheRock builds HRR: host objects from + # the system compiler are not position independent, and the HIP link is. + return [ + "cmake", + "-S", + src / "projects" / "hrr", + "-B", + build / "hrr", + "-GNinja", + "-Wno-dev", + "-DCMAKE_BUILD_TYPE=Release", + f"-DCMAKE_PREFIX_PATH={dist}", + f"-DROCM_PATH={dist}", + # CMake bakes the absolute dist/lib into RUNPATH, which is the build + # container's workspace path. The rocJitsu jobs run on the bare runner, + # whose workspace path differs. These entries reach dist/lib from the + # integration binary and from the playback binary, wherever the + # workspace sits. + f"-DCMAKE_BUILD_RPATH={BUILD_RPATH}", + f"-DCMAKE_C_COMPILER={llvm / 'bin' / 'clang'}", + f"-DCMAKE_CXX_COMPILER={llvm / 'bin' / 'clang++'}", + f"-DCMAKE_HIP_COMPILER={llvm / 'bin' / 'clang++'}", + f"-DCMAKE_HIP_FLAGS=--rocm-device-lib-path={llvm / 'amdgcn' / 'bitcode'}", + f"-DGPU_TARGETS={archs}", + f"-DCMAKE_HIP_ARCHITECTURES={archs}", + "-DHRR_BUILD_PLAYBACK=ON", + "-DHRR_BUILD_TESTS=ON", + "-DHRR_BUILD_INTEGRATION_TESTS=ON", + ] + + +def remove_amdhip64(lib_dir: Path) -> None: + """Remove the release amdhip64 and hiprtc so only the CLR build's remain.""" + for glob in CAPTURE_LIB_GLOBS: + for path in lib_dir.glob(glob): + path.unlink() + + +def copy_amdhip64(src_dir: Path, dst_dir: Path) -> list[str]: + """Copy the CAPTURE_LIB_GLOBS libraries from src_dir to dst_dir, keeping symlinks.""" + dst_dir.mkdir(parents=True, exist_ok=True) + copied = [] + for glob in CAPTURE_LIB_GLOBS: + for path in sorted(src_dir.glob(glob)): + target = dst_dir / path.name + if target.is_symlink() or target.exists(): + target.unlink() + if path.is_symlink(): + os.symlink(os.readlink(path), target) + else: + shutil.copy2(path, target) + copied.append(path.name) + if not any(name.startswith("libamdhip64.so") for name in copied): + raise SystemExit(f"no libamdhip64.so* in {src_dir}") + return copied + + +def linkage_errors(ldd_output: str, lib_dir: Path) -> list[str]: + """Each of LOADED_LIBS must resolve to exactly one file, in lib_dir.""" + resolved: dict[str, set[str]] = {} + for match in re.finditer(r"^\s*(\S+) => (\S+)", ldd_output, re.MULTILINE): + name, path = match.groups() + # A relative RUNPATH entry shows up as .../build/hrr/../../dist/lib. + path = os.path.normpath(path) + for stem in LOADED_LIBS: + if name.startswith(stem): + resolved.setdefault(stem, set()).add(path) + errors = [] + for stem in LOADED_LIBS: + paths = resolved.get(stem, set()) + if len(paths) != 1: + errors.append(f"{stem} resolves to {sorted(paths) or 'nothing'}") + elif Path(next(iter(paths))).parent != lib_dir: + errors.append(f"{stem} resolves to {next(iter(paths))}, not into {lib_dir}") + return errors + + +def staged_library_sources(ldd_output: str, lib_dir: Path) -> list[Path]: + """Libraries ldd resolved outside lib_dir and outside the system directories. + + The integration binary's RUNPATH searches dist/lib, which is where it loads + amdhip64. A dependency that the build tree has and the base artifacts do + not (librocm_kpack.so.0) has to be copied here for the binary to start. + """ + system_roots = (Path("/lib"), Path("/lib64"), Path("/usr/lib"), Path("/usr/lib64")) + sources: list[Path] = [] + for match in re.finditer(r"^\s*\S+ => (\S+)", ldd_output, re.MULTILINE): + raw = match.group(1) + if raw == "not": + continue + src = Path(raw) + if src.parent == lib_dir: + continue + if any(src.is_relative_to(root) for root in system_roots): + continue + sources.append(src) + return sources + + +def stage_runtime_libs(binary: Path, lib_dir: Path) -> list[str]: + """Copy non-system libraries the binary loads into lib_dir. Returns names copied.""" + if not binary.is_file(): + return [] + out = subprocess.run( + ["ldd", str(binary)], check=True, capture_output=True, text=True + ).stdout + lib_dir.mkdir(parents=True, exist_ok=True) + copied: list[str] = [] + for src in staged_library_sources(out, lib_dir): + if not src.is_file(): + continue + dest = lib_dir / src.name + if dest.exists(): + continue + shutil.copy2(src, dest, follow_symlinks=True) + copied.append(dest.name) + print(f"staged {dest.name} from {src}") + return copied + + +def check_linkage(binary: Path, lib_dir: Path) -> int: + out = subprocess.run( + ["ldd", str(binary)], check=True, capture_output=True, text=True + ).stdout + errors = linkage_errors(out, lib_dir) + for error in errors: + print(f"::error::{binary.name}: {error}") + if not errors: + print(f"{binary.name} loads {' and '.join(LOADED_LIBS)} from {lib_dir}") + return 1 if errors else 0 + + +def embedded_targets(objdump_output: str) -> set[str]: + return set(re.findall(r"\bgfx[0-9a-f]+\b", objdump_output)) + + +def missing_targets(archs: str, objdump_output: str) -> list[str]: + found = embedded_targets(objdump_output) + return [a for a in archs.split(";") if a and a not in found] + + +def cmd_fetch_base(args: argparse.Namespace, runner: Runner = run) -> int: + base = json.loads(args.base_json.read_text()) if args.base_json.is_file() else None + if base is None: + base = find_base_run(args.therock) + args.base_json.parent.mkdir(parents=True, exist_ok=True) + args.base_json.write_text(json.dumps(base, indent=2) + "\n") + print(f"Base: {base['repo']} run {base['run_id']} at {base['commit']}") + runner(fetch_command(args.therock, base, args.dist, BUILD_ARTIFACTS)) + return 0 + + +def cmd_build(args: argparse.Namespace, runner: Runner = run) -> int: + runner(clr_configure_command(args.src, args.build, args.dist)) + runner(["cmake", "--build", args.build / "clr"]) + remove_amdhip64(args.dist / "lib") + runner(["cmake", "--install", args.build / "clr"]) + copy_amdhip64(args.dist / "lib", args.capture_lib) + runner(hrr_configure_command(args.src, args.build, args.dist, args.archs)) + runner(["cmake", "--build", args.build / "hrr", "--target", *HRR_TARGETS]) + stage_runtime_libs( + args.build / "hrr" / "tests" / "integration" / "hrr-integration-tests", + args.dist / "lib", + ) + return 0 + + +def cmd_verify(args: argparse.Namespace) -> int: + exe = args.build / "hrr" / "tests" / "integration" / "hrr-integration-tests" + run( + [ + sys.executable, + args.src / "projects/hrr/tests/scripts/hrr_test_config.py", + "check", + "--binary", + exe, + "--config", + args.src / "projects/hrr/tests/config/integration.yaml", + ] + ) + objdump = args.dist / "lib" / "llvm" / "bin" / "llvm-objdump" + out = subprocess.run( + [str(objdump), "--offloading", str(exe)], + check=True, + capture_output=True, + text=True, + ).stdout + missing = missing_targets(args.archs, out) + print("embedded: " + " ".join(sorted(embedded_targets(out)))) + if missing: + print(f"::error::missing from the fat binary: {' '.join(missing)}") + return 1 + return check_linkage(exe, args.dist / "lib") + + +def restore_executable_bits(build: Path) -> None: + """Artifact downloads drop the executable bit.""" + for rel in EXECUTABLES: + path = build / rel + path.chmod(path.stat().st_mode | 0o111) + + +def cmd_install_base(args: argparse.Namespace, runner: Runner = run) -> int: + base = json.loads(args.base_json.read_text()) + print(f"Base: {base['repo']} run {base['run_id']} at {base['commit']}") + runner(fetch_command(args.therock, base, args.dist, RUNTIME_ARTIFACTS)) + remove_amdhip64(args.dist / "lib") + copy_amdhip64(args.capture_lib, args.dist / "lib") + restore_executable_bits(args.build) + return check_linkage( + args.build / "hrr" / "tests" / "integration" / "hrr-integration-tests", + args.dist / "lib", + ) + + +def parse_args(argv: list[str] | None = None) -> argparse.Namespace: + parser = argparse.ArgumentParser(description=__doc__.splitlines()[0]) + parser.add_argument( + "command", + choices=("fetch-base", "build", "verify", "install-base"), + ) + parser.add_argument("--therock", type=Path, default=Path("therock")) + parser.add_argument("--src", type=Path, default=Path.cwd()) + parser.add_argument("--dist", type=Path, default=Path("dist")) + parser.add_argument("--build", type=Path, default=Path("build")) + parser.add_argument("--capture-lib", type=Path, default=Path("build/capture-lib")) + parser.add_argument("--base-json", type=Path, default=Path("build/base.json")) + parser.add_argument( + "--archs", + default="", + help=f"Semicolon-separated GPU targets; default targets.arch of {INTEGRATION_CONFIG}", + ) + args = parser.parse_args(argv) + for name in ("therock", "src", "dist", "build", "capture_lib", "base_json"): + setattr(args, name, getattr(args, name).resolve()) + if args.command in ("build", "verify") and not args.archs: + config = hrr_test_config.load(args.src / INTEGRATION_CONFIG) + args.archs = ";".join(config["targets"]["arch"]) + return args + + +def main(argv: list[str] | None = None) -> int: + args = parse_args(argv) + return { + "fetch-base": cmd_fetch_base, + "build": cmd_build, + "verify": cmd_verify, + "install-base": cmd_install_base, + }[args.command](args) + + +if __name__ == "__main__": + raise SystemExit(main()) diff --git a/projects/hrr/tests/scripts/hrr_ci_configure.py b/projects/hrr/tests/scripts/hrr_ci_configure.py new file mode 100644 index 000000000000..898b6ea409d0 --- /dev/null +++ b/projects/hrr/tests/scripts/hrr_ci_configure.py @@ -0,0 +1,626 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +"""Decide which HRR CI jobs run and on which GPU families. + +Writes these GitHub Actions outputs: + + run_cpu "true" when the API check, script and unit jobs run. + run_gpu "true" when the capture-enabled build and the GPU suite run. + families JSON list of {family, runner, backend, arch, config}. + simulator_families + The subset whose backend is simulator. + gpu_families The subset whose backend is gpu. + skipped_families JSON list of {"family", "reason"}: the other ci.yaml + families, when the suite runs. + reason One line explaining the decision, shown in the summary. + rocm_sdk_version The ROCm SDK the unit job installs. + +The policy lives in projects/hrr/tests/config/ci.yaml. The suite needs an +amdhip64 with HRR capture compiled in, which only the HRR GPU workflow builds. +It runs when a change touches a capture or suite path, on a pull request with +the capture label that touches any HRR path (hrr-ci.yml's paths filter starts +no run otherwise), nightly and on workflow_dispatch. + +The default backend is rocJitsu (simulator). Its jobs are +simulator.default_families, not the families marked enabled. The +workflow_dispatch families input replaces the selection. A simulator family +with no config is an error. ci:gfx labels are TheRock's, so this script does +not read them. The GPU backend +runs every enabled family on TheRock's runner. workflow_dispatch uses its +backend input; any other event uses backend.default. +""" + +from __future__ import annotations + +import argparse +import json +import os +import re +import subprocess +import sys +from dataclasses import dataclass, field +from pathlib import Path +from typing import Callable, Iterable + +import yaml + +DEFAULT_CONFIG = Path(__file__).resolve().parent.parent / "config" / "ci.yaml" +# Its targets.arch is what the GPU suite is built for. +INTEGRATION_CONFIG = DEFAULT_CONFIG.with_name("integration.yaml") + +EVENTS = ("pull_request", "push", "schedule", "workflow_dispatch") + +# The pull request files API returns at most this many files. A larger change +# is treated as touching everything rather than guessed at. +MAX_LISTED_FILES = 3000 + +NO_RUNNER = "no runner label in TheRock's runner config" + + +class ConfigError(ValueError): + pass + + +@dataclass(frozen=True) +class Family: + enabled: bool = True + + +@dataclass(frozen=True) +class Backend: + default: str + + +@dataclass(frozen=True) +class Simulator: + runner: str + default_families: tuple[str, ...] + configs: dict[str, str] + + +@dataclass(frozen=True) +class Config: + rocm_sdk_version: str + cpu_paths: tuple[str, ...] + capture_paths: tuple[str, ...] + suite_paths: tuple[str, ...] + excluded_paths: tuple[str, ...] + capture_label: str + backend: Backend + simulator: Simulator + families: dict[str, Family] + + +@dataclass +class Decision: + run_cpu: bool + run_gpu: bool + reason: str + families: list[dict] = field(default_factory=list) + skipped: list[dict] = field(default_factory=list) + + +def _mapping( + where: str, value, required: Iterable[str], optional: Iterable[str] = () +) -> dict: + if not isinstance(value, dict): + raise ConfigError(f"{where}: expected a mapping") + unknown = sorted(set(value) - set(required) - set(optional)) + if unknown: + raise ConfigError(f"{where}: unknown key {', '.join(map(str, unknown))}") + missing = sorted(set(required) - set(value)) + if missing: + raise ConfigError(f"{where}: missing {', '.join(missing)}") + return value + + +def _string(where: str, value) -> str: + if not isinstance(value, str) or not value: + raise ConfigError(f"{where}: expected a non-empty string") + return value + + +def _strings(where: str, value) -> tuple[str, ...]: + if not isinstance(value, list) or not value: + raise ConfigError(f"{where}: expected a non-empty list") + return tuple(_string(f"{where}[{i}]", v) for i, v in enumerate(value)) + + +def _bool(where: str, value) -> bool: + if not isinstance(value, bool): + raise ConfigError(f"{where}: expected true or false") + return value + + +def _backend(value) -> Backend: + value = _mapping("backend", value, ("default",)) + default = _string("backend.default", value["default"]) + if default not in ("simulator", "gpu"): + raise ConfigError( + f"backend.default: expected simulator or gpu, got {default!r}" + ) + return Backend(default) + + +def _simulator(value, families: dict) -> Simulator: + value = _mapping("simulator", value, ("runner", "default_families", "configs")) + default_families = _strings("simulator.default_families", value["default_families"]) + unknown = [key for key in default_families if key not in families] + if unknown: + raise ConfigError( + f"simulator.default_families: {', '.join(unknown)} not in families" + ) + configs = value["configs"] + if not isinstance(configs, dict) or not configs: + raise ConfigError("simulator.configs: expected a mapping") + unknown = [key for key in configs if key not in families] + if unknown: + raise ConfigError(f"simulator.configs: {', '.join(unknown)} not in families") + parsed = { + _string("simulator.configs", key): _string(f"simulator.configs.{key}", path) + for key, path in configs.items() + } + missing = [key for key in default_families if key not in parsed] + if missing: + raise ConfigError(f"simulator.configs: missing {', '.join(missing)}") + return Simulator( + runner=_string("simulator.runner", value["runner"]), + default_families=default_families, + configs=parsed, + ) + + +def _family(where: str, value) -> Family: + value = _mapping(where, value or {}, (), ("enabled",)) + return Family(enabled=_bool(f"{where}.enabled", value.get("enabled", True))) + + +def parse_config(data) -> Config: + data = _mapping( + "top level", + data, + ( + "version", + "rocm_sdk_version", + "paths", + "capture_label", + "backend", + "simulator", + "families", + ), + ) + if type(data["version"]) is not int or data["version"] != 1: + raise ConfigError(f"version: expected 1, got {data['version']!r}") + paths = _mapping("paths", data["paths"], ("cpu", "capture", "suite", "excluded")) + families = data["families"] + if not isinstance(families, dict) or not families: + raise ConfigError("families: expected a mapping with at least one family") + parsed_families = { + _string("families", key): _family(f"families.{key}", value) + for key, value in families.items() + } + return Config( + rocm_sdk_version=_string("rocm_sdk_version", data["rocm_sdk_version"]), + cpu_paths=_strings("paths.cpu", paths["cpu"]), + capture_paths=_strings("paths.capture", paths["capture"]), + suite_paths=_strings("paths.suite", paths["suite"]), + excluded_paths=_strings("paths.excluded", paths["excluded"]), + capture_label=_string("capture_label", data["capture_label"]), + backend=_backend(data["backend"]), + simulator=_simulator(data["simulator"], parsed_families), + families=parsed_families, + ) + + +def load_archs(path: Path) -> list[str]: + try: + data = yaml.safe_load(path.read_text(encoding="utf-8")) + return list(_strings("targets.arch", data["targets"]["arch"])) + except (ConfigError, KeyError, TypeError) as e: + raise ConfigError(f"{path}: no targets.arch list ({e})") from None + + +def load_config(path: Path) -> Config: + try: + return parse_config(yaml.safe_load(path.read_text(encoding="utf-8"))) + except ConfigError as e: + raise ConfigError(f"{path}: {e}") from None + + +def check_families(config: Config, matrix: dict) -> None: + unknown = [key for key in config.families if key not in matrix] + if unknown: + raise ConfigError( + f"families: {', '.join(unknown)} not in TheRock's family matrix" + f" (known: {', '.join(sorted(matrix))})" + ) + + +def family_targets(linux: dict) -> list[str]: + """The GPU targets of a family's test machines, without feature suffixes. + + TheRock lists them as fetch-gfx-targets; a family that names a single + target (gfx1150) may leave that list empty. + """ + targets = [t.split(":")[0] for t in linux.get("fetch-gfx-targets", [])] + family = str(linux.get("family", "")) + if not targets and re.fullmatch(r"gfx[0-9a-f]+", family): + targets = [family] + return list(dict.fromkeys(targets)) + + +def check_targets(config: Config, matrix: dict, archs: Iterable[str]) -> None: + """The suite must be built for every target of every family ci.yaml lists, + enabled or not, since a label can request any of them.""" + archs = set(archs) + problems = [] + for key in config.families: + targets = family_targets(matrix[key].get("linux", {})) + missing = [t for t in targets if t not in archs] + if not targets: + problems.append(f"TheRock lists no GPU targets for {key}") + elif missing: + problems.append(f"{key} needs {', '.join(missing)}") + if problems: + raise ConfigError( + f"targets.arch of {INTEGRATION_CONFIG.name} does not cover the families: " + + "; ".join(problems) + ) + + +def _matches(path: str, prefixes: Iterable[str]) -> bool: + return any(path == p or path.startswith(p) for p in prefixes) + + +def relevant_paths(paths: Iterable[str], config: Config) -> list[str]: + return [p for p in paths if not _matches(p, config.excluded_paths)] + + +def touches_cpu(paths: Iterable[str], config: Config) -> bool: + return any(_matches(p, config.cpu_paths) for p in relevant_paths(paths, config)) + + +def touches_capture(paths: Iterable[str], config: Config) -> bool: + return any(_matches(p, config.capture_paths) for p in relevant_paths(paths, config)) + + +def touches_suite(paths: Iterable[str], config: Config) -> bool: + return any(_matches(p, config.suite_paths) for p in relevant_paths(paths, config)) + + +def label_matters(label: str, config: Config) -> bool: + """Whether adding this label can change what runs. hrr-ci.yml's concurrency + group makes the same test, so that other labels do not cancel a run. + ci:gfx labels belong to TheRock and must not restart this workflow.""" + label = label.strip().lower() + return label == config.capture_label.lower() + + +def runner_for(linux: dict, pick_weighted: Callable[[list, str], str]) -> str: + labels = linux.get("test-runs-on-labels") + if labels: + return pick_weighted(labels, "hrr") + return linux.get("test-runs-on", "") + + +def resolve_backend(event: str, config: Config, backend_input: str) -> str: + """workflow_dispatch uses its input. Any other event uses the default.""" + if event == "workflow_dispatch" and backend_input: + if backend_input not in ("simulator", "gpu"): + raise ConfigError(f"backend {backend_input!r} is not simulator or gpu") + return backend_input + return config.backend.default + + +def _arch(matrix: dict, key: str) -> str: + return ",".join(family_targets(matrix.get(key, {}).get("linux", {}))) + + +def _job(family: str, runner: str, backend: str, arch: str, config: str = "") -> dict: + return { + "family": family, + "runner": runner, + "backend": backend, + "arch": arch, + "config": config, + } + + +def _requested(config: Config, requested: list[str]) -> set[str]: + unknown = [key for key in requested if key not in config.families] + if unknown: + raise ConfigError( + f"requested families {', '.join(unknown)} are not in ci.yaml" + f" (known: {', '.join(config.families)})" + ) + return set(requested) + + +def select_families( + config: Config, + requested: list[str], + matrix: dict, + pick_weighted: Callable[[list, str], str], +) -> tuple[list[dict], list[dict]]: + """GPU families to test and the other ci.yaml families with the reason, in + ci.yaml order.""" + if requested: + chosen = _requested(config, requested) + else: + chosen = {key for key, family in config.families.items() if family.enabled} + + families, skipped = [], [] + for key, family in config.families.items(): + if key not in chosen: + if requested: + reason = "not requested" + else: + reason = "disabled in ci.yaml" + skipped.append({"family": key, "reason": reason}) + continue + linux = matrix.get(key, {}).get("linux", {}) + runner = runner_for(linux, pick_weighted) + if not runner: + skipped.append({"family": key, "reason": NO_RUNNER}) + continue + families.append(_job(key, runner, "gpu", _arch(matrix, key))) + return families, skipped + + +def select_simulator( + config: Config, + requested: list[str], + matrix: dict, +) -> tuple[list[dict], list[dict]]: + """rocJitsu families. default_families, unless the dispatch input replaces + them. A family with no config is an error.""" + if requested: + chosen = _requested(config, requested) + else: + chosen = set(config.simulator.default_families) + missing = sorted(key for key in chosen if key not in config.simulator.configs) + if missing: + raise ConfigError(f"simulator has no config for {', '.join(missing)}") + + families, skipped = [], [] + for key in config.families: + if key not in chosen: + reason = "not requested" if requested else "not a default simulator family" + skipped.append({"family": key, "reason": reason}) + continue + families.append( + _job( + key, + config.simulator.runner, + "simulator", + _arch(matrix, key), + config.simulator.configs[key], + ) + ) + return families, skipped + + +def decide( + event: str, + paths: list[str] | None, + labels: list[str], + is_fork: bool, + requested: list[str], + config: Config, + matrix: dict, + pick_weighted: Callable[[list, str], str], + added_label: str | None = None, + backend: str = "", +) -> Decision: + """paths is None when the change set could not be listed in full. + + added_label is the label a labeled event added; the run for any other + pull request event already covers a label that cannot change what runs. + """ + if event not in EVENTS: + return Decision(False, False, f"event {event} does not run HRR CI") + if added_label is not None and not label_matters(added_label, config): + return Decision( + False, False, f"the label {added_label} does not change what HRR CI runs" + ) + + if event in ("schedule", "workflow_dispatch"): + run_cpu, capture, why = True, True, f"{event} run" + elif paths is None: + run_cpu, capture, why = True, True, "change set too large to list" + else: + run_cpu = touches_cpu(paths, config) + capture = touches_capture(paths, config) or touches_suite(paths, config) + if touches_capture(paths, config): + why = "the change touches capture or replay" + elif touches_suite(paths, config): + why = "the change touches the suite config" + elif run_cpu: + why = "the change touches HRR" + else: + why = "the change does not touch HRR" + + labelled = event == "pull_request" and config.capture_label.lower() in { + l.strip().lower() for l in labels + } + if labelled and not capture: + run_cpu, capture, why = ( + True, + True, + f"the pull request is labelled {config.capture_label}", + ) + + # Capture paths that the CPU list omits (hip_context.cpp, hip_code_object.*) + # still need the host unit suite. + run_cpu = run_cpu or capture + if not capture: + return Decision(run_cpu, False, why) + # A fork pull request gets no secrets and cannot use the real GPU runners, + # but the simulator jobs need neither (hrr-gpu.yml has read-only + # permissions), so only the gpu backend is refused. + resolved = resolve_backend(event, config, backend) + if is_fork and resolved == "gpu": + return Decision( + run_cpu, False, why + "; GPU suite skipped for a fork pull request" + ) + + if resolved == "simulator": + families, skipped = select_simulator(config, requested, matrix) + else: + families, skipped = select_families(config, requested, matrix, pick_weighted) + no_runner = [s["family"] for s in skipped if s["reason"] == NO_RUNNER] + if no_runner: + why += f"; no runner for {', '.join(no_runner)}" + if not families: + return Decision( + run_cpu, False, why + "; no GPU family with a runner was selected" + ) + return Decision(run_cpu, True, why, families, skipped) + + +def is_fork(event: str, head_repo: str, repo: str) -> bool: + """A pull request from another repository, or from one since deleted.""" + return event == "pull_request" and head_repo != repo + + +def list_pr_files(repo: str, number: str) -> list[str] | None: + out = subprocess.run( + [ + "gh", + "api", + "--paginate", + f"repos/{repo}/pulls/{number}/files", + "--jq", + ".[] | .filename, (.previous_filename // empty)", + ], + check=True, + capture_output=True, + text=True, + ).stdout + files = [line for line in out.splitlines() if line] + return None if len(files) >= MAX_LISTED_FILES else files + + +def list_push_files(repo: str, before: str, after: str) -> list[str] | None: + if not before or set(before) == {"0"}: + return None + out = subprocess.run( + [ + "gh", + "api", + f"repos/{repo}/compare/{before}...{after}", + "--jq", + ".files[] | .filename, (.previous_filename // empty)", + ], + check=True, + capture_output=True, + text=True, + ).stdout + files = [line for line in out.splitlines() if line] + # The compare API lists at most 300 files. + return None if len(files) >= 300 else files + + +def load_matrix(therock_dir: Path): + sys.path.insert(0, str(therock_dir / "build_tools" / "github_actions")) + import amdgpu_family_matrix # noqa: E402 + + families = amdgpu_family_matrix.get_all_families_for_trigger_types([]) + return families, amdgpu_family_matrix.select_weighted_label + + +def _dump(rows: list[dict]) -> str: + return json.dumps(rows, separators=(",", ":")) + + +def write_outputs( + decision: Decision, rocm_sdk_version: str, output_path: str | None +) -> None: + simulator = [row for row in decision.families if row.get("backend") == "simulator"] + gpu = [row for row in decision.families if row.get("backend") == "gpu"] + lines = [ + f"run_cpu={'true' if decision.run_cpu else 'false'}", + f"run_gpu={'true' if decision.run_gpu else 'false'}", + f"families={_dump(decision.families)}", + f"simulator_families={_dump(simulator)}", + f"gpu_families={_dump(gpu)}", + f"skipped_families={_dump(decision.skipped)}", + f"reason={decision.reason}", + f"rocm_sdk_version={rocm_sdk_version}", + ] + text = "\n".join(lines) + "\n" + if output_path: + with open(output_path, "a", encoding="utf-8") as out: + out.write(text) + sys.stdout.write(text) + + +def parse_args(argv: list[str] | None = None) -> argparse.Namespace: + parser = argparse.ArgumentParser(description=__doc__.splitlines()[0]) + parser.add_argument("--event", required=True) + parser.add_argument("--repo", required=True) + parser.add_argument("--pr-number", default="") + parser.add_argument("--before", default="") + parser.add_argument("--after", default="") + parser.add_argument("--labels", default="[]", help="JSON list of label names") + parser.add_argument("--action", default="", help="The pull_request event's action") + parser.add_argument("--label", default="", help="The label a labeled event added") + parser.add_argument( + "--head-repo", + default="", + help="owner/name of the pull request's head repository", + ) + parser.add_argument( + "--families", default="", help="Comma-separated family keys (workflow_dispatch)" + ) + parser.add_argument( + "--backend", + default="", + choices=("", "simulator", "gpu"), + help="workflow_dispatch backend; empty uses ci.yaml", + ) + parser.add_argument("--therock-dir", type=Path, required=True) + parser.add_argument("--config", type=Path, default=DEFAULT_CONFIG) + parser.add_argument("--integration-config", type=Path, default=INTEGRATION_CONFIG) + return parser.parse_args(argv) + + +def main(argv: list[str] | None = None) -> int: + args = parse_args(argv) + labels = json.loads(args.labels or "[]") + requested = [f.strip().lower() for f in args.families.split(",") if f.strip()] + + try: + config = load_config(args.config) + matrix, pick_weighted = load_matrix(args.therock_dir) + check_families(config, matrix) + check_targets(config, matrix, load_archs(args.integration_config)) + + paths: list[str] | None = [] + if args.event == "pull_request" and args.pr_number: + paths = list_pr_files(args.repo, args.pr_number) + elif args.event == "push": + paths = list_push_files(args.repo, args.before, args.after) + + decision = decide( + args.event, + paths, + labels, + is_fork(args.event, args.head_repo, args.repo), + requested, + config, + matrix, + pick_weighted, + added_label=args.label if args.action == "labeled" else None, + backend=args.backend, + ) + except ConfigError as e: + print(f"::error::{e}") + return 1 + write_outputs(decision, config.rocm_sdk_version, os.environ.get("GITHUB_OUTPUT")) + return 0 + + +if __name__ == "__main__": + raise SystemExit(main()) diff --git a/projects/hrr/tests/scripts/hrr_ci_report.py b/projects/hrr/tests/scripts/hrr_ci_report.py new file mode 100644 index 000000000000..a92f5558a18a --- /dev/null +++ b/projects/hrr/tests/scripts/hrr_ci_report.py @@ -0,0 +1,345 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +"""Write the HRR CI summary for a run and decide its overall result. + +The summary has three parts: + + * why this run tested what it did, and which base build the GPU suite used; + * the conclusion of every unit, build and GPU job, so a job that never + produced results (cancelled, timed out, still queued) still has a row; + * the per-suite results table from summarize_junit.py, with a row for every + ci.yaml family, saying why when one did not run. + +Exits 1 when the configure job did not succeed, a needed job failed, was +cancelled, or was skipped after configure selected it, any listed HRR job is +not in a terminal success state, a JUnit suite has failed cases, or a results +file could not be read. Job conclusions from the Actions API are links in the +table. The verdict uses --needs-json (the workflow's toJSON(needs)) and the +JUnit files, because the jobs API can still say queued after those exist. + +With --pass-through true (a label run that ran no HRR job) it reports the +newest other run on --head-sha that has a verdict instead, and exits 1 unless +that run's HRR jobs and its own summary succeeded, or when it cannot be read. +""" + +from __future__ import annotations + +import argparse +import json +import os +import re +import subprocess +import sys +from pathlib import Path + +import summarize_junit + +JOB_PATTERNS = ( + ("API checks", re.compile(r"^API IDs and workloads$")), + ("Script tests", re.compile(r"^HRR script tests$")), + ("Unit", re.compile(r"(?:^|/ )Unit \((?P[^)]+)\)$")), + ("Build", re.compile(r"(?:^|/ )Build capture-enabled HIP and HRR$")), + # The real-GPU job is "Integration ()", the rocJitsu one "rocJitsu ()". + ("Integration", re.compile(r"(?:^|/ )(?:Integration|rocJitsu) \((?P[^)]+)\)$")), +) +# Anything else (queued, in_progress, waiting, unknown) has not finished. +TERMINAL_OK = {"success", "neutral"} +BAD_NEEDS = {"failure", "cancelled"} +CPU_NEEDS = ("api-checks", "scripts", "unit") + + +def classify_jobs(jobs: list[dict]) -> list[dict]: + """HRR jobs from the run's job list, as {kind, name, conclusion, url}.""" + rows = [] + for job in jobs: + # A skipped matrix job keeps its unexpanded name; it is no family. + if "${{" in job.get("name", ""): + continue + for kind, pattern in JOB_PATTERNS: + match = pattern.search(job.get("name", "")) + if match: + rows.append( + { + "kind": kind, + "name": match.groupdict().get("name") or "", + "conclusion": job.get("conclusion") or job.get("status") or "unknown", + "url": job.get("html_url", ""), + } + ) + break + return rows + + +def missing_families(families: list[dict], rows: list[dict]) -> list[str]: + seen = {r["name"] for r in rows if r["kind"] == "Integration"} + return [f["family"] for f in families if f["family"] not in seen] + + +def render_header( + reason: str, run_gpu: bool, families: list[dict], base: dict | None, + configure_result: str = "success", +) -> str: + if configure_result != "success": + return ( + f"## HRR CI\n\n- Configure: {configure_result}, so nothing else ran;" + " see the Configure job's log.\n\n" + ) + lines = ["## HRR CI", "", f"- Scope: {reason}."] + if run_gpu: + names = ", ".join( + f"{f['family']} (`{f['runner']}`)" if f.get("runner") else f["family"] + for f in families + ) or "none" + lines.append(f"- GPU families and runners: {names}.") + if base: + lines.append( + f"- Base ROCm artifacts: [{base['repo']} run {base['run_id']}]({base['run_url']})" + f" at `{base['commit'][:12]}`; amdhip64 rebuilt from this commit with capture enabled." + ) + else: + lines.append("- GPU suite: not run.") + return "\n".join(lines) + "\n\n" + + +def render_jobs(rows: list[dict], missing: list[str]) -> str: + if not rows and not missing: + return "" + lines = ["### Jobs", "", "| Job | Target | Conclusion |", "| --- | --- | --- |"] + for row in rows: + conclusion = f"[{row['conclusion']}]({row['url']})" if row["url"] else row["conclusion"] + lines.append(f"| {row['kind']} | {row['name'] or '—'} | {conclusion} |") + for family in missing: + lines.append(f"| Integration | {family} | _did not run_ |") + return "\n".join(lines) + "\n\n" + + +def overall_ok(rows: list[dict], missing: list[str]) -> bool: + """True when every requested family ran and every listed job finished + successfully. A skipped, failed, or still-queued job is not a pass: the + jobs API has reported queued while the suite had already failed.""" + if missing: + return False + return all(r["conclusion"] in TERMINAL_OK for r in rows) + + +def need_result(needs: dict, name: str) -> str: + info = needs.get(name) + if isinstance(info, dict) and info.get("result"): + return str(info["result"]) + return "missing" + + +def check_needs(needs: dict, *, run_gpu: bool) -> list[str]: + """Problems in GitHub's toJSON(needs). A failure or cancellation fails the + run. On a run whose configure succeeded, the CPU jobs were selected, and + the GPU workflow was selected when run_gpu is true; a skip there is a job + that was supposed to run.""" + problems = [] + for name in needs: + result = need_result(needs, name) + if result in BAD_NEEDS: + problems.append(f"{name} is {result}") + selected = list(CPU_NEEDS) + if run_gpu: + selected.append("gpu") + for name in selected: + result = need_result(needs, name) + if result != "success" and f"{name} is {result}" not in problems: + problems.append(f"{name} is {result}") + return problems + + +SUMMARY_JOB = "HRR CI Summary" + + +def tested_commit(jobs: list[dict], rows: list[dict]) -> bool: + """A run whose configure selected nothing still lists its HRR jobs, as + skipped, and tested nothing. One whose configure failed, or has not + finished, skipped them too but has a verdict of its own to report. A run + that lists no jobs yet is passed over; its own summary comes later and is + the newer check.""" + configure = next((j for j in jobs if j.get("name") == "Configure"), {}) + if (configure.get("conclusion") or configure.get("status")) not in (None, "success", "skipped"): + return True + return any(r["conclusion"] != "skipped" for r in rows) + + +def pass_through_ok(rows: list[dict], summary_conclusion: str) -> tuple[bool, str]: + """Verdict on an earlier run. Its own summary decides, since that is the + check a label run would otherwise hide, and every HRR job must also have + finished without failing: a label run must never turn a failing HRR CI + into a passing one.""" + for row in rows: + if row["conclusion"] not in ("success", "skipped", "neutral"): + return False, f"{row['kind']} {row['name'] or row['conclusion']} is {row['conclusion']}" + if summary_conclusion != "success": + return False, f"{SUMMARY_JOB} is {summary_conclusion or 'missing'}" + return True, "" + + +def gh_api(path: str, jq: str) -> list: + """The values `gh api --paginate --jq` prints for path, one per jq result.""" + out = subprocess.run( + ["gh", "api", "--paginate", path, "--jq", jq], + check=True, capture_output=True, text=True, + ).stdout + decoder = json.JSONDecoder() + values, pos = [], 0 + while pos < len(out): + while pos < len(out) and out[pos].isspace(): + pos += 1 + if pos >= len(out): + break + value, pos = decoder.raw_decode(out, pos) + values.append(value) + return values + + +def list_jobs(repo: str, run_id: str, attempt: str) -> list[dict]: + return gh_api(f"repos/{repo}/actions/runs/{run_id}/attempts/{attempt}/jobs", ".jobs[]") + + +def find_tested_run(repo: str, run_id: str, head_sha: str) -> tuple[dict, list[dict], str] | None: + """The newest other pull_request run of this workflow on head_sha that + tested it, with its HRR jobs and its summary's conclusion. Its latest + attempt counts, as it does for the checks GitHub shows. Label runs that + ran nothing themselves are passed over.""" + [workflow_id] = gh_api(f"repos/{repo}/actions/runs/{run_id}", ".workflow_id") + runs = gh_api( + f"repos/{repo}/actions/workflows/{workflow_id}/runs" + f"?event=pull_request&head_sha={head_sha}&per_page=100", + ".workflow_runs[]", + ) + runs = [r for r in runs if str(r["id"]) != str(run_id)] + # A re-run keeps created_at; run_started_at moves with the latest attempt, + # whose check runs are the newest on the commit. + runs.sort(key=lambda r: r.get("run_started_at") or r["created_at"], reverse=True) + for run in runs: + jobs = list_jobs(repo, str(run["id"]), str(run["run_attempt"])) + rows = classify_jobs(jobs) + if tested_commit(jobs, rows): + summary = next((j for j in jobs if j.get("name") == SUMMARY_JOB), {}) + return run, rows, summary.get("conclusion") or summary.get("status") or "" + return None + + +def pass_through(repo: str, run_id: str, head_sha: str) -> int: + """Summary for a label run that ran no HRR job: the verdict of the run + that tested the commit, and never greener than it. Passes when no run + tested the commit, since there is then no failure to hide. Fails while + that run is still in progress; its own summary, written later, is then + the newer check.""" + try: + found = find_tested_run(repo, run_id, head_sha) + except (subprocess.CalledProcessError, ValueError) as e: + detail = (getattr(e, "stderr", "") or "").strip() or str(e) + print(f"::error::could not find the HRR CI run that tested {head_sha}: {detail}") + return 1 + if found is None: + print(f"No earlier HRR CI run ran HRR jobs on {head_sha}; nothing to report.") + return 0 + run, rows, summary_conclusion = found + markdown = ( + f"## HRR CI\n\n- This label changes nothing that runs; the verdict is that of " + f"[run {run['id']} attempt {run['run_attempt']}]({run.get('html_url', '')}), " + f"which tested this commit.\n\n" + ) + render_jobs(rows, []) + sys.stdout.write(markdown) + summary = os.environ.get("GITHUB_STEP_SUMMARY") + if summary: + with open(summary, "a", encoding="utf-8") as out: + out.write(markdown) + ok, why = pass_through_ok(rows, summary_conclusion) + if not ok: + print(f"::error::HRR CI on this commit reported {why}; " + f"a label event must not turn it into a pass") + return 1 + return 0 + + +def main(argv: list[str] | None = None) -> int: + parser = argparse.ArgumentParser(description=__doc__.splitlines()[0]) + parser.add_argument("--configure-result", default="success") + parser.add_argument("--reason", default="") + parser.add_argument("--run-gpu", default="false") + parser.add_argument("--families", default="[]") + parser.add_argument("--skipped-families", default="[]") + parser.add_argument("--results-dir", type=Path, default=Path("results")) + parser.add_argument("--base-json", type=Path) + parser.add_argument("--repo", default=os.environ.get("GITHUB_REPOSITORY", "")) + parser.add_argument("--run-id", default=os.environ.get("GITHUB_RUN_ID", "")) + parser.add_argument("--run-attempt", default=os.environ.get("GITHUB_RUN_ATTEMPT", "1")) + parser.add_argument("--needs-json", default="", + help="GitHub toJSON(needs). Empty skips the needs check") + parser.add_argument("--pass-through", choices=("true", "false"), default="false", + help="true: report the earlier run that tested --head-sha instead") + # Not GITHUB_SHA: on a pull request that is the merge commit, which no + # earlier run is listed under. + parser.add_argument("--head-sha", default="") + args = parser.parse_args(argv) + + run_gpu = args.run_gpu == "true" + families = json.loads(args.families or "[]") if run_gpu else [] + skipped = json.loads(args.skipped_families or "[]") if run_gpu else [] + base = None + if args.base_json and args.base_json.is_file(): + base = json.loads(args.base_json.read_text()) + + if args.pass_through == "true": + if not args.head_sha: + parser.error("--pass-through true needs --head-sha, the pull request head") + return pass_through(args.repo, args.run_id, args.head_sha) + + try: + jobs = list_jobs(args.repo, args.run_id, args.run_attempt) + except subprocess.CalledProcessError as e: + detail = (e.stderr or "").strip() or f"gh api exited {e.returncode}" + print(f"::error::could not list the jobs of run {args.run_id}: {detail}") + return 1 + rows = classify_jobs(jobs) + missing = missing_families(families, rows) + + configured = args.configure_result == "success" + markdown = render_header(args.reason, run_gpu, families, base, args.configure_result) + markdown += render_jobs(rows, missing) + errors: list[str] = [] + failed_suites: list[str] = [] + args.results_dir.mkdir(parents=True, exist_ok=True) + table_families = [(f["family"], "") for f in families] + table_families += [(s["family"], s["reason"]) for s in skipped] + markdown += summarize_junit.render( + args.results_dir, errors, table_families, failures=failed_suites + ) + + sys.stdout.write(markdown) + summary = os.environ.get("GITHUB_STEP_SUMMARY") + if summary: + with open(summary, "a", encoding="utf-8") as out: + out.write(markdown) + for error in errors: + print(f"::error::unreadable JUnit XML {error}") + for suite in failed_suites: + print(f"::error::failed cases in {suite}") + + needs_problems: list[str] = [] + if args.needs_json and configured: + try: + needs = json.loads(args.needs_json) + except json.JSONDecodeError as error: + print(f"::error::could not read --needs-json: {error}") + return 1 + if not isinstance(needs, dict): + print("::error::--needs-json must be an object") + return 1 + needs_problems = check_needs(needs, run_gpu=run_gpu) + for problem in needs_problems: + print(f"::error::{problem}") + return 0 if ( + configured and not needs_problems and overall_ok(rows, missing) + and not errors and not failed_suites + ) else 1 + + +if __name__ == "__main__": + raise SystemExit(main()) diff --git a/projects/hrr/tests/scripts/hrr_test_config.py b/projects/hrr/tests/scripts/hrr_test_config.py index 16cbfdb64bfd..82faa9caa3bd 100644 --- a/projects/hrr/tests/scripts/hrr_test_config.py +++ b/projects/hrr/tests/scripts/hrr_test_config.py @@ -11,11 +11,13 @@ A rule matches when every key it sets matches. Keys in one ``when`` are ANDed; a list under a key is ORed; a missing key matches anything. Glob patterns (``gfx11*``) are allowed. Several visible GPUs match when any of them does. +``emulator: true`` matches only a run under a GPU emulator (run_catch2.py +``--launcher``), ``emulator: false`` only a run on a real device. ``kind: disabled`` is a temporary regression and needs a public issue URL. ``kind: unsupported`` is permanent and needs only a reason. Usage: - hrr_test_config.py lint CONFIG [CONFIG ...] [--workflow hrr-ci.yml] + hrr_test_config.py lint CONFIG [CONFIG ...] (their targets.arch must agree) hrr_test_config.py check --binary EXE --config CONFIG [--update] """ @@ -37,7 +39,7 @@ ARCH_TOKEN = re.compile(r"gfx[0-9][0-9a-z]*") WORKLOAD_TAGS = {"hrr-direct", "direct"} KINDS = {"disabled", "unsupported"} -WHEN_KEYS = {"os", "arch"} +WHEN_KEYS = {"os", "arch", "emulator"} RULE_KEYS = {"when", "kind", "reason", "issue"} CASE_KEYS = {"skip"} TOP_KEYS = {"version", "suite", "binary", "select", "targets", "skip", "cases"} @@ -47,6 +49,8 @@ # # A skip rule matches when every key under "when" matches (a list is OR, a # missing key matches anything, and a value may be a glob such as gfx11*). +# "emulator: true" matches only a run under a GPU emulator (run_catch2.py +# --launcher), "emulator: false" only a run on a real device. # kind: disabled is a temporary regression and requires a public issue URL. # kind: unsupported is permanent and needs only a reason. # @@ -96,16 +100,21 @@ def value_matches(pattern, value: str) -> bool: return any(fnmatch.fnmatchcase(value, str(item)) for item in _as_list(pattern)) -def rule_matches(rule: dict, os_name: str, archs: list[str]) -> list[str] | None: +def rule_matches( + rule: dict, os_name: str, archs: list[str], emulated: bool = False +) -> list[str] | None: """The architectures that made the rule match, or None. An ``arch`` key matches when any visible architecture matches, and does - not match when no architecture was supplied. An empty ``when`` matches - every machine. + not match when no architecture was supplied. An ``emulator`` key matches + a run under a GPU emulator (true) or on a real device (false). An empty + ``when`` matches every machine. """ when = rule.get("when") or {} if "os" in when and not value_matches(when["os"], os_name): return None + if "emulator" in when and when["emulator"] != emulated: + return None matched = list(archs) if "arch" in when: matched = [arch for arch in archs if value_matches(when["arch"], arch)] @@ -120,7 +129,7 @@ def _rules_for(config: dict, name: str) -> list[dict]: def resolve(config: dict, names: list[str], os_name: str, archs: list[str], - ignore: bool = False) -> tuple[list[str], list[dict]]: + ignore: bool = False, emulated: bool = False) -> tuple[list[str], list[dict]]: """Split ``names`` into cases to run and cases a rule skips. The first matching rule wins, suite rules before the case's own. Each @@ -135,7 +144,7 @@ def resolve(config: dict, names: list[str], os_name: str, archs: list[str], hit = None matched: list[str] = [] for rule in _rules_for(config, name): - matched = rule_matches(rule, os_name, archs) + matched = rule_matches(rule, os_name, archs, emulated) if matched is not None: hit = rule break @@ -195,6 +204,8 @@ def _check_rule(errors: list[str], where: str, rule, targets: dict) -> None: for key in ("os", "arch"): if key in when: _check_pattern(errors, where, key, when[key], targets.get(key, [])) + if "emulator" in when and not isinstance(when["emulator"], bool): + _error(errors, where, "'emulator' must be true or false") kind = rule.get("kind") if kind not in KINDS: _error(errors, where, "'kind' must be disabled or unsupported") @@ -208,12 +219,8 @@ def _check_rule(errors: list[str], where: str, rule, targets: dict) -> None: _error(errors, where, "'issue' must be an https:// URL") -def lint(config: dict, cases: list[dict] | None = None, - workflow_archs: list[str] | None = None) -> list[str]: - """Schema and vocabulary errors. ``cases`` enables the workload-rule check. - - ``workflow_archs``, when given, must be the same set as ``targets.arch``. - """ +def lint(config: dict, cases: list[dict] | None = None) -> list[str]: + """Schema and vocabulary errors. ``cases`` enables the workload-rule check.""" errors: list[str] = [] unknown = set(config) - TOP_KEYS if unknown: @@ -234,14 +241,6 @@ def lint(config: dict, cases: list[dict] | None = None, isinstance(item, str) and item for item in values): _error(errors, "targets", f"'{key}' must be a non-empty list of strings") targets[key] = [] - if workflow_archs is not None and targets.get("arch") is not None: - got = set(targets["arch"]) - want = set(workflow_archs) - if got != want: - missing = ", ".join(sorted(want - got)) or "-" - extra = ", ".join(sorted(got - want)) or "-" - _error(errors, "targets.arch", - f"does not match HRR_ARCHS (missing: {missing}; extra: {extra})") suite_skip = config.get("skip", []) if not isinstance(suite_skip, list): _error(errors, "skip", "must be a list") @@ -276,6 +275,29 @@ def lint(config: dict, cases: list[dict] | None = None, return errors +def arch_mismatches(configs: dict[str, dict]) -> list[str]: + """The suites share one architecture vocabulary, so every ``targets.arch`` + must name the same set as the first config's.""" + archs = { + path: set(config["targets"]["arch"]) + for path, config in configs.items() + if isinstance(config.get("targets"), dict) + and isinstance(config["targets"].get("arch"), list) + } + if len(archs) < 2: + return [] + first, *rest = archs + errors = [] + for path in rest: + if archs[path] != archs[first]: + missing = ", ".join(sorted(archs[first] - archs[path])) or "-" + extra = ", ".join(sorted(archs[path] - archs[first])) or "-" + errors.append( + f"targets.arch of {path} differs from {first} (missing: {missing}; extra: {extra})" + ) + return errors + + def list_cases(binary: str, select: str = "*") -> list[dict]: """Every case ``select`` matches, hidden ones included when ``select`` is ``*``. @@ -334,6 +356,8 @@ def _yaml_pattern(value) -> str: def _emit_rule(lines: list[str], rule: dict, indent: str) -> None: when = rule.get("when") or {} parts = [f"{key}: {_yaml_pattern(when[key])}" for key in ("os", "arch") if key in when] + if "emulator" in when: + parts.append(f"emulator: {'true' if when['emulator'] else 'false'}") lines.append(f"{indent}- when: {{{', '.join(parts)}}}") lines.append(f"{indent} kind: {rule['kind']}") lines.append(f"{indent} reason: {_yaml_str(rule['reason'])}") @@ -382,22 +406,23 @@ def apply_listing(config: dict, names: list[str]) -> tuple[list[str], list[str]] return added, removed -def workflow_archs(path: str | Path) -> list[str]: - """The ``HRR_ARCHS`` list from the HRR CI workflow.""" - try: - text = Path(path).read_text(encoding="utf-8") - except OSError as exc: - raise ConfigError(f"cannot read {path}: {exc}") from exc - match = re.search(r'^ {2}HRR_ARCHS:\s*"([^"]+)"\s*$', text, re.MULTILINE) - if not match: - raise ConfigError(f"no HRR_ARCHS list in {path}") - return [arch for arch in match.group(1).split(";") if arch] - - def detect_os() -> str: return "windows" if os.name == "nt" else "linux" +def compiled_out_here(config: dict, name: str, os_name: str) -> bool: + """True when an OS-wide ``unsupported`` rule covers the case on ``os_name``. + + Such a case may be compiled out of the binary there (an ``#ifndef _WIN32`` + around the whole file), so ``check`` does not call its absence MISSING. + A rule that names an architecture is not OS-wide and does not count. + """ + return any( + rule.get("kind") == "unsupported" and rule_matches(rule, os_name, []) is not None + for rule in _rules_for(config, name) + ) + + def _archs_in(text: str) -> list[str]: found = [] for line in text.splitlines(): @@ -447,8 +472,8 @@ def _print_errors(path: str, errors: list[str]) -> None: def _cmd_lint(args: argparse.Namespace) -> int: - archs = workflow_archs(args.workflow) if args.workflow else None failed = False + loaded = {} for path in args.configs: try: config = load(path) @@ -456,12 +481,16 @@ def _cmd_lint(args: argparse.Namespace) -> int: print(f"FAIL: {exc}", file=sys.stderr) failed = True continue - errors = lint(config, workflow_archs=archs) + loaded[path] = config + errors = lint(config) if errors: _print_errors(path, errors) failed = True else: print(f"OK: {path}") + for error in arch_mismatches(loaded): + print(f"FAIL: {error}", file=sys.stderr) + failed = True return 1 if failed else 0 @@ -484,10 +513,17 @@ def _cmd_check(args: argparse.Namespace) -> int: f"({len(added)} added, {len(removed)} removed)") for name in removed: print(f" removed {name}") - expected = set(config["cases"]) + os_name = detect_os() + # MISSING is measured against the cases an OS-wide unsupported rule may + # have compiled out here; UNEXPECTED against every case the config + # knows. Subtracting the relaxed set from the binary's list would blame + # a case that is legitimately still compiled into the binary. + unsupported = {name for name in config["cases"] + if compiled_out_here(config, name, os_name)} + expected = {name for name in config["cases"] if name not in unsupported} actual = set(names) missing = sorted(expected - actual) - unexpected = sorted(actual - expected) + unexpected = sorted(actual - set(config["cases"])) errors = lint(config, cases=found) if not missing and not unexpected and not errors: print(f"OK: {len(actual)} test cases present, matching {args.config}") @@ -514,7 +550,6 @@ def main(argv: list[str] | None = None) -> int: lint_cmd = commands.add_parser("lint", help="validate config files") lint_cmd.add_argument("configs", nargs="+") - lint_cmd.add_argument("--workflow", help="HRR workflow whose HRR_ARCHS must match targets.arch") lint_cmd.set_defaults(func=_cmd_lint) check_cmd = commands.add_parser("check", help="compare a binary's cases to a config") diff --git a/projects/hrr/tests/scripts/run_catch2.py b/projects/hrr/tests/scripts/run_catch2.py index 5375eb38d971..606296cf6083 100644 --- a/projects/hrr/tests/scripts/run_catch2.py +++ b/projects/hrr/tests/scripts/run_catch2.py @@ -24,15 +24,27 @@ def parse_args(argv: list[str] | None = None) -> argparse.Namespace: type=float, help="seconds before the suite and its workloads are killed", ) - parser.add_argument("--config", type=Path, help="suite YAML; decides which cases run") - parser.add_argument("--os", dest="os_name", choices=("auto", "linux", "windows"), - help="OS the rules see; default auto when --config is set") + parser.add_argument( + "--config", type=Path, help="suite YAML; decides which cases run" + ) + parser.add_argument( + "--os", + dest="os_name", + choices=("auto", "linux", "windows"), + help="OS the rules see; default auto when --config is set", + ) parser.add_argument("--arch", help="auto, none, or comma-separated gfx names") - parser.add_argument("--select", help="Catch2 selector; overrides the one in the config") - parser.add_argument("--ignore-config", action="store_true", - help="run the cases the config would skip") parser.add_argument( - "--launcher", default="", + "--select", help="Catch2 selector; overrides the one in the config" + ) + parser.add_argument( + "--ignore-config", + action="store_true", + help="run the cases the config would skip", + ) + parser.add_argument( + "--launcher", + default="", help="Prefix for the suite process. Case listing still uses the Catch2 binary.", ) parser.add_argument("command", nargs=argparse.REMAINDER) @@ -41,7 +53,9 @@ def parse_args(argv: list[str] | None = None) -> argparse.Namespace: args.command = args.command[1:] if not args.command: parser.error("a Catch2 command is required after --") - if args.config is None and (args.os_name or args.arch or args.select or args.ignore_config): + if args.config is None and ( + args.os_name or args.arch or args.select or args.ignore_config + ): parser.error("--os, --arch, --select and --ignore-config require --config") return args @@ -98,7 +112,10 @@ def _ignore_config(args: argparse.Namespace) -> bool: if args.ignore_config: return True return os.environ.get("HRR_TEST_IGNORE_CONFIG", "").strip().lower() in ( - "1", "true", "yes", "on", + "1", + "true", + "yes", + "on", ) @@ -117,16 +134,22 @@ def _archs_for(arch_arg: str, ignore: bool) -> list[str]: return [] if arch_arg == "auto": import hrr_test_config + archs = hrr_test_config.detect_archs() if not archs: - print("error: could not detect a GPU architecture; " - "pass --arch none or --arch gfxNNNN", file=sys.stderr) + print( + "error: could not detect a GPU architecture; " + "pass --arch none or --arch gfxNNNN", + file=sys.stderr, + ) raise SystemExit(1) return archs archs = _parse_arch(arch_arg) if not archs or not all(arch.startswith("gfx") for arch in archs): - print(f"error: --arch must be auto, none, or gfx names, not {arch_arg!r}", - file=sys.stderr) + print( + f"error: --arch must be auto, none, or gfx names, not {arch_arg!r}", + file=sys.stderr, + ) raise SystemExit(1) return archs @@ -150,13 +173,17 @@ def _add_skip(suite: ET.Element, name: str, message: str) -> None: def _write_junit(path: Path, skips: list[tuple[str, str]]) -> None: root = ET.Element("testsuites") - suite = ET.SubElement(root, "testsuite", { - "name": "hrr-config", - "tests": str(len(skips)), - "failures": "0", - "errors": "0", - "skipped": str(len(skips)), - }) + suite = ET.SubElement( + root, + "testsuite", + { + "name": "hrr-config", + "tests": str(len(skips)), + "failures": "0", + "errors": "0", + "skipped": str(len(skips)), + }, + ) for name, message in skips: _add_skip(suite, name, message) tree = ET.ElementTree(root) @@ -182,7 +209,9 @@ def _inject_skips(path: Path, skips: list[tuple[str, str]]) -> None: tree.write(path, encoding="unicode", xml_declaration=True) -def _apply_config(args: argparse.Namespace) -> tuple[list[str], list[tuple[str, str]], str | None]: +def _apply_config( + args: argparse.Namespace, +) -> tuple[list[str], list[tuple[str, str]], str | None]: """Return the Catch2 command, the config skips, and the case-list file. An empty command means every selected case was skipped, so Catch2 is not @@ -217,16 +246,25 @@ def _apply_config(args: argparse.Namespace) -> tuple[list[str], list[tuple[str, # selector is Catch2's. Listing still uses command[0], which is the binary # in CI. Tests mock list_cases. names = [case["name"] for case in listed] - _run, skipped = hrr_test_config.resolve(config, names, os_name, archs, ignore=ignore) + _run, skipped = hrr_test_config.resolve( + config, names, os_name, archs, ignore=ignore, emulated=bool(args.launcher) + ) records = [ - (item["name"], hrr_test_config.skip_message(os_name, item["archs"], item["rule"])) + ( + item["name"], + hrr_test_config.skip_message(os_name, item["archs"], item["rule"]), + ) for item in skipped ] if not _run: return [], records, None prefix = _strip_filters(args.command) handle = tempfile.NamedTemporaryFile( - "w", prefix="hrr-cases-", suffix=".txt", delete=False, encoding="utf-8", + "w", + prefix="hrr-cases-", + suffix=".txt", + delete=False, + encoding="utf-8", ) with handle: for name in _run: @@ -234,6 +272,13 @@ def _apply_config(args: argparse.Namespace) -> tuple[list[str], list[tuple[str, return prefix + ["--input-file", handle.name], records, handle.name +def launch_command(launcher: str, command: list[str]) -> list[str]: + """Prefix the suite process. Listing uses command[0], the Catch2 binary.""" + if not launcher: + return command + return shlex.split(launcher) + command + + def _report(root: ET.Element) -> tuple[dict[str, int], str]: """Print one line per case. Returns the counts and the joined stdout.""" labels = {"PASS": "PASS", "FAIL": "FAIL", "SKIP": "SKIP", "CONFIG": "SKIP (config)"} @@ -290,16 +335,18 @@ def _run_suite(args: argparse.Namespace, state: dict) -> int: return 0 else: command = list(args.command) - if args.launcher: - command = shlex.split(args.launcher) + command - command = command + [ - "--durations", - "yes", - "--reporter", - "console", - "--reporter", - f"junit::out={args.xml}", - ] + command = launch_command( + args.launcher, + command + + [ + "--durations", + "yes", + "--reporter", + "console", + "--reporter", + f"junit::out={args.xml}", + ], + ) # A session of its own, so a timeout also kills the workloads the suite # spawned: they hold the pipe open and would block the read forever. with subprocess.Popen( diff --git a/projects/hrr/tests/scripts/summarize_junit.py b/projects/hrr/tests/scripts/summarize_junit.py index a183e7430c02..58ab8f081d10 100644 --- a/projects/hrr/tests/scripts/summarize_junit.py +++ b/projects/hrr/tests/scripts/summarize_junit.py @@ -19,19 +19,10 @@ import xml.etree.ElementTree as ET from pathlib import Path -# Display order matches .github/workflows/hrr-ci.yml. -COLUMNS = ( +# The unit job's matrix in .github/workflows/hrr-ci.yml. +UNIT_COLUMNS = ( ("unit-ubuntu-24.04.xml", "Unit", "Linux"), ("unit-windows-2022.xml", "Unit", "Windows"), - ("integration-gfx90a.xml", "Integration", "gfx90a"), - ("integration-gfx94x.xml", "Integration", "gfx94x"), - ("integration-gfx950.xml", "Integration", "gfx950"), - ("integration-gfx103x.xml", "Integration", "gfx103x"), - ("integration-gfx110x.xml", "Integration", "gfx110x"), - ("integration-gfx120x.xml", "Integration", "gfx120x"), - ("integration-gfx1151.xml", "Integration", "gfx1151"), - ("integration-gfx1150.xml", "Integration", "gfx1150"), - ("integration-gfx1153.xml", "Integration", "gfx1153"), ) MAX_FAILED_NAMES = 8 @@ -114,19 +105,41 @@ def failed_cell(names: list[str]) -> str: return text -def render(directory: Path, errors: list[str]) -> str: +def render( + directory: Path, + errors: list[str], + families: list[tuple[str, str]] | None = None, + *, + failures: list[str] | None = None, +) -> str: + """families: (family, why it did not run, or "") per integration row. By + default, one row per integration-.xml present. Suites with failed + cases are appended to failures as "suite platform".""" + if families is None: + families = [ + (path.stem[len("integration-"):], "") + for path in sorted(directory.glob("integration-*.xml")) + ] + columns = [*UNIT_COLUMNS] + not_run = {} + for family, why in families: + columns.append((f"integration-{family}.xml", "Integration", family)) + not_run[family] = why + rows: list[tuple[str, str, dict[str, object] | None]] = [] - for filename, suite, platform in COLUMNS: + for filename, suite, platform in columns: path = directory / filename if path.is_file(): rows.append((suite, platform, try_summarize(path, errors))) + elif not_run.get(platform): + rows.append((suite, platform, {"not_run": not_run[platform]})) else: rows.append((suite, platform, None)) extra = sorted( path.name for path in directory.glob("*.xml") - if path.name not in {filename for filename, _, _ in COLUMNS} + if path.name not in {filename for filename, _, _ in columns} ) for filename in extra: rows.append(("Other", filename, try_summarize(directory / filename, errors))) @@ -141,12 +154,17 @@ def render(directory: Path, errors: list[str]) -> str: if data is None: lines.append(f"| {suite} | {platform} | — | — | — | — | _no results_ |") continue + if "not_run" in data: + lines.append(f"| {suite} | {platform} | — | — | — | — | _not run: {data['not_run']}_ |") + continue if "error" in data: lines.append( f"| {suite} | {platform} | — | — | — | — | _unreadable XML: {data['error']}_ |" ) continue counts = data["counts"] + if failures is not None and counts["FAIL"]: + failures.append(f"{suite} {platform}") lines.append( "| {suite} | {platform} | {passed} | {failed} | {skipped} | {config} | {names} |".format( suite=suite, @@ -162,7 +180,7 @@ def render(directory: Path, errors: list[str]) -> str: lines.extend( [ "", - "Counts are top-level Catch2 cases. Missing XML means that job did not upload results (cancelled or not started). Unreadable XML is usually a job killed mid-run; its results are incomplete.", + "Counts are top-level Catch2 cases. Skip is a case that skipped itself, such as the cases that need two GPUs on a single-GPU runner; Config skip is a case a rule in the suite config skipped. No results means that job did not upload results (cancelled or not started). Unreadable XML is usually a job killed mid-run; its results are incomplete.", "", "### Counts grid", "", @@ -176,9 +194,7 @@ def render(directory: Path, errors: list[str]) -> str: ("Config skip", "CONFIG")): values = [] for _, _, data in rows: - values.append( - cell(None if data is None or "error" in data else data["counts"][key]) - ) + values.append(cell(data["counts"][key] if data and "counts" in data else None)) lines.append("| " + " | ".join([label, *values]) + " |") return "\n".join(lines) + "\n" @@ -197,14 +213,17 @@ def main(argv: list[str] | None = None) -> int: print(f"not a directory: {args.dir}", file=sys.stderr) return 1 errors: list[str] = [] - markdown = render(args.dir, errors) + failures: list[str] = [] + markdown = render(args.dir, errors, failures=failures) sys.stdout.write(markdown) if args.summary is not None: with args.summary.open("a", encoding="utf-8") as out: out.write(markdown) for error in errors: print(f"::error::unreadable JUnit XML {error}") - return 1 if errors else 0 + for suite in failures: + print(f"::error::failed cases in {suite}") + return 1 if errors or failures else 0 if __name__ == "__main__": diff --git a/projects/hrr/tests/scripts/test_hrr_capture_build.py b/projects/hrr/tests/scripts/test_hrr_capture_build.py new file mode 100644 index 000000000000..304e82be4c52 --- /dev/null +++ b/projects/hrr/tests/scripts/test_hrr_capture_build.py @@ -0,0 +1,290 @@ +#!/usr/bin/env python3 +# Copyright (c) Advanced Micro Devices, Inc., or its affiliates. +# SPDX-License-Identifier: MIT +"""Tests for hrr_capture_build.py.""" + +import argparse +import json +import os +import sys +import tempfile +import unittest +from pathlib import Path + +sys.path.insert(0, str(Path(__file__).resolve().parent)) + +import hrr_capture_build as cb # noqa: E402 + +FINDER_FOUND = { + "status": "found", + "results": [ + { + "github_repository_name": "ROCm/rocm-systems", + "workflow_run_id": "123", + "git_commit_sha": "abc", + "workflow_run_html_url": "https://example/runs/123", + "artifact_group": "gfx94X-dcgpu", + } + ], +} + + +class PickBaseRunTest(unittest.TestCase): + def test_found(self): + base = cb.pick_base_run(FINDER_FOUND) + self.assertEqual(base["run_id"], "123") + self.assertEqual(base["commit"], "abc") + self.assertEqual(base["repo"], "ROCm/rocm-systems") + + def test_not_found(self): + with self.assertRaises(SystemExit): + cb.pick_base_run({"status": "not_found", "results": []}) + + +class CommandTest(unittest.TestCase): + def test_fetch_command_uses_recorded_run(self): + base = cb.pick_base_run(FINDER_FOUND) + cmd = [ + str(c) + for c in cb.fetch_command(Path("/t"), base, Path("/d"), ("core-hip_lib",)) + ] + self.assertIn("--flatten", cmd) + self.assertEqual(cmd[cmd.index("--run-id") + 1], "123") + self.assertEqual(cmd[cmd.index("--run-github-repo") + 1], "ROCm/rocm-systems") + self.assertEqual(cmd[-1], "core-hip_lib") + + def test_clr_compiles_capture_in(self): + cmd = [ + str(c) for c in cb.clr_configure_command(Path("/s"), Path("/b"), Path("/d")) + ] + self.assertIn("-DCMAKE_CXX_FLAGS=-DHIP_HRR_CAPTURE_ENABLED", cmd) + self.assertIn("-DCMAKE_INSTALL_PREFIX=/d", cmd) + self.assertIn("-DHIP_COMMON_DIR=/s/projects/hip", cmd) + + def test_hrr_builds_integration_suite_for_archs(self): + cmd = [ + str(c) + for c in cb.hrr_configure_command( + Path("/s"), Path("/b"), Path("/d"), "gfx942;gfx950" + ) + ] + self.assertIn("-DGPU_TARGETS=gfx942;gfx950", cmd) + self.assertIn("-DHRR_BUILD_INTEGRATION_TESTS=ON", cmd) + self.assertIn("-DCMAKE_HIP_COMPILER=/d/lib/llvm/bin/clang++", cmd) + self.assertIn(f"-DCMAKE_BUILD_RPATH={cb.BUILD_RPATH}", cmd) + + def test_build_targets_cover_uploaded_executables(self): + names = {Path(rel).name for rel in cb.EXECUTABLES} + self.assertTrue(names <= set(cb.HRR_TARGETS)) + + def test_runtime_artifacts_are_a_subset_of_build_artifacts(self): + self.assertTrue(set(cb.RUNTIME_ARTIFACTS) <= set(cb.BUILD_ARTIFACTS)) + + +class EmbeddedTargetsTest(unittest.TestCase): + OUTPUT = ( + "OFFLOADING IMAGE [0]:\nkind elf\narch gfx942\ntriple amdgcn-amd-amdhsa\n" + "OFFLOADING IMAGE [1]:\nkind elf\narch gfx1100\n" + ) + + def test_parses_arches(self): + self.assertEqual(cb.embedded_targets(self.OUTPUT), {"gfx942", "gfx1100"}) + + def test_missing(self): + self.assertEqual( + cb.missing_targets("gfx942;gfx950;gfx1100", self.OUTPUT), ["gfx950"] + ) + + def test_none_missing(self): + self.assertEqual(cb.missing_targets("gfx942;gfx1100", self.OUTPUT), []) + + +class LibraryOverlayTest(unittest.TestCase): + def setUp(self): + self._tmp = tempfile.TemporaryDirectory() + self.tmp = Path(self._tmp.name) + + def tearDown(self): + self._tmp.cleanup() + + def make_lib(self, directory: Path, real: str, content: str) -> None: + directory.mkdir(parents=True, exist_ok=True) + (directory / real).write_text(content) + os.symlink(real, directory / "libamdhip64.so.7") + os.symlink("libamdhip64.so.7", directory / "libamdhip64.so") + + def test_overlay_replaces_release_library_and_keeps_rocr(self): + capture = self.tmp / "capture" + dist = self.tmp / "dist" / "lib" + self.make_lib(capture, "libamdhip64.so.7.2.1", "capture") + self.make_lib(dist, "libamdhip64.so.7.2.0", "release") + (dist / "libhiprtc.so.7").write_text("release") + (capture / "libhiprtc.so.7").write_text("capture") + (dist / "libhsa-runtime64.so.1").write_text("base") + + cb.remove_amdhip64(dist) + copied = cb.copy_amdhip64(capture, dist) + + self.assertEqual( + sorted(copied), + [ + "libamdhip64.so", + "libamdhip64.so.7", + "libamdhip64.so.7.2.1", + "libhiprtc.so.7", + ], + ) + self.assertFalse((dist / "libamdhip64.so.7.2.0").exists()) + self.assertEqual((dist / "libamdhip64.so").read_text(), "capture") + self.assertTrue((dist / "libamdhip64.so.7").is_symlink()) + self.assertEqual((dist / "libhiprtc.so.7").read_text(), "capture") + self.assertEqual((dist / "libhsa-runtime64.so.1").read_text(), "base") + + def test_copy_without_library_fails(self): + (self.tmp / "empty").mkdir() + with self.assertRaises(SystemExit): + cb.copy_amdhip64(self.tmp / "empty", self.tmp / "out") + + +class LinkageTest(unittest.TestCase): + LIB = Path("/w/dist/lib") + + def ldd(self, hsa: str, extra: str = "") -> str: + return ( + "\tlinux-vdso.so.1 (0x00007ffd)\n" + "\tlibamdhip64.so.7 => /w/dist/lib/libamdhip64.so.7 (0x00007f1)\n" + f"\tlibhsa-runtime64.so.1 => {hsa} (0x00007f2)\n" + f"{extra}" + "\tlibc.so.6 => /lib64/libc.so.6 (0x00007f3)\n" + ) + + def test_both_from_dist(self): + self.assertEqual( + cb.linkage_errors(self.ldd("/w/dist/lib/libhsa-runtime64.so.1"), self.LIB), + [], + ) + + def test_relative_runpath_entry(self): + text = ( + "\tlibamdhip64.so.7 => /w/build/hrr/tests/integration/../../../../dist/lib/libamdhip64.so.7 (0x1)\n" + "\tlibhsa-runtime64.so.1 => /w/build/hrr/tests/integration/../../../../dist/lib/libhsa-runtime64.so.1 (0x2)\n" + ) + self.assertEqual(cb.linkage_errors(text, self.LIB), []) + + def test_not_found_is_an_error(self): + errors = cb.linkage_errors(self.ldd("not found"), self.LIB) + self.assertEqual(len(errors), 1) + + def test_rocr_from_elsewhere(self): + errors = cb.linkage_errors( + self.ldd("/opt/rocm/lib/libhsa-runtime64.so.1"), self.LIB + ) + self.assertEqual(len(errors), 1) + self.assertIn("libhsa-runtime64.so", errors[0]) + + def test_two_rocr_copies(self): + extra = "\tlibhsa-runtime64.so.2 => /opt/rocm/lib/libhsa-runtime64.so.2 (0x0)\n" + errors = cb.linkage_errors( + self.ldd("/w/dist/lib/libhsa-runtime64.so.1", extra), self.LIB + ) + self.assertEqual(len(errors), 1) + + def test_missing_rocr(self): + text = "\tlibamdhip64.so.7 => /w/dist/lib/libamdhip64.so.7 (0x1)\n" + self.assertEqual(len(cb.linkage_errors(text, self.LIB)), 1) + + +class BuildTest(unittest.TestCase): + def test_builds_clr_and_hrr_over_the_prebuilt_rocr(self): + with tempfile.TemporaryDirectory() as tmp: + root = Path(tmp) + dist = root / "dist" + (dist / "lib").mkdir(parents=True) + (dist / "lib" / "libhsa-runtime64.so.1").write_text("base") + calls = [] + + def runner(cmd): + cmd = [str(c) for c in cmd] + calls.append(cmd) + if cmd[:2] == ["cmake", "--install"]: + (dist / "lib" / "libamdhip64.so.7").write_text("capture") + + args = argparse.Namespace( + src=root / "src", + build=root / "build", + dist=dist, + capture_lib=root / "build" / "capture-lib", + archs="gfx942", + ) + cb.cmd_build(args, runner=runner) + sources = [c[c.index("-S") + 1] for c in calls if "-S" in c] + self.assertEqual( + sources, [f"{root}/src/projects/clr", f"{root}/src/projects/hrr"] + ) + self.assertEqual( + [p.name for p in args.capture_lib.iterdir()], ["libamdhip64.so.7"] + ) + self.assertEqual( + (dist / "lib" / "libhsa-runtime64.so.1").read_text(), "base" + ) + + +class StageRuntimeLibsTest(unittest.TestCase): + def test_kpack_outside_the_install_is_staged(self): + text = "\n".join( + [ + "libamdhip64.so.7 => /work/dist/lib/libamdhip64.so.7 (0x1)", + "libhsa-runtime64.so.1 => /work/dist/lib/libhsa-runtime64.so.1 (0x2)", + "librocm_kpack.so.0 => /work/build/clr/lib/librocm_kpack.so.0 (0x3)", + "libc.so.6 => /lib64/libc.so.6 (0x4)", + ] + ) + sources = cb.staged_library_sources(text, Path("/work/dist/lib")) + self.assertEqual(sources, [Path("/work/build/clr/lib/librocm_kpack.so.0")]) + + +class ArchsTest(unittest.TestCase): + SRC = Path(__file__).resolve().parents[4] + + def test_default_is_the_integration_config(self): + config = cb.hrr_test_config.load(self.SRC / cb.INTEGRATION_CONFIG) + args = cb.parse_args(["build", "--src", str(self.SRC)]) + self.assertEqual(args.archs, ";".join(config["targets"]["arch"])) + + def test_override(self): + args = cb.parse_args( + ["verify", "--src", str(self.SRC), "--archs", "gfx942;gfx1152"] + ) + self.assertEqual(args.archs, "gfx942;gfx1152") + + +class ExecutableBitsTest(unittest.TestCase): + def test_restores_bits(self): + with tempfile.TemporaryDirectory() as tmp: + build = Path(tmp) + for rel in cb.EXECUTABLES: + (build / rel).parent.mkdir(parents=True, exist_ok=True) + (build / rel).write_text("") + (build / rel).chmod(0o644) + cb.restore_executable_bits(build) + for rel in cb.EXECUTABLES: + self.assertEqual((build / rel).stat().st_mode & 0o777, 0o755) + + +class FetchBaseTest(unittest.TestCase): + def test_reuses_recorded_base(self): + with tempfile.TemporaryDirectory() as tmp: + base_json = Path(tmp) / "base.json" + base_json.write_text(json.dumps(cb.pick_base_run(FINDER_FOUND))) + calls = [] + args = argparse.Namespace( + base_json=base_json, therock=Path("/t"), dist=Path("/d") + ) + cb.cmd_fetch_base(args, runner=calls.append) + cmd = [str(c) for c in calls[0]] + self.assertEqual(cmd[cmd.index("--run-id") + 1], "123") + self.assertIn("core-runtime_dev", cmd) + + +if __name__ == "__main__": + unittest.main() diff --git a/projects/hrr/tests/scripts/test_hrr_check_sources.py b/projects/hrr/tests/scripts/test_hrr_check_sources.py new file mode 100644 index 000000000000..9f1164b658cd --- /dev/null +++ b/projects/hrr/tests/scripts/test_hrr_check_sources.py @@ -0,0 +1,161 @@ +#!/usr/bin/env python3 +# Copyright (c) Advanced Micro Devices, Inc., or its affiliates. +# SPDX-License-Identifier: MIT +"""Tests for hrr_check_sources.py.""" + +import contextlib +import io +import subprocess +import sys +import tempfile +import unittest +from pathlib import Path + +sys.path.insert(0, str(Path(__file__).resolve().parents[2] / "tools")) + +import hrr_check_sources as chk # noqa: E402 + + +def write(path: Path, text: str) -> None: + path.parent.mkdir(parents=True, exist_ok=True) + path.write_text(text, encoding="utf-8") + + +class Tree(unittest.TestCase): + """A fake checkout whose runner records the commands it is given.""" + + def setUp(self): + self._tmp = tempfile.TemporaryDirectory() + base = Path(self._tmp.name) + self.root = base / "src" + self.out = base / "out" + self.out.mkdir() + write(self.root / chk.API_ARGS_HEADER, "// current header\n") + self.calls = [] + self.rc = {} + + def tearDown(self): + self._tmp.cleanup() + + def runner(self, cmd): + self.calls.append(list(cmd)) + return self.rc.get(Path(cmd[1]).name, 0) + + def quiet(self, fn, *args, **kwargs): + buf = io.StringIO() + with contextlib.redirect_stdout(buf): + result = fn(*args, **kwargs) + return result, buf.getvalue() + + +class ApiCompatTest(Tree): + def test_compares_against_base(self): + ok, _ = self.quiet( + chk.check_api_compat, self.root, self.out, "abc", + self.runner, lambda root, sha: ("base header\n", None), + ) + self.assertTrue(ok) + cmd = self.calls[0] + self.assertEqual(Path(cmd[1]).name, "check_hrr_api_compat.py") + current = Path(cmd[cmd.index("--current-header") + 1]) + self.assertEqual(current, self.root / chk.API_ARGS_HEADER) + base = Path(cmd[cmd.index("--base-header") + 1]) + self.assertEqual(base.read_text(), "base header\n") + + def test_incompatible_ids_fail(self): + self.rc["check_hrr_api_compat.py"] = 1 + ok, output = self.quiet( + chk.check_api_compat, self.root, self.out, "abc", + self.runner, lambda root, sha: ("base\n", None), + ) + self.assertFalse(ok) + self.assertIn("not compatible with abc", output) + + def test_missing_base_header_passes(self): + ok, output = self.quiet( + chk.check_api_compat, self.root, self.out, "abc", + self.runner, lambda root, sha: (None, None), + ) + self.assertTrue(ok) + self.assertEqual(self.calls, []) + self.assertIn("does not exist at abc", output) + + def test_git_failure_fails_instead_of_passing(self): + # A base blob the blob:none clone cannot fetch, or a base commit it + # does not have: not "the file is absent", so no silent pass. + failure = (None, "cannot read projects/hrr/include/hrr/hrr_api_args.h " + "at abc: fatal: could not read from remote repository") + ok, output = self.quiet( + chk.check_api_compat, self.root, self.out, "abc", + self.runner, lambda root, sha: failure, + ) + self.assertFalse(ok) + self.assertEqual(self.calls, []) + self.assertIn("::error::cannot read", output) + self.assertIn("do not pass it silently", output) + + +class BaseHeaderTest(unittest.TestCase): + """base_header's telling "the path is absent" from "git broke".""" + + def setUp(self): + self._tmp = tempfile.TemporaryDirectory() + self.root = Path(self._tmp.name) + self.addCleanup(self._tmp.cleanup) + for argv in (["init", "-q"], ["config", "user.email", "t@example.com"], + ["config", "user.name", "t"]): + subprocess.run(["git", "-C", str(self.root), *argv], check=True, + capture_output=True, text=True) + + def commit(self, relpath: str, text: str) -> str: + path = self.root / relpath + path.parent.mkdir(parents=True, exist_ok=True) + path.write_text(text, encoding="utf-8") + subprocess.run(["git", "-C", str(self.root), "add", relpath], check=True) + subprocess.run(["git", "-C", str(self.root), "commit", "-qm", relpath], check=True) + return subprocess.run( + ["git", "-C", str(self.root), "rev-parse", "HEAD"], + check=True, capture_output=True, text=True, + ).stdout.strip() + + def test_content_at_base(self): + sha = self.commit(chk.API_ARGS_HEADER, "// base\n") + text, failure = chk.base_header(self.root, sha) + self.assertEqual(text, "// base\n") + self.assertIsNone(failure) + + def test_path_genuinely_absent_passes_as_none(self): + # Committed, but never this path: the only "nothing to compare" case. + sha = self.commit("README.md", "x\n") + text, failure = chk.base_header(self.root, sha) + self.assertIsNone(text) + self.assertIsNone(failure) + + def test_base_commit_the_clone_does_not_have_fails(self): + text, failure = chk.base_header(self.root, "0" * 40) + self.assertIsNone(text) + self.assertIn("cannot read", failure) + self.assertIn("at " + "0" * 40, failure) + + +class WorkloadsTest(Tree): + def test_runs_the_workload_check(self): + ok, _ = self.quiet(chk.check_workloads, self.root, self.runner) + self.assertTrue(ok) + self.assertEqual(Path(self.calls[0][1]).name, "check_matrix.py") + self.assertIn("--check-workloads", self.calls[0]) + + def test_failure_is_reported(self): + self.rc["check_matrix.py"] = 2 + ok, output = self.quiet(chk.check_workloads, self.root, self.runner) + self.assertFalse(ok) + self.assertIn("--check-workloads failed (exit 2)", output) + + +class ResolveBaseTest(unittest.TestCase): + def test_explicit_sha(self): + self.assertEqual(chk.resolve_base(Path("."), "1234abcd"), "1234abcd") + + +if __name__ == "__main__": + unittest.main() diff --git a/projects/hrr/tests/scripts/test_hrr_ci_configure.py b/projects/hrr/tests/scripts/test_hrr_ci_configure.py new file mode 100644 index 000000000000..4efd60fd464e --- /dev/null +++ b/projects/hrr/tests/scripts/test_hrr_ci_configure.py @@ -0,0 +1,729 @@ +#!/usr/bin/env python3 +# Copyright (c) Advanced Micro Devices, Inc., or its affiliates. +# SPDX-License-Identifier: MIT +"""Tests for hrr_ci_configure.py.""" + +import copy +import io +import os +import sys +import re +import tempfile +import unittest +from contextlib import redirect_stdout +from pathlib import Path +from unittest import mock + +import yaml + +sys.path.insert(0, str(Path(__file__).resolve().parent)) + +import hrr_ci_configure as cfg # noqa: E402 + +MATRIX = { + "gfx94x": { + "linux": { + "family": "gfx94X-dcgpu", + "test-runs-on": "gpu-94x", + "test-runs-on-labels": [{"label": "gpu-94x-a", "count": 1}], + "fetch-gfx-targets": ["gfx942"], + } + }, + "gfx90a": { + "linux": { + "family": "gfx90a", + "test-runs-on": "gpu-90a", + "fetch-gfx-targets": ["gfx90a"], + } + }, + "gfx103x": { + "linux": { + "family": "gfx103X-all", + "test-runs-on": "gpu-103x", + "fetch-gfx-targets": ["gfx1030"], + } + }, + "gfx110x": { + "linux": { + "family": "gfx110X-all", + "test-runs-on": "gpu-110x", + "fetch-gfx-targets": ["gfx1100", "gfx1101"], + } + }, + "gfx950": { + "linux": { + "family": "gfx950-dcgpu", + "test-runs-on": "gpu-950-1", + "test-runs-on-multi-gpu": "gpu-950-8", + "fetch-gfx-targets": ["gfx950"], + } + }, + "gfx1150": { + "linux": { + "family": "gfx1150", + "test-runs-on": "", + } + }, + "gfx125x": { + "linux": { + "family": "gfx125X-dcgpu", + "test-runs-on": "gpu-125x", + "fetch-gfx-targets": ["gfx1250"], + } + }, +} + +REPO_CONFIG = yaml.safe_load(cfg.DEFAULT_CONFIG.read_text(encoding="utf-8")) + +# The repository's paths with families that exercise every option. gfx103x is +# in MATRIX but not here. +FAMILIES = { + "gfx90a": {}, + "gfx94x": None, + "gfx950": {}, + "gfx110x": {"enabled": False}, + "gfx1150": {}, + "gfx125x": {"enabled": False}, +} + + +def config_data(**changes): + data = copy.deepcopy(REPO_CONFIG) + data["families"] = copy.deepcopy(FAMILIES) + data.update(changes) + families = data.get("families") + simulator = data.get("simulator") + if isinstance(families, dict) and isinstance(simulator, dict): + present = [ + key for key in simulator.get("default_families", []) if key in families + ] + if not present and families: + present = [next(iter(families))] + simulator["default_families"] = present + configs = simulator.setdefault("configs", {}) + if isinstance(configs, dict): + for key in [key for key in configs if key not in families]: + del configs[key] + for key in present: + configs.setdefault(key, f"emulation/rocjitsu/configs/{key}.json") + return data + + +def gpu_data(**changes): + """The family-selection tests predate the simulator default.""" + data = config_data(**changes) + data["backend"] = {"default": "gpu"} + return data + + +CONFIG = cfg.parse_config(gpu_data()) +REPO = cfg.load_config(cfg.DEFAULT_CONFIG) + + +def pick_first(labels, _context): + return labels[0]["label"] + + +def decide(event, paths, labels=(), is_fork=False, requested=(), config=CONFIG): + return cfg.decide( + event, paths, list(labels), is_fork, list(requested), config, MATRIX, pick_first + ) + + +def family_keys(decision): + return [f["family"] for f in decision.families] + + +class RepoConfigTest(unittest.TestCase): + def test_a_gfx_label_does_not_add_a_simulator_family(self): + """ci:gfx is a TheRock label. It must not widen the rocJITsu matrix.""" + data = copy.deepcopy(REPO_CONFIG) + data["simulator"]["default_families"] = ["gfx950", "gfx125x"] + d = decide( + "pull_request", + ["projects/hrr/playback/x.cpp"], + ["ci:gfx110x"], + config=cfg.parse_config(data), + ) + self.assertEqual(family_keys(d), ["gfx950", "gfx125x"]) + self.assertTrue(all(row["backend"] == "simulator" for row in d.families)) + +class PathClassificationTest(unittest.TestCase): + def test_playback_change_is_capture(self): + self.assertTrue( + cfg.touches_capture(["projects/hrr/playback/hip_playback.cpp"], REPO) + ) + + def test_capture_source_is_capture(self): + self.assertTrue( + cfg.touches_capture(["projects/clr/hipamd/src/hrr/hip_capture.cpp"], REPO) + ) + + def test_ci_config_runs_the_suite_and_is_not_capture(self): + paths = ["projects/hrr/tests/config/ci.yaml"] + self.assertFalse(cfg.touches_capture(paths, REPO)) + self.assertTrue(cfg.touches_suite(paths, REPO)) + decision = decide("pull_request", paths, config=REPO) + self.assertTrue(decision.run_gpu) + self.assertEqual(family_keys(decision), list(REPO.simulator.default_families)) + + def test_public_header_is_cpu_only(self): + paths = ["projects/hip/include/hip/hip_runtime_api.h"] + self.assertTrue(cfg.touches_cpu(paths, REPO)) + self.assertFalse(cfg.touches_capture(paths, REPO)) + + def test_unit_test_is_cpu_only(self): + paths = ["projects/hrr/tests/unit/hrr_format_test.cc"] + self.assertTrue(cfg.touches_cpu(paths, REPO)) + self.assertFalse(cfg.touches_capture(paths, REPO)) + + def test_skills_are_excluded(self): + paths = ["projects/hrr/skills/hrr/SKILL.md", "projects/hrr/README.md"] + self.assertFalse(cfg.touches_cpu(paths, REPO)) + + def test_unrelated_change(self): + self.assertFalse(cfg.touches_cpu(["projects/rocminfo/rocminfo.cc"], REPO)) + + def test_prefix_must_match_whole_component(self): + self.assertFalse(cfg.touches_cpu(["projects/hrrx/foo.cc"], REPO)) + + +class ConfigValidationTest(unittest.TestCase): + def assertRejected(self, data, message): + with self.assertRaises(cfg.ConfigError) as caught: + cfg.parse_config(data) + self.assertIn(message, str(caught.exception)) + + def test_unknown_top_level_key(self): + self.assertRejected(config_data(gpu_families=[]), "unknown key gpu_families") + + def test_missing_key(self): + data = config_data() + del data["capture_label"] + self.assertRejected(data, "missing capture_label") + + def test_wrong_version(self): + self.assertRejected(config_data(version=2), "version: expected 1") + + def test_unknown_family_option(self): + families = {"gfx94x": {"runs_on": "x"}} + self.assertRejected( + config_data(families=families), "families.gfx94x: unknown key runs_on" + ) + + def test_enabled_must_be_a_bool(self): + families = {"gfx94x": {"enabled": "no"}} + self.assertRejected(config_data(families=families), "families.gfx94x.enabled") + + def test_empty_path_list(self): + data = config_data() + data["paths"]["capture"] = [] + self.assertRejected(data, "paths.capture: expected a non-empty list") + + def test_no_families(self): + self.assertRejected(config_data(families={}), "families: expected a mapping") + + def test_family_missing_from_matrix(self): + config = cfg.parse_config(config_data(families={"gfx94x": {}, "gfx999": {}})) + with self.assertRaises(cfg.ConfigError) as caught: + cfg.check_families(config, MATRIX) + self.assertIn("gfx999 not in TheRock's family matrix", str(caught.exception)) + + def test_family_targets(self): + self.assertEqual( + cfg.family_targets(MATRIX["gfx110x"]["linux"]), ["gfx1100", "gfx1101"] + ) + # gfx1150 lists no fetch targets; its family name is the target. + self.assertEqual(cfg.family_targets(MATRIX["gfx1150"]["linux"]), ["gfx1150"]) + self.assertEqual( + cfg.family_targets({"fetch-gfx-targets": ["gfx942", "gfx942:xnack+"]}), + ["gfx942"], + ) + self.assertEqual(cfg.family_targets({"family": "gfx101X-dgpu"}), []) + + def test_targets_cover_every_family(self): + archs = [ + "gfx90a", + "gfx942", + "gfx950", + "gfx1100", + "gfx1101", + "gfx1150", + "gfx1250", + ] + cfg.check_targets(CONFIG, MATRIX, archs) + with self.assertRaises(cfg.ConfigError) as caught: + cfg.check_targets( + CONFIG, MATRIX, [a for a in archs if a not in ("gfx1101", "gfx1250")] + ) + # gfx125x is disabled, but a label can still request it. + self.assertIn( + "gfx110x needs gfx1101; gfx125x needs gfx1250", str(caught.exception) + ) + + def test_family_without_targets_is_an_error(self): + config = cfg.parse_config(config_data(families={"gfx94x": {}})) + matrix = {"gfx94x": {"linux": {"family": "gfx94X-dcgpu"}}} + with self.assertRaises(cfg.ConfigError) as caught: + cfg.check_targets(config, matrix, ["gfx942"]) + self.assertIn("no GPU targets for gfx94x", str(caught.exception)) + + def test_repo_integration_config_lists_archs(self): + self.assertIn("gfx1103", cfg.load_archs(cfg.INTEGRATION_CONFIG)) + + def test_load_error_names_the_file(self): + with tempfile.TemporaryDirectory() as tmp: + path = Path(tmp) / "ci.yaml" + path.write_text("version: 1\n") + with self.assertRaises(cfg.ConfigError) as caught: + cfg.load_config(path) + self.assertIn(str(path), str(caught.exception)) + + def test_main_reports_a_bad_config(self): + with tempfile.TemporaryDirectory() as tmp: + path = Path(tmp) / "ci.yaml" + path.write_text(yaml.safe_dump(config_data(version=2))) + out = io.StringIO() + with redirect_stdout(out): + status = cfg.main( + [ + "--event", + "schedule", + "--repo", + "r", + "--therock-dir", + tmp, + "--config", + str(path), + ] + ) + self.assertEqual(status, 1) + self.assertIn("::error::", out.getvalue()) + + +class DecideTest(unittest.TestCase): + def test_unrelated_pull_request_runs_nothing(self): + d = decide("pull_request", ["projects/rocminfo/rocminfo.cc"]) + self.assertFalse(d.run_cpu) + self.assertFalse(d.run_gpu) + + def test_cpu_only_pull_request(self): + d = decide("pull_request", ["projects/hrr/tests/unit/hrr_format_test.cc"]) + self.assertTrue(d.run_cpu) + self.assertFalse(d.run_gpu) + + def test_capture_pull_request_runs_enabled_families(self): + d = decide("pull_request", ["projects/clr/hipamd/src/hrr/hip_capture.cpp"]) + self.assertTrue(d.run_gpu) + # gfx110x and gfx125x are disabled and gfx1150 has no runner. + self.assertEqual(family_keys(d), ["gfx90a", "gfx94x", "gfx950"]) + self.assertEqual( + d.families[1], + { + "family": "gfx94x", + "runner": "gpu-94x-a", + "backend": "gpu", + "arch": "gfx942", + "config": "", + }, + ) + self.assertIn("no runner for gfx1150", d.reason) + self.assertEqual( + d.skipped, + [ + {"family": "gfx110x", "reason": "disabled in ci.yaml"}, + {"family": "gfx1150", "reason": cfg.NO_RUNNER}, + {"family": "gfx125x", "reason": "disabled in ci.yaml"}, + ], + ) + + def test_rename_out_of_capture_counts(self): + # The files API lists previous_filename alongside filename. + d = decide( + "pull_request", ["projects/other/x.cpp", "projects/hrr/playback/x.cpp"] + ) + self.assertTrue(d.run_gpu) + + def test_label_requests_gpu_on_unrelated_pull_request(self): + d = decide( + "pull_request", ["projects/rocminfo/x.cc"], labels=["ci:hrr-capture"] + ) + self.assertTrue(d.run_cpu) + self.assertTrue(d.run_gpu) + + def test_label_match_ignores_case(self): + d = decide( + "pull_request", ["projects/rocminfo/x.cc"], labels=["CI:HRR-Capture"] + ) + self.assertTrue(d.run_gpu) + + def test_capture_only_path_also_runs_cpu(self): + d = decide("pull_request", ["projects/clr/hipamd/src/hip_context.cpp"]) + self.assertTrue(d.run_cpu) + self.assertTrue(d.run_gpu) + + def test_gfx_label_does_not_add_a_family(self): + d = decide( + "pull_request", + ["projects/hrr/playback/x.cpp"], + labels=["ci:gfx110X-all", "ci:gfx125x"], + ) + self.assertEqual(family_keys(d), ["gfx90a", "gfx94x", "gfx950"]) + + def test_label_for_family_not_in_config_is_ignored(self): + d = decide( + "pull_request", ["projects/hrr/playback/x.cpp"], labels=["ci:gfx103x"] + ) + self.assertNotIn("gfx103x", family_keys(d)) + + def test_fork_still_runs_the_simulator(self): + d = decide( + "pull_request", + ["projects/hrr/playback/x.cpp"], + is_fork=True, + config=cfg.parse_config(config_data()), + ) + self.assertTrue(d.run_cpu) + self.assertTrue(d.run_gpu) + self.assertTrue(d.families) + self.assertTrue(all(row["backend"] == "simulator" for row in d.families)) + + def test_fork_skips_the_real_gpu_backend(self): + d = decide("pull_request", ["projects/hrr/playback/x.cpp"], is_fork=True) + self.assertTrue(d.run_cpu) + self.assertFalse(d.run_gpu) + self.assertIn("fork", d.reason) + + def test_push_without_capture_change_has_no_gpu(self): + d = decide("push", ["projects/hrr/tests/unit/x.cc"]) + self.assertTrue(d.run_cpu) + self.assertFalse(d.run_gpu) + + def test_unlisted_change_set_runs_everything(self): + d = decide("push", None) + self.assertTrue(d.run_cpu) + self.assertTrue(d.run_gpu) + + def test_schedule_runs_the_enabled_families(self): + d = decide("schedule", []) + self.assertEqual(family_keys(d), ["gfx90a", "gfx94x", "gfx950"]) + + def test_dispatch_families_replace_the_selection(self): + d = decide("workflow_dispatch", [], requested=["gfx125x", "gfx1150"]) + self.assertEqual(family_keys(d), ["gfx125x"]) + self.assertIn("no runner for gfx1150", d.reason) + self.assertIn({"family": "gfx94x", "reason": "not requested"}, d.skipped) + + def test_dispatch_unknown_family(self): + with self.assertRaises(cfg.ConfigError) as caught: + decide("workflow_dispatch", [], requested=["gfx103x"]) + self.assertIn("gfx103x", str(caught.exception)) + + def test_unknown_event(self): + d = decide("merge_group", []) + self.assertFalse(d.run_cpu) + self.assertFalse(d.run_gpu) + + +class SimulatorBackendTest(unittest.TestCase): + def test_default_runs_no_real_gpu_job_on_any_event(self): + """Pull requests, pushes and the nightly run only rocJitsu jobs; none + queues on a GPU runner.""" + for event in ("pull_request", "push", "schedule", "workflow_dispatch"): + decision = decide(event, ["projects/hrr/playback/x.cpp"], config=REPO) + self.assertTrue(decision.run_gpu, event) + self.assertEqual( + {row["backend"] for row in decision.families}, {"simulator"}, event + ) + + def test_every_default_simulator_family_has_a_config_file(self): + root = Path(__file__).resolve().parents[4] + for key in REPO.simulator.default_families: + self.assertTrue((root / REPO.simulator.configs[key]).is_file(), key) + + def test_the_label_does_not_move_a_run_to_a_real_gpu(self): + for label, path in ( + ("ci:hrr-capture", "projects/hrr/README.md"), + ("ci:gfx94x", "projects/hrr/playback/x.cpp"), + ): + decision = decide("pull_request", [path], [label], config=REPO) + self.assertTrue(decision.run_gpu, label) + self.assertEqual( + {row["backend"] for row in decision.families}, {"simulator"}, label + ) + + def test_default_event_is_the_rocjitsu_jobs(self): + decision = decide("pull_request", ["projects/hrr/playback/x.cpp"], config=REPO) + self.assertIn("gfx950", family_keys(decision)) + self.assertIn("gfx125x", family_keys(decision)) + by_family = {row["family"]: row for row in decision.families} + self.assertEqual(by_family["gfx950"]["runner"], "ubuntu-24.04") + self.assertEqual(by_family["gfx950"]["backend"], "simulator") + self.assertEqual(by_family["gfx950"]["arch"], "gfx950") + self.assertTrue( + by_family["gfx950"]["config"].endswith("gfx950_mi355x_kmd_2gpu.json") + ) + self.assertEqual(by_family["gfx125x"]["runner"], "ubuntu-24.04") + self.assertEqual(by_family["gfx125x"]["arch"], "gfx1250") + self.assertTrue( + by_family["gfx125x"]["config"].endswith("gfx1250_mi455x_kmd_4gpu.json") + ) + self.assertNotEqual( + by_family["gfx125x"]["runner"], MATRIX["gfx125x"]["linux"]["test-runs-on"] + ) + + def test_gpu_dispatch_selects_the_therock_runner(self): + decision = cfg.decide( + "workflow_dispatch", + [], + [], + False, + ["gfx94x"], + REPO, + MATRIX, + pick_first, + backend="gpu", + ) + self.assertEqual(family_keys(decision), ["gfx94x"]) + self.assertEqual(decision.families[0]["backend"], "gpu") + self.assertEqual(decision.families[0]["runner"], "gpu-94x-a") + self.assertNotIn("gfx125x", family_keys(decision)) + + def test_simulator_rejects_a_requested_family_without_a_config(self): + with self.assertRaises(cfg.ConfigError) as caught: + decide("workflow_dispatch", [], requested=["gfx103x"], config=REPO) + self.assertIn("no config for gfx103x", str(caught.exception)) + + def test_unknown_backend_value(self): + data = config_data() + data["backend"]["default"] = "metal" + with self.assertRaises(cfg.ConfigError) as caught: + cfg.parse_config(data) + self.assertIn("backend.default", str(caught.exception)) + + +class LabelEventTest(unittest.TestCase): + CAPTURE = ["projects/hrr/playback/x.cpp"] + WORKFLOW = ( + Path(__file__).resolve().parents[4] / ".github" / "workflows" / "hrr-ci.yml" + ) + + def test_which_labels_matter(self): + for label in ("ci:hrr-capture", "CI:HRR-Capture"): + self.assertTrue(cfg.label_matters(label, REPO), label) + for label in ( + "documentation", + "ci:skip", + "hrr", + "ci:gfx110x", + "ci:gfx110X-all", + ): + self.assertFalse(cfg.label_matters(label, REPO), label) + + def test_other_label_runs_nothing(self): + d = cfg.decide( + "pull_request", + self.CAPTURE, + ["ci:hrr-capture", "documentation"], + False, + [], + CONFIG, + MATRIX, + pick_first, + added_label="documentation", + ) + self.assertFalse(d.run_cpu) + self.assertFalse(d.run_gpu) + self.assertIn("documentation", d.reason) + + def test_capture_label_reruns_and_a_gfx_label_does_not(self): + d = cfg.decide( + "pull_request", + self.CAPTURE, + ["ci:hrr-capture"], + False, + [], + CONFIG, + MATRIX, + pick_first, + added_label="ci:hrr-capture", + ) + self.assertTrue(d.run_gpu) + d = cfg.decide( + "pull_request", + self.CAPTURE, + ["ci:gfx110x"], + False, + [], + CONFIG, + MATRIX, + pick_first, + added_label="ci:gfx110x", + ) + self.assertFalse(d.run_cpu) + self.assertFalse(d.run_gpu) + + def test_labeled_trigger_and_summary_report_a_verdict(self): + """Regression for the label that can turn a failing HRR CI green. + + github keeps the most recent check run of a name on a commit, and a + skipped check satisfies a required-status rule, so a labeled run + that configures nothing must not let its summary skip on top of the + run that failed: the summary must run for every labeled event and + answer with the verdict of the run that tested the commit (the + label run knows nothing itself), never a blanket pass. + """ + text = self.WORKFLOW.read_text(encoding="utf-8") + types = re.search(r"types:\s*\[([^\]]*)\]", text).group(1) + self.assertIn("labeled", types) + + summary = text[text.index(" summary:") :] + # The summary also runs when the label run configured nothing. + self.assertIn("github.event.action == 'labeled'", summary) + # And it answers with an actual verdict, never a skip of its own. + self.assertNotIn( + "needs.configure.result != 'success' || needs.configure.outputs.run_cpu == 'true' ||", + summary, + ) + # Only a label run that configured nothing passes the verdict through; + # every other run reports its own jobs. + step = summary[summary.index("PASS_THROUGH:") : summary.index("run: >-")] + self.assertIn("github.event.action == 'labeled'", step) + self.assertIn("needs.configure.result == 'success'", step) + self.assertIn("run_cpu == 'false'", step) + self.assertIn("run_gpu == 'false'", step) + self.assertIn('--pass-through "${PASS_THROUGH}"', summary) + # The runs API keys a run by the pull request head, not the merge commit. + self.assertIn("github.event.pull_request.head.sha", summary) + + def test_concurrency_group_makes_the_same_test(self): + text = self.WORKFLOW.read_text(encoding="utf-8") + group = text[text.index("concurrency:") : text.index("cancel-in-progress:")] + self.assertIn(f"github.event.label.name != '{REPO.capture_label}'", group) + self.assertNotIn("ci:gfx", group) + + +class ForkTest(unittest.TestCase): + REPO = "ROCm/rocm-systems" + + def test_same_repository_is_not_a_fork(self): + self.assertFalse(cfg.is_fork("pull_request", self.REPO, self.REPO)) + + def test_other_repository_is_a_fork(self): + self.assertTrue(cfg.is_fork("pull_request", "someone/rocm-systems", self.REPO)) + + def test_deleted_head_repository_is_a_fork(self): + self.assertTrue(cfg.is_fork("pull_request", "", self.REPO)) + + def test_only_pull_requests_are_forks(self): + self.assertFalse(cfg.is_fork("push", "", self.REPO)) + + def test_main_skips_real_gpu_for_a_deleted_head_repository(self): + with tempfile.TemporaryDirectory() as tmp, mock.patch.object( + cfg, "list_pr_files", return_value=["projects/hrr/playback/x.cpp"] + ), mock.patch.object( + cfg, "load_matrix", return_value=(MATRIX, pick_first) + ), mock.patch.dict( + os.environ, {"GITHUB_OUTPUT": ""} + ): + path = Path(tmp) / "ci.yaml" + path.write_text(yaml.safe_dump(gpu_data())) + out = io.StringIO() + with redirect_stdout(out): + status = cfg.main( + [ + "--event", + "pull_request", + "--repo", + self.REPO, + "--pr-number", + "1", + "--head-repo", + "", + "--therock-dir", + tmp, + "--config", + str(path), + ] + ) + self.assertEqual(status, 0) + self.assertIn("run_cpu=true\n", out.getvalue()) + self.assertIn("run_gpu=false\n", out.getvalue()) + self.assertIn("fork", out.getvalue()) + + +class OutputTest(unittest.TestCase): + def test_write_outputs(self): + d = cfg.Decision( + True, + True, + "why", + [{"family": "gfx94x", "runner": "r"}], + [{"family": "gfx125x", "reason": "disabled in ci.yaml"}], + ) + with tempfile.TemporaryDirectory() as tmp: + out = Path(tmp) / "out" + with redirect_stdout(io.StringIO()): + cfg.write_outputs(d, "1.2.3", str(out)) + text = out.read_text() + self.assertIn("run_cpu=true\n", text) + self.assertIn('families=[{"family":"gfx94x","runner":"r"}]\n', text) + self.assertIn( + 'skipped_families=[{"family":"gfx125x","reason":"disabled in ci.yaml"}]\n', + text, + ) + self.assertIn("reason=why\n", text) + self.assertIn("rocm_sdk_version=1.2.3\n", text) + + +class WorkflowPathsFilterTest(unittest.TestCase): + ROOT = Path(__file__).resolve().parents[4] + + @staticmethod + def glob(path: str) -> str: + return path + "**" if path.endswith("/") else path + + def expected(self, config: cfg.Config) -> tuple[set[str], set[str]]: + """ci.yaml's path lists as a GitHub paths filter: entries under a listed + directory fold into that directory's glob.""" + union = set(config.cpu_paths + config.capture_paths + config.suite_paths) + dirs = [p for p in union if p.endswith("/")] + kept = {p for p in union if not any(p != d and p.startswith(d) for d in dirs)} + return ( + {self.glob(p) for p in kept}, + {"!" + self.glob(p) for p in config.excluded_paths}, + ) + + def test_hrr_ci_paths_filter_matches_ci_yaml(self): + config = cfg.load_config(self.ROOT / "projects/hrr/tests/config/ci.yaml") + # ci.yaml matches every entry as a prefix; the filter matches an entry + # without a trailing "/" only as that exact file. + for path in ( + config.cpu_paths + + config.capture_paths + + config.suite_paths + + config.excluded_paths + ): + if not path.endswith("/"): + self.assertTrue((self.ROOT / path).is_file(), path) + include, exclude = self.expected(config) + workflow = yaml.safe_load( + (self.ROOT / ".github/workflows/hrr-ci.yml").read_text(encoding="utf-8") + ) + # PyYAML reads the bare key `on` as True. + triggers = workflow.get("on", workflow.get(True)) + for event in ("pull_request", "push"): + with self.subTest(event=event): + paths = triggers[event]["paths"] + self.assertEqual(len(paths), len(set(paths))) + self.assertEqual({p for p in paths if not p.startswith("!")}, include) + self.assertEqual({p for p in paths if p.startswith("!")}, exclude) + # A negated pattern only drops what an earlier pattern matched. + first_negated = next( + i for i, p in enumerate(paths) if p.startswith("!") + ) + self.assertTrue(all(p.startswith("!") for p in paths[first_negated:])) + + +if __name__ == "__main__": + unittest.main() diff --git a/projects/hrr/tests/scripts/test_hrr_ci_report.py b/projects/hrr/tests/scripts/test_hrr_ci_report.py new file mode 100644 index 000000000000..e974700de2d9 --- /dev/null +++ b/projects/hrr/tests/scripts/test_hrr_ci_report.py @@ -0,0 +1,457 @@ +#!/usr/bin/env python3 +# Copyright (c) Advanced Micro Devices, Inc., or its affiliates. +# SPDX-License-Identifier: MIT +"""Tests for hrr_ci_report.py.""" + +import contextlib +import io +import json +import os +import sys +import subprocess +import tempfile +import unittest +from pathlib import Path +from unittest import mock + +sys.path.insert(0, str(Path(__file__).resolve().parent)) + +import hrr_ci_report as rpt # noqa: E402 + +JOBS = [ + {"name": "Configure", "conclusion": "success", "html_url": "u0"}, + {"name": "API IDs and workloads", "conclusion": "success", "html_url": "c0"}, + {"name": "Unit (ubuntu-24.04)", "conclusion": "success", "html_url": "u1"}, + {"name": "Unit (windows-2022)", "conclusion": "failure", "html_url": "u2"}, + {"name": "gpu / Build capture-enabled HIP and HRR", "conclusion": "success", "html_url": "u3"}, + {"name": "gpu / Integration (gfx94x)", "conclusion": "success", "html_url": "u4"}, + {"name": "gpu / Integration (gfx950)", "conclusion": "cancelled", "html_url": "u5"}, +] + +FAMILIES = [{"family": "gfx94x"}, {"family": "gfx950"}, {"family": "gfx110x"}] + + +class ClassifyTest(unittest.TestCase): + def test_picks_hrr_jobs(self): + rows = rpt.classify_jobs(JOBS) + self.assertEqual( + [(r["kind"], r["name"]) for r in rows], + [ + ("API checks", ""), + ("Unit", "ubuntu-24.04"), + ("Unit", "windows-2022"), + ("Build", ""), + ("Integration", "gfx94x"), + ("Integration", "gfx950"), + ], + ) + + def test_missing_family(self): + rows = rpt.classify_jobs(JOBS) + self.assertEqual(rpt.missing_families(FAMILIES, rows), ["gfx110x"]) + + def test_skipped_matrix_job_with_unexpanded_name_is_dropped(self): + rows = rpt.classify_jobs( + [ + {"name": "GPU / Integration (${{ matrix.family }})", "conclusion": "skipped"}, + {"name": "GPU / Integration (gfx94x)", "conclusion": "success"}, + ] + ) + self.assertEqual([(r["kind"], r["name"]) for r in rows], [("Integration", "gfx94x")]) + + def test_queued_job_reports_status(self): + rows = rpt.classify_jobs([{"name": "gpu / Integration (gfx90a)", "status": "queued"}]) + self.assertEqual(rows[0]["conclusion"], "queued") + self.assertFalse(rpt.overall_ok(rows, [])) + + +class RocJitsuJobTest(unittest.TestCase): + def test_rocjitsu_job_counts_as_the_family_integration_job(self): + rows = rpt.classify_jobs( + [ + {"name": "GPU / rocJitsu (gfx90a)", "conclusion": "success", "html_url": "u1"}, + {"name": "GPU / rocJitsu (gfx94x)", "conclusion": "failure", "html_url": "u2"}, + ] + ) + self.assertEqual( + [(r["kind"], r["name"], r["conclusion"]) for r in rows], + [("Integration", "gfx90a", "success"), ("Integration", "gfx94x", "failure")], + ) + families = [{"family": "gfx90a"}, {"family": "gfx94x"}] + self.assertEqual(rpt.missing_families(families, rows), []) + + + def test_a_failed_rocjitsu_build_leaves_its_families_missing(self): + # The simulator jobs skip, and a skipped matrix job keeps its + # unexpanded name, so the build's failure shows as missing families. + rows = rpt.classify_jobs( + [ + {"name": "GPU / Build rocJitsu", "conclusion": "failure", "html_url": "u0"}, + {"name": "GPU / rocJitsu (${{ matrix.family }})", "conclusion": "skipped"}, + ] + ) + missing = rpt.missing_families([{"family": "gfx90a"}], rows) + self.assertEqual(missing, ["gfx90a"]) + self.assertFalse(rpt.overall_ok(rows, missing)) + + +class GpuDispatchTest(unittest.TestCase): + def test_skipped_rocjitsu_build_does_not_fail_a_gpu_dispatch(self): + # A gpu dispatch runs the real-GPU jobs; "Build rocJitsu" is skipped. + # It is no HRR job row, so the green Integration rows must pass. + rows = rpt.classify_jobs( + [ + {"name": "GPU / Build rocJitsu", "conclusion": "skipped"}, + {"name": "GPU / Integration (gfx94x)", "conclusion": "success", "html_url": "u1"}, + {"name": "GPU / Integration (gfx950)", "conclusion": "success", "html_url": "u2"}, + ] + ) + families = [{"family": "gfx94x"}, {"family": "gfx950"}] + missing = rpt.missing_families(families, rows) + self.assertEqual(missing, []) + self.assertTrue(rpt.overall_ok(rows, missing)) + + +class OverallTest(unittest.TestCase): + def test_failure_or_cancel_fails(self): + self.assertFalse(rpt.overall_ok(rpt.classify_jobs(JOBS), [])) + + def test_all_green(self): + rows = rpt.classify_jobs([JOBS[1], JOBS[2], JOBS[5]]) + self.assertTrue(rpt.overall_ok(rows, [])) + + def test_missing_family_fails(self): + rows = rpt.classify_jobs([JOBS[5]]) + self.assertFalse(rpt.overall_ok(rows, ["gfx950"])) + + +class RenderTest(unittest.TestCase): + def test_header_names_base(self): + base = {"repo": "ROCm/rocm-systems", "run_id": "7", "run_url": "https://r/7", "commit": "0123456789abcdef"} + families = [{"family": "gfx94x", "runner": "linux-gfx942-1gpu"}, {"family": "gfx950"}] + text = rpt.render_header("the change touches capture or replay", True, families, base) + self.assertIn("GPU families and runners: gfx94x (`linux-gfx942-1gpu`), gfx950.", text) + self.assertIn("[ROCm/rocm-systems run 7](https://r/7) at `0123456789ab`", text) + + def test_header_without_gpu(self): + self.assertIn("GPU suite: not run.", rpt.render_header("why", False, [], None)) + + def test_jobs_table_lists_missing(self): + rows = rpt.classify_jobs(JOBS) + text = rpt.render_jobs(rows, ["gfx110x"]) + self.assertIn("| Integration | gfx950 | [cancelled](u5) |", text) + self.assertIn("| Integration | gfx110x | _did not run_ |", text) + + def test_main_appends_results_table(self): + with tempfile.TemporaryDirectory() as tmp, \ + mock.patch.object(rpt, "list_jobs", return_value=JOBS[:3]), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()) as out: + rc = rpt.main(["--reason", "why", "--results-dir", tmp]) + self.assertEqual(rc, 0) + self.assertIn("## HRR results by platform", out.getvalue()) + + def test_main_gives_skipped_families_a_row(self): + with tempfile.TemporaryDirectory() as tmp, \ + mock.patch.object(rpt, "list_jobs", return_value=JOBS[5:6]), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()) as out: + rc = rpt.main([ + "--run-gpu", "true", "--results-dir", tmp, + "--families", '[{"family":"gfx94x","runner":"r"}]', + "--skipped-families", '[{"family":"gfx125x","reason":"disabled in ci.yaml"}]', + ]) + self.assertEqual(rc, 0) + self.assertIn("| Integration | gfx94x | — | — | — | — | _no results_ |", out.getvalue()) + self.assertIn("| Integration | gfx125x | — | — | — | — | _not run: disabled in ci.yaml_ |", + out.getvalue()) + + def test_main_reports_job_listing_failure(self): + err = subprocess.CalledProcessError(1, "gh", stderr="HTTP 403: rate limited\n") + with tempfile.TemporaryDirectory() as tmp, \ + mock.patch.object(rpt, "list_jobs", side_effect=err), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()) as out: + rc = rpt.main(["--run-id", "7", "--results-dir", tmp]) + self.assertEqual(rc, 1) + self.assertIn("::error::could not list the jobs of run 7: HTTP 403: rate limited", out.getvalue()) + + def test_main_fails_when_configure_did_not_succeed(self): + for result in ("failure", "cancelled"): + with tempfile.TemporaryDirectory() as tmp, \ + mock.patch.object(rpt, "list_jobs", return_value=JOBS[:1]), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()) as out: + rc = rpt.main(["--configure-result", result, "--run-gpu", "", "--families", "", + "--results-dir", tmp]) + self.assertEqual(rc, 1, result) + self.assertIn(f"Configure: {result}, so nothing else ran", out.getvalue()) + + def test_run_39793856571_does_not_pass(self): + """The jobs API was still queued, needs.gpu had failed, and the gfx950 + JUnit file already contained Unit_HRR_PinnedHostFork. All three must + fail the summary; any one of them is enough.""" + jobs = [ + {"name": "API IDs and workloads", "conclusion": "success", "html_url": "u0"}, + {"name": "HRR script tests", "conclusion": "success"}, + {"name": "Unit (ubuntu-24.04)", "conclusion": "success"}, + {"name": "Unit (windows-2022)", "conclusion": "success"}, + {"name": "GPU / Build capture-enabled HIP and HRR", "conclusion": "success"}, + {"name": "GPU / rocJitsu (gfx950)", "status": "queued", "html_url": "u2"}, + ] + needs = { + "configure": {"result": "success"}, + "api-checks": {"result": "success"}, + "scripts": {"result": "success"}, + "unit": {"result": "success"}, + "gpu": {"result": "failure"}, + } + failed = ( + '' + '' + "" + ) + with tempfile.TemporaryDirectory() as tmp, \ + mock.patch.object(rpt, "list_jobs", return_value=jobs), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()) as out: + (Path(tmp) / "integration-gfx950.xml").write_text(failed, encoding="utf-8") + rc = rpt.main([ + "--run-gpu", "true", "--results-dir", tmp, + "--families", '[{"family":"gfx950","runner":"ubuntu-24.04"}]', + "--needs-json", json.dumps(needs), + ]) + self.assertEqual(rc, 1) + text = out.getvalue() + self.assertIn("gpu is failure", text) + self.assertIn("failed cases in Integration gfx950", text) + self.assertIn("| Integration | gfx950 | [queued](u2) |", text) + + def test_junit_failure_fails_when_jobs_and_needs_succeeded(self): + jobs = [ + {"name": "API IDs and workloads", "conclusion": "success"}, + {"name": "HRR script tests", "conclusion": "success"}, + {"name": "Unit (ubuntu-24.04)", "conclusion": "success"}, + {"name": "GPU / rocJitsu (gfx94x)", "conclusion": "success"}, + ] + needs = {name: {"result": "success"} for name in ( + "configure", "api-checks", "scripts", "unit", "gpu")} + failed = ( + '' + '' + "" + ) + with tempfile.TemporaryDirectory() as tmp, \ + mock.patch.object(rpt, "list_jobs", return_value=jobs), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()): + (Path(tmp) / "integration-gfx94x.xml").write_text(failed, encoding="utf-8") + rc = rpt.main([ + "--run-gpu", "true", "--results-dir", tmp, + "--families", '[{"family":"gfx94x"}]', + "--needs-json", json.dumps(needs), + ]) + self.assertEqual(rc, 1) + + def test_a_selected_cpu_job_that_was_skipped_fails(self): + needs = { + "configure": {"result": "success"}, + "api-checks": {"result": "skipped"}, + "scripts": {"result": "success"}, + "unit": {"result": "success"}, + "gpu": {"result": "skipped"}, + } + with tempfile.TemporaryDirectory() as tmp, \ + mock.patch.object(rpt, "list_jobs", return_value=JOBS[:3]), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()) as out: + rc = rpt.main(["--results-dir", tmp, "--needs-json", json.dumps(needs)]) + self.assertEqual(rc, 1) + self.assertIn("api-checks is skipped", out.getvalue()) + + def test_green_needs_jobs_and_junit_pass(self): + jobs = [ + {"name": "API IDs and workloads", "conclusion": "success"}, + {"name": "HRR script tests", "conclusion": "success"}, + {"name": "Unit (ubuntu-24.04)", "conclusion": "success"}, + {"name": "Unit (windows-2022)", "conclusion": "success"}, + {"name": "GPU / rocJitsu (gfx94x)", "conclusion": "success"}, + ] + needs = {name: {"result": "success"} for name in ( + "configure", "api-checks", "scripts", "unit", "gpu")} + passed = ( + '' + '' + ) + with tempfile.TemporaryDirectory() as tmp, \ + mock.patch.object(rpt, "list_jobs", return_value=jobs), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()): + (Path(tmp) / "integration-gfx94x.xml").write_text(passed, encoding="utf-8") + rc = rpt.main([ + "--run-gpu", "true", "--results-dir", tmp, + "--families", '[{"family":"gfx94x"}]', + "--needs-json", json.dumps(needs), + ]) + self.assertEqual(rc, 0) + + +class GhApiTest(unittest.TestCase): + def test_reads_every_value_across_pages(self): + out = subprocess.CompletedProcess([], 0, stdout='{"a": 1}\n{"a": 2}{"a": 3}\n7\n') + with mock.patch.object(rpt.subprocess, "run", return_value=out) as run: + self.assertEqual(rpt.gh_api("p", ".x[]"), [{"a": 1}, {"a": 2}, {"a": 3}, 7]) + self.assertEqual(run.call_args.args[0], ["gh", "api", "--paginate", "p", "--jq", ".x[]"]) + + def test_empty_output_is_no_values(self): + out = subprocess.CompletedProcess([], 0, stdout="\n") + with mock.patch.object(rpt.subprocess, "run", return_value=out): + self.assertEqual(rpt.gh_api("p", ".x[]"), []) + + def test_unreadable_workflow_id_fails_the_summary(self): + with mock.patch.object(rpt, "gh_api", return_value=[]), \ + contextlib.redirect_stdout(io.StringIO()) as out: + rc = rpt.main(["--repo", "R", "--pass-through", "true", "--head-sha", "abc", + "--run-id", "100"]) + self.assertEqual(rc, 1) + self.assertIn("could not find the HRR CI run that tested abc", out.getvalue()) + + +class PassThroughTest(unittest.TestCase): + """A label run with nothing to run reports the run that tested the + commit; it must never report greener than that run.""" + LABEL_ARGS = ["--pass-through", "true", "--head-sha", "abc", + "--run-id", "100", "--run-attempt", "1"] + GREEN = [JOBS[1], JOBS[2], JOBS[5]] # generated, unit ubuntu, integration gfx94x + SUMMARY_OK = {"name": "HRR CI Summary", "conclusion": "success", "html_url": "s0"} + SUMMARY_RED = {"name": "HRR CI Summary", "conclusion": "failure", "html_url": "s1"} + GREEN = GREEN + [SUMMARY_OK] + # What a run whose configure selected nothing lists: its jobs, skipped. + LABEL_ONLY = [ + {"name": "Configure", "conclusion": "success", "html_url": "u0"}, + {"name": "API IDs and workloads", "conclusion": "skipped", "html_url": "k1"}, + {"name": "HRR script tests", "conclusion": "skipped", "html_url": "k2"}, + {"name": "Unit (${{ matrix.os }})", "conclusion": "skipped", "html_url": "k3"}, + {"name": "GPU", "conclusion": "skipped", "html_url": "k4"}, + SUMMARY_OK, + ] + + def run_main(self, runs, extra=()): + """runs: [(run id, attempt, created_at, jobs[, run_started_at])], as + the API lists them.""" + listing = [{"id": rid, "run_attempt": att, "created_at": at, + "run_started_at": started[0] if started else at, "html_url": f"r{rid}"} + for rid, att, at, _, *started in runs] + jobs = {(str(rid), str(att)): j for rid, att, _, j, *_ in runs} + + def gh_api(path, jq): + if path == "repos/R/actions/runs/100": + return [7] + if path.startswith("repos/R/actions/workflows/7/runs?event=pull_request&head_sha=abc&"): + return listing + rid, att = path.split("/runs/")[1].split("/attempts/") + return jobs[(rid, att.split("/")[0])] + + tmp = tempfile.TemporaryDirectory() + self.addCleanup(tmp.cleanup) + with mock.patch.object(rpt, "gh_api", side_effect=gh_api), \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()) as out: + rc = rpt.main(["--results-dir", tmp.name, "--repo", "R", + *self.LABEL_ARGS, *extra]) + return rc, out.getvalue() + + def test_delivers_the_earlier_run_jobs(self): + rc, text = self.run_main([(99, 1, "2026-10-15T10:00:00Z", self.GREEN)]) + self.assertEqual(rc, 0, text) + self.assertIn("[run 99 attempt 1](r99)", text) + self.assertIn("| API checks | — | [success](c0) |", text) + + def test_fails_when_the_earlier_run_failed(self): + rc, text = self.run_main([(99, 1, "2026-10-15T10:00:00Z", JOBS + [self.SUMMARY_RED])]) + self.assertEqual(rc, 1) + self.assertIn("a label event must not turn it into a pass", text) + + def test_fails_when_only_the_earlier_summary_failed(self): + # e.g. a family that never reported: every HRR job is green. + jobs = self.GREEN[:-1] + [self.SUMMARY_RED] + rc, text = self.run_main([(99, 1, "2026-10-15T10:00:00Z", jobs)]) + self.assertEqual(rc, 1) + self.assertIn("HRR CI Summary is failure", text) + + def test_skips_label_runs_and_takes_the_newest_tested_run(self): + rc, text = self.run_main([ + (98, 1, "2026-10-15T09:00:00Z", self.GREEN), + (99, 2, "2026-10-15T10:00:00Z", JOBS + [self.SUMMARY_RED]), # newest that tested + (101, 1, "2026-10-15T11:00:00Z", self.LABEL_ONLY), # an earlier label run + # This run, still in progress, with a green-looking job of its own. + (100, 1, "2026-10-15T12:00:00Z", self.GREEN[:-1]), + ]) + self.assertEqual(rc, 1, text) + self.assertIn("[run 99 attempt 2](r99)", text) + + def test_fails_on_a_job_that_has_not_finished(self): + jobs = [{"name": "HRR script tests", "conclusion": None, "status": "in_progress", + "html_url": "u9"}, + {"name": "HRR CI Summary", "conclusion": None, "status": "queued"}] + rc, text = self.run_main([(99, 1, "2026-10-15T10:00:00Z", jobs)]) + self.assertEqual(rc, 1) + self.assertIn("is in_progress", text) + + def test_fails_when_the_earlier_configure_failed(self): + jobs = [{**j, "conclusion": "failure"} if j["name"] == "Configure" else j + for j in self.LABEL_ONLY if j["name"] != "HRR CI Summary"] + [self.SUMMARY_RED] + rc, text = self.run_main([(99, 1, "2026-10-15T10:00:00Z", jobs)]) + self.assertEqual(rc, 1, text) + self.assertIn("[run 99 attempt 1](r99)", text) + self.assertIn("HRR CI Summary is failure", text) + + def test_an_unfinished_configure_has_a_verdict_to_wait_for(self): + jobs = [{"name": "Configure", "conclusion": None, "status": "in_progress"}, + {"name": "HRR CI Summary", "conclusion": None, "status": "queued"}] + rc, text = self.run_main([(99, 1, "2026-10-15T10:00:00Z", jobs)]) + self.assertEqual(rc, 1, text) + self.assertIn("[run 99 attempt 1](r99)", text) + + def test_a_rerun_counts_from_its_latest_attempt(self): + rc, text = self.run_main([ + (98, 2, "2026-10-15T09:00:00Z", JOBS + [self.SUMMARY_RED], "2026-10-15T12:00:00Z"), + (99, 1, "2026-10-15T10:00:00Z", self.GREEN), + ]) + self.assertEqual(rc, 1, text) + self.assertIn("[run 98 attempt 2](r98)", text) + + def test_needs_the_head_sha(self): + with self.assertRaises(SystemExit), contextlib.redirect_stderr(io.StringIO()): + rpt.main(["--repo", "R", "--pass-through", "true", "--run-id", "100"]) + + def test_passes_when_no_run_tested_the_commit(self): + # A pull request that touches no HRR path: no run ran an HRR job. + rc, text = self.run_main([(99, 1, "2026-10-15T10:00:00Z", self.LABEL_ONLY)]) + self.assertEqual(rc, 0, text) + self.assertIn("nothing to report", text) + + def test_fails_when_the_lookup_fails(self): + err = subprocess.CalledProcessError(1, "gh", stderr="HTTP 502") + tmp = tempfile.TemporaryDirectory() + self.addCleanup(tmp.cleanup) + with mock.patch.object(rpt, "gh_api", side_effect=err), \ + contextlib.redirect_stdout(io.StringIO()) as out: + rc = rpt.main(["--results-dir", tmp.name, "--repo", "R", *self.LABEL_ARGS]) + self.assertEqual(rc, 1) + self.assertIn("HTTP 502", out.getvalue()) + + def test_ordinary_run_reports_its_own_jobs(self): + tmp = tempfile.TemporaryDirectory() + self.addCleanup(tmp.cleanup) + with mock.patch.object(rpt, "list_jobs", return_value=self.GREEN) as listed, \ + mock.patch.dict(os.environ, {"GITHUB_STEP_SUMMARY": ""}), \ + contextlib.redirect_stdout(io.StringIO()): + rc = rpt.main(["--results-dir", tmp.name, "--repo", "R", "--pass-through", "false", + "--run-id", "100", "--run-attempt", "2"]) + self.assertEqual(rc, 0) + listed.assert_called_once_with("R", "100", "2") + + +if __name__ == "__main__": + unittest.main() diff --git a/projects/hrr/tests/scripts/test_hrr_test_config.py b/projects/hrr/tests/scripts/test_hrr_test_config.py index af5c45560a6b..ffd11cd78062 100644 --- a/projects/hrr/tests/scripts/test_hrr_test_config.py +++ b/projects/hrr/tests/scripts/test_hrr_test_config.py @@ -110,6 +110,25 @@ def test_no_device_does_not_match_an_arch_rule(self): run, _ = cfg.resolve(suite(cases), ["Filtered"], "linux", []) self.assertEqual(run, ["Filtered"]) + def test_emulator_rule_matches_only_an_emulated_run(self): + cases = {"Filtered": {"skip": [rule({"emulator": True, "arch": "gfx1151"})]}} + config = suite(cases) + run, skipped = cfg.resolve( + config, ["Filtered"], "linux", ["gfx1151"], emulated=True + ) + self.assertEqual(run, []) + self.assertEqual(skipped[0]["archs"], ["gfx1151"]) + run, _ = cfg.resolve(config, ["Filtered"], "linux", ["gfx1151"]) + self.assertEqual(run, ["Filtered"]) + + def test_emulator_false_rule_matches_only_a_real_device(self): + cases = {"Filtered": {"skip": [rule({"emulator": False})]}} + config = suite(cases) + run, _ = cfg.resolve(config, ["Filtered"], "linux", ["gfx90a"], emulated=True) + self.assertEqual(run, ["Filtered"]) + run, _ = cfg.resolve(config, ["Filtered"], "linux", ["gfx90a"]) + self.assertEqual(run, []) + def test_ignore_runs_everything(self): cases = {"Filtered": {"skip": [rule({"os": "linux"})]}} config = suite(cases, skip=[rule({"arch": "gfx90a"})]) @@ -126,6 +145,13 @@ def test_a_valid_config_is_quiet(self): ]}} self.assertEqual(cfg.lint(suite(cases)), []) + def test_emulator_must_be_a_bool(self): + cases = {"Filtered": {"skip": [rule({"emulator": "yes"})]}} + errors = cfg.lint(suite(cases)) + self.assertTrue(any("emulator" in error for error in errors)) + cases = {"Filtered": {"skip": [rule({"emulator": True, "arch": "gfx90a"})]}} + self.assertEqual(cfg.lint(suite(cases)), []) + def test_disabled_needs_an_issue_and_every_rule_needs_a_reason(self): cases = {"Filtered": {"skip": [{"when": {"os": "linux"}, "kind": "disabled"}]}} errors = cfg.lint(suite(cases)) @@ -155,11 +181,6 @@ def test_a_direct_workload_cannot_carry_rules(self): self.assertTrue(any("workload" in error for error in errors)) self.assertEqual(cfg.lint(suite({"Unit_HRR_Foo_Direct": {}}), cases=listed), []) - def test_workflow_archs_must_match(self): - errors = cfg.lint(suite({"Runs": {}}), workflow_archs=["gfx90a"]) - self.assertTrue(any("HRR_ARCHS" in error for error in errors)) - self.assertEqual(cfg.lint(suite({"Runs": {}}), workflow_archs=list(ARCHS)), []) - class ListingTest(unittest.TestCase): def test_json_listing_keeps_tags_and_passes_the_selector(self): @@ -204,6 +225,56 @@ def test_check_reports_both_directions(self): self.assertIn("UNEXPECTED", err.getvalue()) self.assertIn("New", err.getvalue()) + def test_check_allows_a_case_compiled_out_where_it_is_unsupported(self): + self.write(suite({ + "Kept": {}, + "PosixOnly": {"skip": [rule({"os": "windows"})]}, + "OneArch": {"skip": [rule({"os": "windows", "arch": "gfx90a"})]}, + })) + payload = listing(("Kept", [])) + for os_name, rc, text in (("windows", 1, "OneArch"), ("linux", 1, "PosixOnly")): + with self.subTest(os=os_name): + with mock.patch.object(cfg, "detect_os", return_value=os_name), \ + mock.patch.object(cfg.subprocess, "run", return_value=completed(payload)): + err = io.StringIO() + with contextlib.redirect_stderr(err): + got = cfg.main(["check", "--binary", "./bin", "--config", str(self.path)]) + self.assertEqual(got, rc) + self.assertIn("MISSING", err.getvalue()) + self.assertIn(text, err.getvalue()) + if os_name == "windows": + self.assertNotIn("PosixOnly", err.getvalue()) + + def test_supported_case_compiled_in_is_not_unexpected(self): + # An OS-wide unsupported rule (even an empty one) only relaxes the + # MISSING side: the case stays in the config, so a binary that still + # compiles it is correct and must not be reported UNEXPECTED. + self.write(suite({ + "Kept": {}, + "NoWhen": {"skip": [rule({})]}, + "LinuxOnly": {"skip": [rule({"os": "linux"})]}, + })) + payload = listing(("Kept", []), ("NoWhen", []), ("LinuxOnly", [])) + with mock.patch.object(cfg, "detect_os", return_value="linux"), \ + mock.patch.object(cfg.subprocess, "run", return_value=completed(payload)): + out, err = io.StringIO(), io.StringIO() + with contextlib.redirect_stdout(out), contextlib.redirect_stderr(err): + rc = cfg.main(["check", "--binary", "./bin", "--config", str(self.path)]) + self.assertEqual(rc, 0, err.getvalue()) + self.assertNotIn("UNEXPECTED", err.getvalue()) + self.assertIn("OK: 3 test cases present", out.getvalue()) + + def test_case_not_in_the_config_is_still_unexpected(self): + self.write(suite({"Kept": {}, "LinuxOnly": {"skip": [rule({"os": "linux"})]}})) + payload = listing(("Kept", []), ("LinuxOnly", []), ("Bogus", [])) + with mock.patch.object(cfg, "detect_os", return_value="linux"), \ + mock.patch.object(cfg.subprocess, "run", return_value=completed(payload)): + err = io.StringIO() + with contextlib.redirect_stderr(err): + rc = cfg.main(["check", "--binary", "./bin", "--config", str(self.path)]) + self.assertEqual(rc, 1) + self.assertIn("UNEXPECTED (added without updating the config): Bogus", err.getvalue()) + def test_update_keeps_rules_and_adds_empty_entries(self): kept = rule({"os": "linux", "arch": "gfx1151"}, "disabled", "npu", "https://github.com/ROCm/rocm-systems/issues/1") @@ -259,15 +330,6 @@ def test_nothing_visible_is_an_empty_list(self): self.assertEqual(cfg.detect_archs(lambda argv: None), []) -class WorkflowTest(unittest.TestCase): - def test_reads_hrr_archs(self): - tmp = tempfile.NamedTemporaryFile("w", delete=False, encoding="utf-8") - self.addCleanup(lambda: Path(tmp.name).unlink(missing_ok=True)) - tmp.write('env:\n HRR_ARCHS: "gfx90a;gfx942"\n') - tmp.close() - self.assertEqual(cfg.workflow_archs(tmp.name), ["gfx90a", "gfx942"]) - - class LintCommandTest(unittest.TestCase): def test_lint_command_fails_on_a_bad_file(self): tmp = tempfile.NamedTemporaryFile("w", suffix=".yaml", delete=False, encoding="utf-8") @@ -280,6 +342,29 @@ def test_lint_command_fails_on_a_bad_file(self): self.assertEqual(rc, 1) self.assertIn("version", err.getvalue()) + def test_lint_command_fails_when_targets_arch_differ(self): + other = suite({"Runs": {}}) + other["targets"]["arch"] = ARCHS[:-1] + ["gfx1200"] + paths = [] + for config in (suite({"Runs": {}}), other): + tmp = tempfile.NamedTemporaryFile("w", suffix=".yaml", delete=False, encoding="utf-8") + self.addCleanup(lambda name=tmp.name: Path(name).unlink(missing_ok=True)) + tmp.write(cfg.render(config)) + tmp.close() + paths.append(tmp.name) + err = io.StringIO() + with contextlib.redirect_stderr(err), contextlib.redirect_stdout(io.StringIO()): + rc = cfg.main(["lint", *paths]) + self.assertEqual(rc, 1) + self.assertIn("missing: gfx1151; extra: gfx1200", err.getvalue()) + with contextlib.redirect_stdout(io.StringIO()): + self.assertEqual(cfg.main(["lint", paths[0], paths[0]]), 0) + + def test_repo_suites_share_targets_arch(self): + config_dir = Path(__file__).resolve().parent.parent / "config" + configs = {name: cfg.load(config_dir / name) for name in ("unit.yaml", "integration.yaml")} + self.assertEqual(cfg.arch_mismatches(configs), []) + if __name__ == "__main__": unittest.main() diff --git a/projects/hrr/tests/scripts/test_run_catch2.py b/projects/hrr/tests/scripts/test_run_catch2.py index 6e7b5552a758..c8468f492f1d 100644 --- a/projects/hrr/tests/scripts/test_run_catch2.py +++ b/projects/hrr/tests/scripts/test_run_catch2.py @@ -68,10 +68,20 @@ def test_timeout_is_optional(self): def test_launcher_stays_off_the_catch2_binary(self): args = run_catch2.parse_args( - ["--xml", "r.xml", "--launcher", "rocjitsu --daemon --config c.json --", "--", "./bin"] + [ + "--xml", + "r.xml", + "--launcher", + "rocjitsu --daemon --config c.json --", + "--", + "./bin", + ] ) self.assertEqual(args.command, ["./bin"]) - self.assertEqual(args.launcher, "rocjitsu --daemon --config c.json --") + self.assertEqual( + run_catch2.launch_command(args.launcher, args.command), + ["rocjitsu", "--daemon", "--config", "c.json", "--", "./bin"], + ) def test_requires_a_command(self): with contextlib.redirect_stderr(io.StringIO()): @@ -90,13 +100,25 @@ def parse(self, text: str) -> ET.Element: def test_test_result(self): self.assertEqual(run_catch2.test_result(self.parse(record("a"))), "PASS") - self.assertEqual(run_catch2.test_result(self.parse(record("a", "skipped"))), "SKIP") - self.assertEqual(run_catch2.test_result(self.parse(record("a", "failure"))), "FAIL") - self.assertEqual(run_catch2.test_result(self.parse(record("a", "error"))), "FAIL") + self.assertEqual( + run_catch2.test_result(self.parse(record("a", "skipped"))), "SKIP" + ) + self.assertEqual( + run_catch2.test_result(self.parse(record("a", "failure"))), "FAIL" + ) + self.assertEqual( + run_catch2.test_result(self.parse(record("a", "error"))), "FAIL" + ) def test_group_cases_folds_sections_into_their_case(self): root = self.parse( - junit(record("A"), record("B/x"), record("B/y"), record("C/s/t"), '') + junit( + record("A"), + record("B/x"), + record("B/y"), + record("C/s/t"), + "", + ) ) groups = run_catch2.group_cases(root) self.assertEqual(list(groups), ["A", "B", "C"]) @@ -104,7 +126,9 @@ def test_group_cases_folds_sections_into_their_case(self): def test_group_result_precedence(self): def result(*kinds): - root = self.parse(junit(*(record("A/" + str(i), k) for i, k in enumerate(kinds)))) + root = self.parse( + junit(*(record("A/" + str(i), k) for i, k in enumerate(kinds))) + ) return run_catch2.group_result(run_catch2.group_cases(root)["A"]) self.assertEqual(result(None, None), "PASS") @@ -125,7 +149,15 @@ def setUp(self): def run_main(self, spec: dict, *extra: str): spec_path = self.dir / "spec.json" spec_path.write_text(json.dumps(spec)) - argv = ["--xml", str(self.xml), *extra, "--", sys.executable, str(self.fake), str(spec_path)] + argv = [ + "--xml", + str(self.xml), + *extra, + "--", + sys.executable, + str(self.fake), + str(spec_path), + ] out, err = io.StringIO(), io.StringIO() with contextlib.redirect_stdout(out), contextlib.redirect_stderr(err): rc = run_catch2.main(argv) @@ -182,7 +214,9 @@ def test_exit_4_with_a_failure_stays_failed(self): self.assertIn("1 failed", out) def test_crash_after_skips_keeps_the_transcript(self): - rc, out, _ = self.run_main({"xml": junit(record("A", "skipped"), record("B")), "rc": 139}) + rc, out, _ = self.run_main( + {"xml": junit(record("A", "skipped"), record("B")), "rc": 139} + ) self.assertEqual(rc, 139) self.assertIn("Catch2 terminated without a JUnit failure", out) self.assertIn("BANNER", out) @@ -205,53 +239,128 @@ def test_missing_junit_with_clean_exit_still_fails(self): def _config(self, cases): path = self.dir / "suite.yaml" - path.write_text(hrr_test_config.render({ - "version": 1, - "suite": "integration", - "binary": "hrr-integration-tests", - "select": "[hrr]~[direct]", - "targets": {"os": ["linux", "windows"], - "arch": ["gfx90a", "gfx1151"]}, - "cases": cases, - })) + path.write_text( + hrr_test_config.render( + { + "version": 1, + "suite": "integration", + "binary": "hrr-integration-tests", + "select": "[hrr]~[direct]", + "targets": { + "os": ["linux", "windows"], + "arch": ["gfx90a", "gfx1151"], + }, + "cases": cases, + } + ) + ) return path def test_config_skips_a_case_and_records_why(self): - path = self._config({ - "Keep": {}, - "Drop": {"skip": [{ - "when": {"os": "linux", "arch": "gfx1151"}, - "kind": "disabled", - "reason": "npu", - "issue": "https://github.com/ROCm/rocm-systems/issues/1", - }]}, - }) - listed = [{"name": "Keep", "tags": ["[hrr]"]}, {"name": "Drop", "tags": ["[hrr]"]}] + path = self._config( + { + "Keep": {}, + "Drop": { + "skip": [ + { + "when": {"os": "linux", "arch": "gfx1151"}, + "kind": "disabled", + "reason": "npu", + "issue": "https://github.com/ROCm/rocm-systems/issues/1", + } + ] + }, + } + ) + listed = [ + {"name": "Keep", "tags": ["[hrr]"]}, + {"name": "Drop", "tags": ["[hrr]"]}, + ] with mock.patch("hrr_test_config.list_cases", return_value=listed): rc, out, _ = self.run_main( {"xml": junit(record("Keep"))}, - "--config", str(path), "--os", "linux", "--arch", "gfx1151", + "--config", + str(path), + "--os", + "linux", + "--arch", + "gfx1151", ) self.assertEqual(rc, 0) self.assertIn("PASS: Keep", out) self.assertIn("SKIP (config): Drop", out) - self.assertIn("hrr-config: linux/gfx1151: npu " - "(https://github.com/ROCm/rocm-systems/issues/1)", out) + self.assertIn( + "hrr-config: linux/gfx1151: npu " + "(https://github.com/ROCm/rocm-systems/issues/1)", + out, + ) self.assertIn("1 config-skipped", out) root = ET.parse(self.xml).getroot() skipped = root.find(".//testcase[@name='Drop']/skipped") self.assertIsNotNone(skipped) self.assertTrue(skipped.attrib["message"].startswith("hrr-config:")) + def test_emulator_rule_applies_only_with_a_launcher(self): + path = self._config( + { + "Keep": {}, + "Drop": { + "skip": [ + { + "when": {"emulator": True}, + "kind": "unsupported", + "reason": "emulator only", + } + ] + }, + } + ) + listed = [ + {"name": "Keep", "tags": ["[hrr]"]}, + {"name": "Drop", "tags": ["[hrr]"]}, + ] + with mock.patch("hrr_test_config.list_cases", return_value=listed): + rc, out, _ = self.run_main( + {"xml": junit(record("Keep"))}, + "--config", str(path), "--os", "linux", "--arch", "gfx1151", + "--launcher", "env", + ) + self.assertEqual(rc, 0) + self.assertIn("SKIP (config): Drop", out) + with mock.patch("hrr_test_config.list_cases", return_value=listed): + rc, out, _ = self.run_main( + {"xml": junit(record("Keep"), record("Drop"))}, + "--config", str(path), "--os", "linux", "--arch", "gfx1151", + ) + self.assertEqual(rc, 0) + self.assertNotIn("SKIP (config)", out) + def test_config_runs_nothing_when_every_case_is_skipped(self): - path = self._config({"Drop": {"skip": [{ - "when": {"os": "linux"}, "kind": "unsupported", "reason": "all", - }]}}) - with mock.patch("hrr_test_config.list_cases", - return_value=[{"name": "Drop", "tags": ["[hrr]"]}]) as listed: + path = self._config( + { + "Drop": { + "skip": [ + { + "when": {"os": "linux"}, + "kind": "unsupported", + "reason": "all", + } + ] + } + } + ) + with mock.patch( + "hrr_test_config.list_cases", + return_value=[{"name": "Drop", "tags": ["[hrr]"]}], + ) as listed: rc, out, _ = self.run_main( {"xml": junit(record("Drop")), "banner": "SHOULD NOT RUN"}, - "--config", str(path), "--os", "linux", "--arch", "none", + "--config", + str(path), + "--os", + "linux", + "--arch", + "none", ) self.assertEqual(rc, 0) self.assertIn("SKIP (config): Drop", out) @@ -260,14 +369,31 @@ def test_config_runs_nothing_when_every_case_is_skipped(self): self.assertEqual(listed.call_args.args[1], "[hrr]~[direct]") def test_ignore_config_runs_the_skipped_case(self): - path = self._config({"Drop": {"skip": [{ - "when": {"os": "linux"}, "kind": "unsupported", "reason": "all", - }]}}) - with mock.patch("hrr_test_config.list_cases", - return_value=[{"name": "Drop", "tags": ["[hrr]"]}]) : + path = self._config( + { + "Drop": { + "skip": [ + { + "when": {"os": "linux"}, + "kind": "unsupported", + "reason": "all", + } + ] + } + } + ) + with mock.patch( + "hrr_test_config.list_cases", + return_value=[{"name": "Drop", "tags": ["[hrr]"]}], + ): rc, out, _ = self.run_main( {"xml": junit(record("Drop")), "rc": 0}, - "--config", str(path), "--os", "linux", "--arch", "gfx1151", + "--config", + str(path), + "--os", + "linux", + "--arch", + "gfx1151", "--ignore-config", ) self.assertEqual(rc, 0) @@ -280,16 +406,36 @@ def test_select_overrides_the_config_and_a_trailing_filter_is_dropped(self): spec_path = self.dir / "spec.json" # A non-zero status with no JUnit failure dumps the transcript, which is # the only place the wrapper's argv (and the case file) are visible. - spec_path.write_text(json.dumps({ - "xml": junit(record("Keep")), "echo_args": True, "rc": 139, - })) + spec_path.write_text( + json.dumps( + { + "xml": junit(record("Keep")), + "echo_args": True, + "rc": 139, + } + ) + ) argv = [ - "--xml", str(self.xml), "--config", str(path), "--os", "linux", - "--arch", "none", "--select", "[cpu]", - "--", sys.executable, str(self.fake), str(spec_path), "[hrr]~[direct]", + "--xml", + str(self.xml), + "--config", + str(path), + "--os", + "linux", + "--arch", + "none", + "--select", + "[cpu]", + "--", + sys.executable, + str(self.fake), + str(spec_path), + "[hrr]~[direct]", ] - with mock.patch("hrr_test_config.list_cases", - return_value=[{"name": "Keep", "tags": ["[cpu]"]}]) as listed: + with mock.patch( + "hrr_test_config.list_cases", + return_value=[{"name": "Keep", "tags": ["[cpu]"]}], + ) as listed: out = io.StringIO() with contextlib.redirect_stdout(out): rc = run_catch2.main(argv) @@ -305,7 +451,12 @@ def test_auto_arch_with_no_device_fails_closed(self): with mock.patch("hrr_test_config.detect_archs", return_value=[]): rc, _, err = self.run_main( {"xml": junit(record("Keep"))}, - "--config", str(path), "--os", "linux", "--arch", "auto", + "--config", + str(path), + "--os", + "linux", + "--arch", + "auto", ) self.assertEqual(rc, 1) self.assertIn("could not detect a GPU architecture", err) @@ -313,11 +464,15 @@ def test_auto_arch_with_no_device_fails_closed(self): def test_config_flags_without_a_config_are_rejected(self): with contextlib.redirect_stderr(io.StringIO()): with self.assertRaises(SystemExit): - run_catch2.parse_args(["--xml", "r.xml", "--arch", "none", "--", "./bin"]) + run_catch2.parse_args( + ["--xml", "r.xml", "--arch", "none", "--", "./bin"] + ) @unittest.skipUnless(os.name == "posix", "process groups are POSIX only") def test_timeout_kills_the_suite_and_reports(self): - rc, out, _ = self.run_main({"sleep": 60, "banner": "LAST CASE"}, "--timeout", "1") + rc, out, _ = self.run_main( + {"sleep": 60, "banner": "LAST CASE"}, "--timeout", "1" + ) self.assertEqual(rc, 1) self.assertIn("timed out after 1 s", out) diff --git a/projects/hrr/tests/scripts/test_summarize_junit.py b/projects/hrr/tests/scripts/test_summarize_junit.py index 07bab8206249..2df153c39158 100644 --- a/projects/hrr/tests/scripts/test_summarize_junit.py +++ b/projects/hrr/tests/scripts/test_summarize_junit.py @@ -130,18 +130,31 @@ def test_render_marks_missing_unreadable_and_extra(self): self.write("integration-gfx90a.xml", " int: + print("+ " + " ".join(cmd), flush=True) + return subprocess.run(list(cmd)).returncode + + +def error(message: str) -> None: + print(f"::error::{message}", flush=True) + + +def base_header(root: Path, base_sha: str) -> tuple[str | None, str | None]: + """hrr_api_args.h at base_sha. + + Returns (content, None) on success, (None, None) when the path genuinely + does not exist at base_sha -- the only case with nothing to compare -- + and (None, reason) for any other git failure. The base commit must be + verified to resolve first: in a plain checkout an unresolvable rev and + a missing path print the same 'does not exist in' text, and the + codegen job's ``blob:none`` clone can also report 'not our ref' for a + commit it does not have -- an unreadable base must fail the check + instead of passing an incompatible API-ID renumbering. + """ + verify = subprocess.run( + ["git", "-C", str(root), "rev-parse", "--verify", "--quiet", + f"{base_sha}^{{commit}}"], + capture_output=True, + text=True, + ) + if verify.returncode != 0: + detail = (verify.stderr or verify.stdout).strip() + return None, ( + f"cannot read {API_ARGS_HEADER} at {base_sha}: the base commit does not " + f"resolve ({detail or f'git rev-parse exited {verify.returncode}'})" + ) + proc = subprocess.run( + ["git", "-C", str(root), "show", f"{base_sha}:{API_ARGS_HEADER}"], + capture_output=True, + text=True, + ) + if proc.returncode == 0: + return proc.stdout, None + detail = (proc.stderr or proc.stdout).strip() + if re.search(r"does not exist in|exists on disk, but not in", detail): + return None, None + if re.search(r"not our ref|error:Server does not allow request", detail): + return None, ( + f"cannot read {API_ARGS_HEADER} at {base_sha}: " + f"the blob could not be fetched ({detail.splitlines()[-1]})" + ) + return None, ( + f"cannot read {API_ARGS_HEADER} at {base_sha}: " + f"{detail or f'git show exited {proc.returncode}'}" + ) + + +def resolve_base(root: Path, base_sha: str) -> str: + """The commit to compare API IDs against: base_sha, else the parent commit.""" + if base_sha and set(base_sha) != {"0"}: + return base_sha + return subprocess.run( + ["git", "-C", str(root), "rev-parse", "HEAD^"], + check=True, capture_output=True, text=True, + ).stdout.strip() + + +def check_api_compat( + root: Path, + out: Path, + base_sha: str, + runner: Runner = run, + read_base: Callable[[Path, str], tuple[str | None, str | None]] = base_header, +) -> bool: + current = root / API_ARGS_HEADER + text, failure = read_base(root, base_sha) + if failure: + error(failure) + error("the API-ID compatibility with the base could not be verified; do not pass it silently") + return False + if text is None: + print(f"{API_ARGS_HEADER} does not exist at {base_sha}; nothing to compare") + return True + base = out / "base_hrr_api_args.h" + base.write_text(text, encoding="utf-8") + rc = runner( + [ + sys.executable, str(root / COMPAT_CHECKER), + "--base-header", str(base), + "--current-header", str(current), + ] + ) + if rc != 0: + error(f"HRR API IDs are not compatible with {base_sha}; see {COMPAT_CHECKER}") + return False + return True + + +def check_workloads(root: Path, runner: Runner = run) -> bool: + rc = runner( + [sys.executable, str(root / MATRIX_TOOLS / "check_matrix.py"), "--check-workloads"] + ) + if rc != 0: + error(f"check_matrix.py --check-workloads failed (exit {rc})") + return rc == 0 + + +def main(argv: list[str] | None = None) -> int: + parser = argparse.ArgumentParser(description=__doc__.splitlines()[0]) + parser.add_argument("check", choices=("api-compat", "workloads", "all")) + parser.add_argument("--root", type=Path, default=Path.cwd()) + parser.add_argument("--out", type=Path, default=Path("build/hrr-codegen")) + parser.add_argument("--base-sha", default="") + args = parser.parse_args(argv) + + root = args.root.resolve() + out = args.out.resolve() + out.mkdir(parents=True, exist_ok=True) + + ok = True + if args.check in ("api-compat", "all"): + ok &= check_api_compat(root, out, resolve_base(root, args.base_sha)) + if args.check in ("workloads", "all"): + ok &= check_workloads(root) + return 0 if ok else 1 + + +if __name__ == "__main__": + raise SystemExit(main()) diff --git a/projects/hrr/tools/tests/test_generated_files_current.py b/projects/hrr/tools/tests/test_generated_files_current.py index d386874a0737..202aa873838a 100644 --- a/projects/hrr/tools/tests/test_generated_files_current.py +++ b/projects/hrr/tools/tests/test_generated_files_current.py @@ -11,7 +11,10 @@ API matrix and the dispatch tables disagree about what an API does at replay. This regenerates each into a scratch directory and compares it with the -checked-in copy. Nothing is written into the tree. The matrix half needs PyYAML, +checked-in copy, with the generator's --check-hrr-coverage on: every API outside +MANUAL_CAPTURE_APIS must have a parseable dispatch slot and every manual one its +hand-written shim in hip_capture.cpp. The CLR build runs that check only with +CLR_HRR_CODEGEN_CHECK, which is off by default. Nothing is written into the tree. The matrix half needs PyYAML, as the matrix tools do, and is skipped without it. Run: python3 -m unittest discover -s projects/hrr/tools/tests @@ -58,7 +61,7 @@ class GeneratedFilesCurrent(unittest.TestCase): def test_generator_outputs_match_the_checked_in_files(self): with tempfile.TemporaryDirectory() as tmp: out = Path(tmp) - _run([TOOLS / "gen_hrr_api_args.py", "--silent", + _run([TOOLS / "gen_hrr_api_args.py", "--silent", "--check-hrr-coverage", "--output-header", out / "hrr_api_args.h", "--output-capture", out / "capture.cpp", "--output-playback", out / "playback.cpp"]) @@ -73,6 +76,19 @@ def test_generator_outputs_match_the_checked_in_files(self): ) if d] self.assertEqual(stale, [], "\n\n".join(stale)) + def test_coverage_check_rejects_a_missing_manual_capture_shim(self): + # The check above only means something if --check-hrr-coverage can fail. + with tempfile.TemporaryDirectory() as tmp: + out = Path(tmp) + empty = out / "hip_capture.cpp" + empty.write_text("") + with self.assertRaises(RuntimeError): + _run([TOOLS / "gen_hrr_api_args.py", "--silent", "--check-hrr-coverage", + "--manual-capture-source", empty, + "--output-header", out / "hrr_api_args.h", + "--output-capture", out / "capture.cpp", + "--output-playback", out / "playback.cpp"]) + @unittest.skipUnless(HAVE_YAML, "PyYAML is required for the matrix tools") def test_api_matrix_header_matches_the_derived_manifest(self): with tempfile.TemporaryDirectory() as tmp: