diff --git a/.github/actions/setup_test_environment/action.yml b/.github/actions/setup_test_environment/action.yml index 8fd0b182164..c853f65663a 100644 --- a/.github/actions/setup_test_environment/action.yml +++ b/.github/actions/setup_test_environment/action.yml @@ -21,6 +21,13 @@ inputs: ARTIFACT_RUN_ID: description: (string) Run ID to download artifacts from (`env.ARTIFACT_RUN_ID` or a value like `19716407405`) required: true + RUN_GITHUB_REPO: + description: >- + (string) Repository owning ARTIFACT_RUN_ID, in 'owner/repo' form. Only + needed when the run belongs to another repository, such as a rockrel + release nightly; the bucket lookup queries the run through the API and + fails if it looks in the wrong place. Defaults to this repository. + default: "" FETCH_ARTIFACT_ARGS: description: (string) Extra arguments to install_rocm_from_artifacts.py (e.g. `--blas --tests`) RELEASE_TYPE: @@ -62,6 +69,7 @@ runs: ARTIFACT_GROUP: ${{ inputs.ARTIFACT_GROUP }} AMDGPU_TARGETS: ${{ inputs.AMDGPU_TARGETS }} ARTIFACT_RUN_ID: ${{ inputs.ARTIFACT_RUN_ID }} + RUN_GITHUB_REPO: ${{ inputs.RUN_GITHUB_REPO }} FETCH_ARTIFACT_ARGS: ${{ inputs.FETCH_ARTIFACT_ARGS }} RELEASE_TYPE: ${{ inputs.RELEASE_TYPE }} GITHUB_TOKEN: ${{ github.token }} @@ -70,6 +78,7 @@ runs: run: | python ./build_tools/install_rocm_from_artifacts.py \ --run-id=${ARTIFACT_RUN_ID} \ + ${RUN_GITHUB_REPO:+--run-github-repo=${RUN_GITHUB_REPO}} \ --artifact-group=${ARTIFACT_GROUP} \ --amdgpu-targets=${AMDGPU_TARGETS} \ --output-dir=${OUTPUT_ARTIFACTS_DIR} \ diff --git a/.github/workflows/multi_arch_build_portable_linux_artifacts.yml b/.github/workflows/multi_arch_build_portable_linux_artifacts.yml index 3f8dac649aa..ccc288c72d5 100644 --- a/.github/workflows/multi_arch_build_portable_linux_artifacts.yml +++ b/.github/workflows/multi_arch_build_portable_linux_artifacts.yml @@ -49,6 +49,13 @@ on: required: false default: "" description: "CMake preset name for this build variant (optional)" + extra_cmake_options: + type: string + required: false + default: "" + description: >- + Extra -D options appended to the configure command, after the + stage-derived arguments. Used by variant builds such as coverage. release_type: description: 'Release type; controls artifact bucket and IAM role' type: string @@ -73,6 +80,14 @@ on: description: "Comma-separated artifact names to build (granular reuse). Empty = build all artifacts in stage." type: string default: "" + artifact_run_id: + description: >- + Run id to fetch inbound artifacts from and push this stage's artifacts + and logs under. Empty uses github.run_id. A variant build sharing a + workflow run with a regular one (coverage) passes its own id, since the + two produce artifacts with the same names. + type: string + default: "" quartz_tracking_id: type: string default: "" @@ -100,6 +115,7 @@ jobs: ROCM_PACKAGE_VERSION: ${{ inputs.rocm_package_version }} RELEASE_TYPE: ${{ inputs.release_type }} REBUILD_ARTIFACTS: ${{ inputs.rebuild_artifacts }} + ARTIFACT_RUN_ID: ${{ inputs.artifact_run_id != '' && inputs.artifact_run_id || github.run_id }} STAGE_NAME: ${{ inputs.stage_name }} CCACHE_CONFIGPATH: ${{ github.workspace }}/.ccache/ccache.conf TEATIME_FORCE_INTERACTIVE: 0 @@ -162,7 +178,7 @@ jobs: - name: Fetch inbound artifacts run: | python build_tools/artifact_manager.py fetch \ - --run-id=${{ github.run_id }} \ + --run-id="${ARTIFACT_RUN_ID}" \ --stage="${STAGE_NAME}" \ --amdgpu-families="${AMDGPU_FAMILIES}" \ --output-dir="${BUILD_DIR}" \ @@ -212,7 +228,8 @@ jobs: ${{ inputs.build_variant_cmake_preset && format('--preset={0}', inputs.build_variant_cmake_preset) }} \ ${{ inputs.external_repo_config != '' && format('-DTHEROCK_{0}_SOURCE_DIR={1}', fromJSON(inputs.external_repo_config).source_package, fromJSON(inputs.external_repo_config).checkout_path) || '' }} \ ${{ inputs.external_repo_config != '' && fromJSON(inputs.external_repo_config).extra_cmake_options || '' }} \ - ${{ steps.stage_config.outputs.cmake_args }} + ${{ steps.stage_config.outputs.cmake_args }} \ + ${{ inputs.extra_cmake_options }} - name: Build stage run: | @@ -265,7 +282,7 @@ jobs: - name: Push stage artifacts run: | python build_tools/artifact_manager.py push \ - --run-id ${{ github.run_id }} \ + --run-id "${ARTIFACT_RUN_ID}" \ --stage="${STAGE_NAME}" \ --build-dir="${BUILD_DIR}" @@ -273,7 +290,7 @@ jobs: if: ${{ always() }} run: | python build_tools/github_actions/post_stage_upload.py \ - --run-id=${{ github.run_id }} \ + --run-id="${ARTIFACT_RUN_ID}" \ --stage="${STAGE_NAME}" \ --build-dir="${BUILD_DIR}" \ --amdgpu-family="${AMDGPU_FAMILIES}" diff --git a/.github/workflows/multi_arch_ci_coverage_linux.yml b/.github/workflows/multi_arch_ci_coverage_linux.yml new file mode 100644 index 00000000000..8b7790406c4 --- /dev/null +++ b/.github/workflows/multi_arch_ci_coverage_linux.yml @@ -0,0 +1,338 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +# Multi-Arch CI Coverage - Linux +# +# Self-contained Linux coverage entry point. Computes the coverage matrix, +# builds one instrumented Linux stack for every coverage-enabled project, then +# fans out to multi_arch_ci_coverage_report.yml once per project. +# +# The rocm-libraries nightly invokes this via a cross-repo reusable `uses:` call. +# A future Windows mirror copies this file, swapping the build reusable and +# runners; multi_arch_ci_coverage_report.yml is reused unchanged. +# +# Currently targets gfx94X-dcgpu; change-based selection and multi-arch come later. + +name: Multi-Arch CI Coverage - Linux + +on: + workflow_dispatch: + inputs: + baseline_run_id: + type: string + description: "Regular nightly run id supplying the non-instrumented dependencies. Empty tests this run's own instrumented stack end to end." + default: "" + baseline_release_type: + type: string + description: "Release channel baseline_run_id published to; artifacts are bucketed by channel." + default: "nightly" + baseline_repository: + type: string + description: "Repository owning baseline_run_id (owner/repo). Release nightlies come from ROCm/rockrel; empty means this repository." + default: "" + projects_to_test: + type: string + description: "Comma-separated coverage projects (e.g. hiprand), case-insensitive. Alias: rocm_libraries_all. Empty = every buildable project. rccl/rocshmem require a comm-libs build job (not present), so rocm_systems_all and all are rejected." + default: "" + amdgpu_families: + type: string + description: "Comma-separated GPU families to collect coverage on." + default: "gfx94X-dcgpu" + coverage_config_source: + type: string + description: "Repository and ref holding the per-project coverage metadata." + default: "ROCm/rocm-libraries@main" + test_type: + type: string + description: "Test type to run against the instrumented build." + default: "standard" + external_repo_config: + type: string + description: "JSON from detect_external_repo_config.py naming the external checkout to build (e.g. a rocm-libraries commit). Empty builds TheRock's pinned submodules." + default: "" + quartz_tracking_id: + type: string + description: "Leave empty. Reserved for release tracking." + default: "" + repository: + type: string + description: "Repository holding this workflow's scripts. Empty defaults to github.repository; cross-repo callers pass ROCm/TheRock." + default: "" + ref: + type: string + description: "Ref/SHA of this workflow's scripts. Cross-repo callers pass the ref they pinned in uses:." + default: "" + + # Required so cross-repo callers (rocm-libraries nightly) can invoke via `uses:`. + workflow_call: + inputs: + baseline_run_id: + type: string + description: "Regular nightly run id supplying the non-instrumented dependencies. Empty tests this run's own instrumented stack end to end." + default: "" + baseline_release_type: + type: string + description: "Release channel baseline_run_id published to; artifacts are bucketed by channel." + default: "nightly" + baseline_repository: + type: string + description: "Repository owning baseline_run_id (owner/repo). Release nightlies come from ROCm/rockrel; empty means this repository." + default: "" + projects_to_test: + type: string + description: "Comma-separated coverage projects (e.g. hiprand), case-insensitive. Alias: rocm_libraries_all. Empty = every buildable project. rccl/rocshmem require a comm-libs build job (not present), so rocm_systems_all and all are rejected." + default: "" + amdgpu_families: + type: string + description: "Comma-separated GPU families to collect coverage on." + default: "gfx94X-dcgpu" + coverage_config_source: + type: string + description: "Repository and ref holding the per-project coverage metadata." + default: "ROCm/rocm-libraries@main" + test_type: + type: string + description: "Test type to run against the instrumented build." + default: "standard" + external_repo_config: + type: string + description: "JSON from detect_external_repo_config.py naming the external checkout to build. The rocm-libraries nightly passes its own, so coverage measures the commit under test; empty builds TheRock's pinned submodules." + default: "" + quartz_tracking_id: + type: string + description: "Leave empty. Reserved for release tracking." + default: "" + repository: + type: string + description: "Repository holding this workflow's scripts. Empty defaults to github.repository; cross-repo callers pass ROCm/TheRock." + default: "" + ref: + type: string + description: "Ref/SHA of this workflow's scripts. Cross-repo callers pass the ref they pinned in uses:." + default: "" + +permissions: + contents: read + +run-name: >- + Multi-Arch CI Coverage - Linux + | baseline: ${{ inputs.baseline_run_id || 'none (self-contained)' }} + | ${{ inputs.amdgpu_families || 'gfx94X-dcgpu' }} + +concurrency: + group: ${{ github.workflow }}-${{ github.run_id }} + cancel-in-progress: false + +jobs: + setup_coverage_matrix: + name: Setup Coverage Matrix + runs-on: ubuntu-24.04 + outputs: + coverage_matrix: ${{ steps.matrix.outputs.coverage_matrix }} + dist_amdgpu_families: ${{ steps.matrix.outputs.dist_amdgpu_families }} + families_matrix_json: ${{ steps.matrix.outputs.families_matrix_json }} + coverage_cmake_options: ${{ steps.matrix.outputs.coverage_cmake_options }} + needs_math_libs: ${{ steps.matrix.outputs.needs_math_libs }} + rocm_package_version: ${{ steps.rocm_package_version.outputs.rocm_package_version }} + # S3 namespace for the instrumented stack. A caller that builds a regular + # stack in the same run (the rocm-libraries nightly does, and uses it as + # the baseline) publishes artifacts with the same names under + # github.run_id, so sharing that namespace overwrites them. + artifact_run_id: ${{ github.run_id }}-coverage + steps: + - name: Checking out repository + id: checkout + uses: actions/checkout@3d3c42e5aac5ba805825da76410c181273ba90b1 # v7.0.1 + timeout-minutes: 15 + with: + # Cross-repo callers pass ROCm/TheRock + ref; same-repo defaults to github.repository. + repository: ${{ inputs.repository || github.repository }} + ref: ${{ inputs.ref }} + + - name: Setting up Python + uses: actions/setup-python@5fda3b95a4ea91299a34e894583c3862153e4b97 # v7.0.0 + with: + python-version: 3.12 + + - name: Generate coverage matrix + id: matrix + env: + PROJECTS_TO_TEST: ${{ inputs.projects_to_test || '' }} + AMDGPU_FAMILIES: ${{ inputs.amdgpu_families || 'gfx94X-dcgpu' }} + COVERAGE_CONFIG_SOURCE: ${{ inputs.coverage_config_source || 'ROCm/rocm-libraries@main' }} + run: python ./build_tools/github_actions/configure_coverage_ci.py + + - name: Compute package version + id: rocm_package_version + run: | + python ./build_tools/compute_rocm_package_version.py \ + --release-type=ci \ + --override-git-sha=${{ steps.checkout.outputs.commit }} + + # Built first so later stages have a compiler-runtime to link against and + # fetch as an inbound artifact. + build_compiler_runtime: + name: Build Compiler Runtime + needs: setup_coverage_matrix + if: ${{ needs.setup_coverage_matrix.outputs.coverage_matrix != '[]' }} + uses: ./.github/workflows/multi_arch_build_portable_linux_artifacts.yml + secrets: inherit + with: + stage_name: compiler-runtime + stage_display_name: "Stage - Coverage Compiler Runtime" + timeout_minutes: 480 # 8 hours (compiler is big) + dist_amdgpu_families: ${{ needs.setup_coverage_matrix.outputs.dist_amdgpu_families }} + rocm_package_version: ${{ needs.setup_coverage_matrix.outputs.rocm_package_version }} + release_type: ci + artifact_run_id: ${{ needs.setup_coverage_matrix.outputs.artifact_run_id }} + external_repo_config: ${{ inputs.external_repo_config }} + quartz_tracking_id: ${{ inputs.quartz_tracking_id }} + # TheRock scripts for the instrumented build; cross-repo callers pass ROCm/TheRock + ref. + repository: ${{ inputs.repository }} + ref: ${{ inputs.ref }} + permissions: + contents: read + id-token: write + + # Build the whole stage: test jobs need the full dependency set, and + # instrumenting as a unit keeps the per-project build cost flat. + # coverage_cmake_options is the THEROCK_COVERAGE_*_ALL group option for a whole + # group, else individual project flags; unselected projects build normally. + build_instrumented_math_libs: + name: Build Instrumented Math Libs + needs: [setup_coverage_matrix, build_compiler_runtime] + if: >- + ${{ + !failure() && !cancelled() && + needs.setup_coverage_matrix.outputs.needs_math_libs == 'true' + }} + strategy: + fail-fast: false + matrix: + family_info: ${{ fromJSON(needs.setup_coverage_matrix.outputs.families_matrix_json) }} + uses: ./.github/workflows/multi_arch_build_portable_linux_artifacts.yml + secrets: inherit + with: + stage_name: math-libs + stage_display_name: "Stage - Coverage Math Libs (${{ matrix.family_info.amdgpu_family }})" + timeout_minutes: 480 # 8 hours; instrumented builds are slower than normal + amdgpu_family: ${{ matrix.family_info.amdgpu_family }} + dist_amdgpu_families: ${{ needs.setup_coverage_matrix.outputs.dist_amdgpu_families }} + extra_cmake_options: ${{ needs.setup_coverage_matrix.outputs.coverage_cmake_options }} + rocm_package_version: ${{ needs.setup_coverage_matrix.outputs.rocm_package_version }} + release_type: ci + artifact_run_id: ${{ needs.setup_coverage_matrix.outputs.artifact_run_id }} + external_repo_config: ${{ inputs.external_repo_config }} + quartz_tracking_id: ${{ inputs.quartz_tracking_id }} + # TheRock scripts for the instrumented build; cross-repo callers pass ROCm/TheRock + ref. + repository: ${{ inputs.repository }} + ref: ${{ inputs.ref }} + permissions: + contents: read + id-token: write + + # A caller that builds the baseline in this same run (the rocm-libraries + # nightly) can start this workflow alongside that build. The tests install + # the baseline from S3, so hold them until that build's math-libs stage (which + # needs compiler-runtime) has succeeded, or its compiler-runtime stage when no + # math-libs project is selected. + wait_for_baseline: + name: Wait for Baseline + needs: setup_coverage_matrix + if: >- + ${{ + inputs.baseline_run_id == github.run_id && + needs.setup_coverage_matrix.outputs.coverage_matrix != '[]' + }} + runs-on: ubuntu-24.04 + # Above the script's own 300-minute limit, so it reports the timeout. + timeout-minutes: 330 + steps: + - name: Checking out repository + uses: actions/checkout@3d3c42e5aac5ba805825da76410c181273ba90b1 # v7.0.1 + timeout-minutes: 15 + with: + # Cross-repo callers pass ROCm/TheRock + ref; same-repo defaults to github.repository. + repository: ${{ inputs.repository || github.repository }} + ref: ${{ inputs.ref }} + - name: Wait for the baseline stage + env: + GITHUB_TOKEN: ${{ github.token }} + BASELINE_STAGE: ${{ needs.setup_coverage_matrix.outputs.needs_math_libs == 'true' && 'math-libs' || 'compiler-runtime' }} + AMDGPU_FAMILIES: ${{ needs.setup_coverage_matrix.outputs.dist_amdgpu_families }} + run: | + python build_tools/github_actions/wait_for_baseline_stage.py \ + --stage="${BASELINE_STAGE}" \ + --amdgpu-families="${AMDGPU_FAMILIES}" + + # One test + report job per project. Each installs baseline_run_id and replaces + # only its own project from this run's instrumented stack, so dependencies stay + # non-instrumented even though the build instrumented them all at once. + coverage_report: + name: Coverage::${{ matrix.coverage.project_name }} (${{ matrix.coverage.amdgpu_families }}) + # Name compiler-runtime explicitly, not via math-libs: a selection with no + # math-libs build still tests against binaries that stage produced. + needs: + [ + setup_coverage_matrix, + build_compiler_runtime, + build_instrumented_math_libs, + wait_for_baseline, + ] + # Also guard the empty matrix: fromJSON('') would hard-error. + if: >- + ${{ + !failure() && !cancelled() && + needs.setup_coverage_matrix.outputs.coverage_matrix != '[]' + }} + strategy: + fail-fast: false + matrix: + coverage: ${{ fromJSON(needs.setup_coverage_matrix.outputs.coverage_matrix) }} + uses: ./.github/workflows/multi_arch_ci_coverage_report.yml + secrets: inherit + with: + project_name: ${{ matrix.coverage.project_name }} + test_component: ${{ matrix.coverage.test_component }} + object_globs: ${{ matrix.coverage.object_globs }} + fetch_artifact_args: ${{ matrix.coverage.fetch_artifact_args }} + build_stage: ${{ matrix.coverage.build_stage }} + codecov_flag: ${{ matrix.coverage.codecov_flag }} + amdgpu_families: ${{ matrix.coverage.amdgpu_families }} + baseline_run_id: ${{ inputs.baseline_run_id }} + baseline_release_type: ${{ inputs.baseline_release_type }} + baseline_repository: ${{ inputs.baseline_repository }} + artifact_run_id: ${{ needs.setup_coverage_matrix.outputs.artifact_run_id }} + test_type: ${{ inputs.test_type || 'standard' }} + release_type: ci + # Forwarded to L2's checkouts and the component jobs. + repository: ${{ inputs.repository }} + ref: ${{ inputs.ref }} + permissions: + contents: read + id-token: write + + ci_summary: + name: Coverage Summary + if: always() + needs: + - setup_coverage_matrix + - build_compiler_runtime + - build_instrumented_math_libs + - wait_for_baseline + - coverage_report + runs-on: ubuntu-24.04 + steps: + - name: Checking out repository + uses: actions/checkout@3d3c42e5aac5ba805825da76410c181273ba90b1 # v7.0.1 + timeout-minutes: 15 + with: + # Cross-repo callers pass ROCm/TheRock + ref; same-repo defaults to github.repository. + repository: ${{ inputs.repository || github.repository }} + ref: ${{ inputs.ref }} + - name: Evaluate workflow results + env: + GITHUB_TOKEN: ${{ github.token }} + run: | + python build_tools/github_actions/workflow_summary.py \ + --needs-json '${{ toJSON(needs) }}' diff --git a/.github/workflows/multi_arch_ci_coverage_report.yml b/.github/workflows/multi_arch_ci_coverage_report.yml new file mode 100644 index 00000000000..893f4fce9f5 --- /dev/null +++ b/.github/workflows/multi_arch_ci_coverage_report.yml @@ -0,0 +1,321 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +# Multi-Arch CI Coverage - Report +# +# Runs the test and report half of the coverage pipeline for a single project on +# a single GPU family: +# +# test (collect profraw) -> merge -> report +# +# The instrumented stack is built by the caller +# (multi_arch_ci_coverage_linux.yml); this workflow only consumes it. +# +# baseline_run_id drives the hybrid install: the tree comes from a regular +# nightly and only the project under test is replaced from the instrumented run, +# keeping instrumented dependencies out of the report. Without it, the +# instrumented artifacts are tested as built, measuring dependencies alongside it. + +name: Multi-Arch CI Coverage - Report + +on: + workflow_call: + inputs: + project_name: + type: string + description: "Coverage project key (e.g. hiprand)" + test_component: + type: string + description: "Test matrix key for the project's tests" + object_globs: + type: string + description: "Comma-separated globs matching the instrumented binaries" + fetch_artifact_args: + type: string + default: "" + description: "Extra install_rocm_from_artifacts.py args for the report job" + build_stage: + type: string + default: "" + description: >- + Build stage owning the project. Its sources are fetched so the HTML + report can annotate them; lcov and text outputs don't need them. + Empty skips the fetch and the HTML. + baseline_run_id: + type: string + default: "" + description: >- + Regular nightly run supplying non-instrumented dependencies. Test jobs + install it and overlay only the project under test. Empty tests the + instrumented artifacts as built. + baseline_repository: + type: string + default: "" + description: >- + Repository owning baseline_run_id, in 'owner/repo' form. Needed when + the baseline is a release nightly, which ROCm/rockrel publishes. + baseline_release_type: + type: string + default: nightly + artifact_run_id: + type: string + default: "" + description: >- + Run id the instrumented stack was pushed under, which the overlay and + the report read from. Empty uses github.run_id. + codecov_flag: + type: string + default: "" + description: "Codecov flag to file the report under" + amdgpu_families: + type: string + description: "GPU family to test coverage on" + amdgpu_targets: + type: string + default: "" + description: "Specific GPU targets to test, passed through to the installer. Empty derives them from amdgpu_families." + test_type: + type: string + default: standard + release_type: + type: string + default: ci + repository: + description: "Repository to checkout. Defaults to github.repository." + type: string + default: "" + ref: + description: "Branch, tag, or SHA to checkout. Defaults to the triggering ref." + type: string + default: "" + +permissions: + contents: read + +jobs: + configure_test_matrix: + name: Configure Coverage Tests (${{ inputs.project_name }}) + runs-on: ubuntu-24.04 + outputs: + components: ${{ steps.configure.outputs.components }} + platform: ${{ steps.configure.outputs.platform }} + steps: + - name: Checking out repository + uses: actions/checkout@3d3c42e5aac5ba805825da76410c181273ba90b1 # v7.0.1 + timeout-minutes: 15 + with: + repository: ${{ inputs.repository || github.repository }} + ref: ${{ inputs.ref }} + + - name: Checkout CI config + uses: actions/checkout@3d3c42e5aac5ba805825da76410c181273ba90b1 # v7.0.1 + timeout-minutes: 15 + with: + repository: ROCm/therock-ci-config + ref: main + path: ci-config + continue-on-error: true + + - name: Setting up Python + uses: actions/setup-python@5fda3b95a4ea91299a34e894583c3862153e4b97 # v7.0.0 + with: + python-version: 3.12 + + - name: Configuring coverage test options + id: configure + env: + ARTIFACT_GROUP: multi-arch-coverage + AMDGPU_FAMILIES: ${{ inputs.amdgpu_families }} + TEST_TYPE: ${{ inputs.test_type }} + TEST_LABELS: "[]" + # One project per report, so only its tests run here; other projects + # get their own job. + PROJECTS_TO_TEST: ${{ inputs.test_component }} + CI_CONFIG_PATH: ci-config + BUILD_VARIANT: coverage + run: | + python ./build_tools/github_actions/fetch_test_configurations.py \ + --platform=linux + + test_coverage: + name: Test Coverage (${{ inputs.project_name }}) + needs: configure_test_matrix + if: >- + ${{ + !failure() && !cancelled() && + needs.configure_test_matrix.outputs.components != '' && + needs.configure_test_matrix.outputs.components != '[]' + }} + strategy: + fail-fast: false + matrix: + components: ${{ fromJSON(needs.configure_test_matrix.outputs.components || '[]') }} + uses: ./.github/workflows/test_code_coverage_component.yml + secrets: inherit + with: + # Installer takes the generic tree from baseline_run_id and replaces only + # this project's libs from this run's instrumented artifacts, deriving the + # artifact name and folder from the component job_name. + generic_run_id: ${{ inputs.baseline_run_id }} + baseline_repository: ${{ inputs.baseline_repository }} + code_coverage_run_id: ${{ inputs.artifact_run_id }} + artifact_group: multi-arch-coverage + amdgpu_families: ${{ inputs.amdgpu_families }} + amdgpu_targets: ${{ inputs.amdgpu_targets }} + test_runs_on: ${{ matrix.components.test_runner }} + platform: ${{ needs.configure_test_matrix.outputs.platform }} + component: ${{ toJSON(matrix.components) }} + release_type: ${{ inputs.baseline_release_type }} + repository: ${{ inputs.repository }} + ref: ${{ inputs.ref }} + build_variant: coverage + coverage_enabled: true + coverage_artifact_prefix: coverage-profraw-${{ inputs.project_name }}-${{ inputs.amdgpu_families }} + + aggregate_coverage: + name: Coverage Report (${{ inputs.project_name }}) + needs: [configure_test_matrix, test_coverage] + # Runs even if some shards failed: a partial report beats none, and a + # failing test still exercises code. + if: >- + ${{ + !cancelled() && + needs.configure_test_matrix.result == 'success' && + needs.configure_test_matrix.outputs.components != '[]' + }} + runs-on: ubuntu-24.04 + permissions: + contents: read + id-token: write + env: + OUTPUT_ARTIFACTS_DIR: ${{ github.workspace }}/build + PROFRAW_DIR: ${{ github.workspace }}/coverage-profraw + REPORT_DIR: ${{ github.workspace }}/coverage-report + steps: + - name: Checking out repository + uses: actions/checkout@3d3c42e5aac5ba805825da76410c181273ba90b1 # v7.0.1 + timeout-minutes: 15 + with: + repository: ${{ inputs.repository || github.repository }} + ref: ${{ inputs.ref }} + + - name: Setting up Python + uses: actions/setup-python@5fda3b95a4ea91299a34e894583c3862153e4b97 # v7.0.0 + with: + python-version: 3.12 + + - name: Download profraw files from all shards + uses: actions/download-artifact@37930b1c2abaa49bbe596cd826c3c89aef350131 # v7.0.0 + with: + pattern: coverage-profraw-${{ inputs.project_name }}-${{ inputs.amdgpu_families }}-* + path: ${{ env.PROFRAW_DIR }} + continue-on-error: true + + - name: Install python deps + run: pip install -r requirements.txt + + - name: Download instrumented artifacts + env: + GITHUB_TOKEN: ${{ github.token }} + RELEASE_TYPE: ${{ inputs.release_type }} + FETCH_ARTIFACT_ARGS: ${{ inputs.fetch_artifact_args }} + run: | + python ./build_tools/install_rocm_from_artifacts.py \ + --run-id=${{ inputs.artifact_run_id || github.run_id }} \ + --artifact-group=${{ inputs.amdgpu_families }} \ + --output-dir=${OUTPUT_ARTIFACTS_DIR} \ + ${FETCH_ARTIFACT_ARGS} + + # Only HTML rendering reads source, so fetch after the artifacts are down: + # a failure here (e.g. out of disk) costs only the annotated view, leaving + # lcov and the text table. + - name: Fetch sources for the annotated report + if: ${{ inputs.build_stage != '' }} + id: fetch_sources + timeout-minutes: 30 + continue-on-error: true + env: + GH_TOKEN: ${{ github.token }} + STAGE_NAME: ${{ inputs.build_stage }} + run: | + gh auth setup-git --hostname github.com + ./build_tools/fetch_sources.py --stage "${STAGE_NAME}" --jobs 12 --depth 1 + + - name: Merge profraw and generate reports + env: + # Sources are recorded at the build container's path, not where they + # sit here. The HTML and text renderings need to reach them on disk, + # so they map the build workspace to this checkout. Codecov instead + # matches the lcov against the repository tree, so that export strips + # the build-time repo root to emit repo-relative paths. The repo is a + # submodule one level below the superproject workspace, so its root is + # BUILD_WORKSPACE plus the repo name. + BUILD_WORKSPACE: /__w/${{ github.event.repository.name }}/${{ github.event.repository.name }} + BUILD_REPO_ROOT: /__w/${{ github.event.repository.name }}/${{ github.event.repository.name }}/${{ github.event.repository.name }} + HTML_ENABLED: ${{ steps.fetch_sources.outcome == 'success' }} + run: | + html_args=() + if [[ "${HTML_ENABLED}" == "true" ]]; then + html_args=( + --html-output="${REPORT_DIR}/html" + --project-title="${{ inputs.project_name }} (${{ inputs.amdgpu_families }})" + ) + else + echo "::warning::Sources unavailable; skipping the annotated HTML report." + fi + + # Strip through the submodule root so Codecov SF: paths are relative + # to rocm-libraries/ (matching its working-directory). path-equivalence + # maps the build-time submodule path to GITHUB_WORKSPACE for HTML. + BUILD_SUBMODULE_PATH="${BUILD_REPO_ROOT}/rocm-libraries" + python ./build_tools/github_actions/merge_coverage_report.py \ + --profraw-dir="${PROFRAW_DIR}" \ + --rocm-dir="${OUTPUT_ARTIFACTS_DIR}" \ + --object-globs="${{ inputs.object_globs }}" \ + --profdata-output="${REPORT_DIR}/coverage.profdata" \ + --lcov-output="${REPORT_DIR}/coverage.info" \ + --summary-output="${REPORT_DIR}/coverage_summary.txt" \ + --path-equivalence="${BUILD_SUBMODULE_PATH},${GITHUB_WORKSPACE}" \ + --lcov-strip-prefix="${BUILD_SUBMODULE_PATH}/" \ + --ignore-filename-regex=".*/build/.*" \ + "${html_args[@]}" + + # Put totals on the run summary page, readable without downloading. + - name: Publish the summary to the job page + if: ${{ always() }} + run: | + summary="${REPORT_DIR}/coverage_summary.txt" + [[ -f "${summary}" ]] || exit 0 + { + echo "## Coverage: ${{ inputs.project_name }} (${{ inputs.amdgpu_families }})" + echo '```' + tail -n 3 "${summary}" + echo '```' + } >> "${GITHUB_STEP_SUMMARY}" + + - name: Upload coverage report + if: ${{ always() }} + uses: actions/upload-artifact@043fb46d1a93c77aae656e7c1c64a875d1fc6a0a # v7.0.1 + with: + name: coverage-report-${{ inputs.project_name }}-${{ inputs.amdgpu_families }} + # The whole directory: coverage.info for Codecov, coverage_summary.txt + # for the numbers, and html/index.html for the annotated report. + path: | + ${{ env.REPORT_DIR }}/coverage.info + ${{ env.REPORT_DIR }}/coverage_summary.txt + ${{ env.REPORT_DIR }}/html + if-no-files-found: warn + + # Skip (not fail) when there's no Codecov token, so forks and first-time + # setups still get the lcov artifact above. + - name: Upload report to Codecov + uses: codecov/codecov-action@fb8b3582c8e4def4969c97caa2f19720cb33a72f # v7.0.0 + with: + token: ${{ secrets.CODECOV_TOKEN }} + working-directory: ${{ github.workspace }}/rocm-libraries + files: ${{ env.REPORT_DIR }}/coverage.info + flags: "${{ inputs.codecov_flag || inputs.project_name }} - Nightly" + name: ${{ inputs.project_name }}-${{ inputs.amdgpu_families }} + disable_search: true + recurse_submodules: true + fail_ci_if_error: false diff --git a/.github/workflows/test_code_coverage_component.yml b/.github/workflows/test_code_coverage_component.yml index 4b115526cc4..99e7ec58cba 100644 --- a/.github/workflows/test_code_coverage_component.yml +++ b/.github/workflows/test_code_coverage_component.yml @@ -9,6 +9,20 @@ on: generic_run_id: type: string default: "" + baseline_repository: + description: >- + Repository owning generic_run_id (owner/repo), passed as + --run-github-repo. Empty defaults to the current repository. The + instrumented replacement always resolves against this coverage run's + own repository. + type: string + default: "" + code_coverage_run_id: + description: >- + Run id the instrumented stack was pushed under, passed as + --code-coverage-run-id. Empty uses github.run_id. + type: string + default: "" artifact_group: type: string amdgpu_families: @@ -37,16 +51,14 @@ on: default: "release" coverage_enabled: description: >- - Collect llvm profraw files from the test run and upload them as a - workflow artifact. Only meaningful against a coverage instrumented - build. + Collect llvm profraw files from the test run and upload them. + Only meaningful against a coverage-instrumented build. type: boolean default: false coverage_artifact_prefix: description: >- - Prefix for the uploaded profraw artifact name. A coverage run tests - several projects in one workflow run, so the caller sets this to - something project-specific to keep each project profiles apart. + Prefix for the uploaded profraw artifact name. Set per-project so one + coverage run's projects keep their profiles apart. type: string default: "coverage-profraw" default_container_image: @@ -78,15 +90,14 @@ jobs: (fromJSON(inputs.component).container_image || inputs.default_container_image) || null }} - # Container options are configured in build_tools/github_actions/fetch_test_configurations.py - # and include base options (ipc, user, ulimit, seccomp), GPU options (devices, groups), - # and any job-specific options + # Container options come from fetch_test_configurations.py: base (ipc, + # user, ulimit, seccomp), GPU (devices, groups), and job-specific options. options: ${{ fromJSON(inputs.component).container_options }} strategy: fail-fast: false matrix: - # The shard array is based on "total_shards" from "fetch_test_configurations.py" - # The test executable will shard based on the array. (ex: [1, 2, 3, 4] = four test shards) + # Shard array from fetch_test_configurations.py's total_shards. + # The test executable shards on it (e.g. [1,2,3,4] = four shards). shard: ${{ fromJSON(inputs.component).shard_arr }} defaults: run: @@ -94,7 +105,7 @@ jobs: env: GENERIC_RUN_ID: ${{ inputs.generic_run_id }} TEST_COMPONENT: ${{ fromJSON(inputs.component).job_name }} - CODE_COVERAGE_RUN_ID: ${{ github.run_id }} + CODE_COVERAGE_RUN_ID: ${{ inputs.code_coverage_run_id || github.run_id }} OUTPUT_ARTIFACTS_DIR: "./build" THEROCK_BIN_DIR: "./build/bin" AMDGPU_FAMILIES: ${{ inputs.amdgpu_families }} @@ -131,17 +142,39 @@ jobs: repository: ${{ inputs.repository || github.repository }} ref: ${{ inputs.ref }} + - name: Setting up Python + uses: actions/setup-python@5fda3b95a4ea91299a34e894583c3862153e4b97 # v7.0.0 + with: + python-version: 3.12 + - name: Install TheRock Code Coverage Build timeout-minutes: 15 env: RELEASE_TYPE: ${{ inputs.release_type }} FETCH_ARTIFACT_ARGS: ${{ fromJSON(inputs.component).fetch_artifact_args }} + BASELINE_REPOSITORY: ${{ inputs.baseline_repository }} run: | pip3 install -r requirements-test.txt + # Baseline tree comes from generic_run_id. When empty (self-contained + # dispatch), fall back to this run's own instrumented artifacts; + # --replace then re-extracts them, leaving the whole stack instrumented. + generic_run_id="${GENERIC_RUN_ID}" + if [[ -z "${generic_run_id}" ]]; then + generic_run_id="${CODE_COVERAGE_RUN_ID}" + fi + # Pass --run-github-repo only when set, so empty defaults to the + # current repository. The generic baseline lives in a different repo + # than this coverage run (rockrel for the PR check, rocm-libraries for + # the nightly); the S3 bucket lookup 404s if that repo is not named. + run_github_repo_arg=() + if [[ -n "${BASELINE_REPOSITORY}" ]]; then + run_github_repo_arg=(--run-github-repo="${BASELINE_REPOSITORY}") + fi # FETCH_ARTIFACT_ARGS intentionally splits arguments supplied by fetch_test_configurations.py. python3 build_tools/install_rocm_code_coverage_build.py \ --code-coverage-run-id="${CODE_COVERAGE_RUN_ID}" \ - --run-id="${GENERIC_RUN_ID}" \ + --run-id="${generic_run_id}" \ + "${run_github_repo_arg[@]}" \ --artifact-group="${AMDGPU_FAMILIES}" \ --amdgpu-targets="${AMDGPU_TARGETS}" \ --tests \ @@ -166,7 +199,9 @@ jobs: python ./build_tools/print_driver_gpu_info.py - name: Setup Additional Requirements - if: ${{ fromJSON(inputs.component).additional_requirements_files != '' }} + # First-element check skips cleanly when the list is empty ([]) or the + # key is absent, without depending on how an empty list coerces to ''. + if: ${{ fromJSON(inputs.component).additional_requirements_files[0] != '' }} env: REQUIREMENTS_FILES: ${{ join(fromJSON(inputs.component).additional_requirements_files, ',') }} run: | @@ -182,8 +217,8 @@ jobs: echo "ASAN_SYMBOLIZER_PATH=${OUTPUT_ARTIFACTS_DIR}/llvm/bin/llvm-symbolizer" >> $GITHUB_ENV # Instrumented binaries write one profraw per process. %p (pid) and %m - # (binary signature) keep the files distinct so a shard that forks or - # loads several instrumented libraries does not overwrite itself. + # (binary signature) keep files distinct so a forking/multi-lib shard + # doesn't overwrite itself. - name: Prepare coverage profile directory if: ${{ inputs.coverage_enabled }} env: @@ -192,8 +227,8 @@ jobs: mkdir -p "${GITHUB_WORKSPACE}/coverage-report/profraw" echo "LLVM_PROFILE_FILE=${GITHUB_WORKSPACE}/coverage-report/profraw/${TEST_COMPONENT}-shard${SHARD_INDEX}-%p-%m.profraw" >> "${GITHUB_ENV}" - # Set thread limits from KUBE_CPU_REQUEST (Kubernetes-injected) to avoid oversubscription. - # Must be done at runtime since container env vars aren't in GitHub's env context. + # Limit threads from KUBE_CPU_REQUEST (Kubernetes-injected) to avoid + # oversubscription. Done at runtime; container env isn't in GitHub's context. - name: Set thread limits run: | if [[ -n "${KUBE_CPU_REQUEST:-}" ]]; then @@ -243,8 +278,23 @@ jobs: run: | # zizmor: ignore[template-injection] ${{ fromJSON(inputs.component).test_script }} - # Uploaded even when the test step failed: partial coverage from a failing - # shard is still useful, and the aggregation job tolerates missing shards. + # A device-side profile is named after its GPU target + # (gfx942:sramecc+:xnack-.0..profraw). upload-artifact rejects the + # colons and with them every other file in the upload, and llvm-cov cannot + # use the profile against host objects anyway. Its presence means kernels + # were instrumented, which the build is meant to prevent, so say so. + - name: Drop device-side profiles + if: ${{ always() && inputs.coverage_enabled }} + run: | + shopt -s nullglob + device_profiles=("${GITHUB_WORKSPACE}"/coverage-report/profraw/*:*.profraw) + if (( ${#device_profiles[@]} )); then + echo "::warning::Kernels were instrumented; dropping ${#device_profiles[@]} device-side profile(s): ${device_profiles[*]##*/}" + rm -f -- "${device_profiles[@]}" + fi + + # Upload even on test failure: partial coverage is useful and the + # aggregation job tolerates missing shards. - name: Upload coverage profraw files if: ${{ always() && inputs.coverage_enabled }} uses: actions/upload-artifact@cf430e030ddbb5b0abf93d22962f4752f3646cd9 # v7.0.2 diff --git a/CMakeLists.txt b/CMakeLists.txt index 1bab4b4ed8b..7d40ebcdd36 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -212,6 +212,59 @@ option(THEROCK_QUIET_INSTALL "Enable quiet install logging (install logs only go option(THEROCK_WINDOWS_LINK_CPU_AFFINITY "On Windows, pin each link step of sub-projects built with the amd-llvm/amd-hip toolchains to one random CPU (works around lld-link silently dropping inputs on AMD Turin hosts, see #9469)" OFF) option(THEROCK_USE_SAFE_DEPENDENCY_PROVIDER "Enable the safe dependency provider, performing strict package checks" ON) +################################################################################ +# Coverage Options +# Enable code-coverage instrumentation for a whole component group by setting the +# per-project _ENABLE_COVERAGE flags. therock_subproject.cmake turns +# each one into whichever option that subproject actually implements, using the +# THEROCK_COVERAGE_OPTION_ map generated below. Both the group lists +# and that map come from the coverage registry (COVERAGE_PROJECTS in +# build_tools/github_actions/configure_coverage_ci.py) at configure time, and +# cover only the projects that emit LLVM profile data. +################################################################################ + +option(THEROCK_COVERAGE_ROCM_LIBRARIES_ALL + "Enable coverage instrumentation for all rocm-libraries projects" OFF) +option(THEROCK_COVERAGE_ROCM_SYSTEMS_ALL + "Enable coverage instrumentation for all rocm-systems projects" OFF) +option(THEROCK_COVERAGE_ALL + "Enable coverage instrumentation for all coverage-enabled projects" OFF) + +# Emit the per-group project lists from COVERAGE_PROJECTS. CMAKE_CONFIGURE_DEPENDS +# re-runs configure when that file changes. +set_property(DIRECTORY APPEND PROPERTY CMAKE_CONFIGURE_DEPENDS + "${CMAKE_CURRENT_SOURCE_DIR}/build_tools/github_actions/configure_coverage_ci.py") +file(MAKE_DIRECTORY "${CMAKE_CURRENT_BINARY_DIR}/cmake") +execute_process( + COMMAND ${Python3_EXECUTABLE} + "${CMAKE_CURRENT_SOURCE_DIR}/build_tools/github_actions/configure_coverage_ci.py" + "--emit-cmake" "${CMAKE_CURRENT_BINARY_DIR}/cmake/therock_coverage_projects.cmake" + COMMAND_ERROR_IS_FATAL ANY) +include("${CMAKE_CURRENT_BINARY_DIR}/cmake/therock_coverage_projects.cmake") + +set(_therock_coverage_selected "") +if(THEROCK_COVERAGE_ALL OR THEROCK_COVERAGE_ROCM_LIBRARIES_ALL) + list(APPEND _therock_coverage_selected ${THEROCK_COVERAGE_ROCM_LIBRARIES_PROJECTS}) +endif() +if(THEROCK_COVERAGE_ALL OR THEROCK_COVERAGE_ROCM_SYSTEMS_ALL) + list(APPEND _therock_coverage_selected ${THEROCK_COVERAGE_ROCM_SYSTEMS_PROJECTS}) +endif() + +if(THEROCK_COVERAGE_ROCM_SYSTEMS_ALL AND NOT THEROCK_COVERAGE_ROCM_SYSTEMS_PROJECTS) + message(WARNING + "THEROCK_COVERAGE_ROCM_SYSTEMS_ALL is ON but no rocm-systems projects are " + "onboarded to coverage yet; nothing will be instrumented for that group.") +endif() + +foreach(_proj IN LISTS _therock_coverage_selected) + string(TOUPPER "${_proj}" _proj_upper) + # Respect an explicit -D_ENABLE_COVERAGE=OFF from the command line. + if(NOT DEFINED ${_proj_upper}_ENABLE_COVERAGE) + set(${_proj_upper}_ENABLE_COVERAGE ON) + message(STATUS "Coverage: enabling ${_proj_upper}_ENABLE_COVERAGE") + endif() +endforeach() + ################################################################################ # Feature selection # Each feature added via therock_add_feature produces a boolean cache variable diff --git a/build_tools/_therock_utils/artifact_backend.py b/build_tools/_therock_utils/artifact_backend.py index 357839ba537..182c2851bbb 100644 --- a/build_tools/_therock_utils/artifact_backend.py +++ b/build_tools/_therock_utils/artifact_backend.py @@ -334,6 +334,7 @@ def create_backend_from_env( run_id: Optional[str] = None, github_repository: Optional[str] = None, platform: Optional[str] = None, + release_type: Optional[str] = None, ) -> ArtifactBackend: """Create the appropriate backend based on environment variables. @@ -344,6 +345,10 @@ def create_backend_from_env( For S3 backend (when THEROCK_LOCAL_STAGING_DIR is not set): - Uses WorkflowOutputRoot.from_workflow_run() for bucket selection + + Args: + release_type: Release type override (e.g. "ci", "nightly"). If None, + falls back to the RELEASE_TYPE environment variable. """ import platform as platform_module @@ -363,6 +368,9 @@ def create_backend_from_env( ) output_root = WorkflowOutputRoot.from_workflow_run( - run_id=run_id, platform=platform_name, github_repository=github_repository + run_id=run_id, + platform=platform_name, + github_repository=github_repository, + release_type=release_type, ) return S3Backend(output_root=output_root) diff --git a/build_tools/_therock_utils/elf_phdr.py b/build_tools/_therock_utils/elf_phdr.py new file mode 100644 index 00000000000..ebdb66ca4d4 --- /dev/null +++ b/build_tools/_therock_utils/elf_phdr.py @@ -0,0 +1,175 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +"""Puts a relocated ELF program header table where in-process readers look. + +kpack moves the program header table of a binary it rewrites to the end of the +file, mapped by a PT_LOAD of its own whose p_vaddr is only page-congruent with +the table's file offset. ld.so copes, but code that finds the table at +__ehdr_start + e_phoff reads memory that is not mapped. compiler-rt's profile +writer does exactly that to record binary IDs, so an instrumented library that +kpack has rewritten segfaults at exit before writing a single counter. + +kpack pins p_vaddr == p_offset for executables (normalize_phdr_vaddr in +rocm_kpack/elf/phdr_manager.py) and leaves shared libraries alone. +pin_phdr_table() applies the same edit to a shared library. +""" + +import os +import struct +from pathlib import Path + +_EHDR = struct.Struct("<16sHHIQQQIHHHHHH") +_PHDR = struct.Struct(" bytes: + f.seek(offset) + data = f.read(size) + if len(data) != size: + raise ValueError(f"truncated ELF: wanted {size} bytes at {offset:#x}") + return data + + +def _round_up(value: int, alignment: int) -> int: + return (value + alignment - 1) // alignment * alignment + + +def _section_names(f, e_shoff: int, e_shentsize: int, e_shnum: int, e_shstrndx: int): + if not e_shoff or e_shentsize < _SHDR.size: + return set() + first = _SHDR.unpack(_read_at(f, e_shoff, _SHDR.size)) + if e_shnum == 0: + e_shnum = first[5] + if e_shstrndx == _SHN_XINDEX: + e_shstrndx = first[6] + if e_shstrndx >= e_shnum: + return set() + table = _read_at(f, e_shoff, e_shnum * e_shentsize) + sections = [_SHDR.unpack_from(table, i * e_shentsize) for i in range(e_shnum)] + strtab = sections[e_shstrndx] + names = _read_at(f, strtab[4], strtab[5]) + return { + names[s[0] : names.find(b"\0", s[0])].decode(errors="replace") + for s in sections + if s[0] < len(names) + } + + +def pin_phdr_table(path: Path, *, require_section: str | None = None) -> bool: + """Pins a shared library's relocated program header table to p_vaddr == p_offset. + + When the table already sits above every other PT_LOAD's address range this + only rewrites two program headers; otherwise the table is copied to a new + end-of-file offset above that range and e_phoff follows it, as kpack does + for executables. + + Leaves the file untouched, returning False, unless it is a little-endian + ELF64 x86-64 shared library without a PT_INTERP whose table lives in a + PT_LOAD of its own, the last one, starting at e_phoff and mapped somewhere + else, and, when require_section is given, it has a section of that name. + Returns True if the file was modified. + """ + with open(path, "r+b") as f: + header = f.read(_EHDR.size) + if len(header) < _EHDR.size or header[:4] != b"\x7fELF": + return False + if header[4] != 2 or header[5] != 1: # ELFCLASS64, ELFDATA2LSB + return False + ( + _, + e_type, + e_machine, + _, + _, + e_phoff, + e_shoff, + _, + _, + e_phentsize, + e_phnum, + e_shentsize, + e_shnum, + e_shstrndx, + ) = _EHDR.unpack(header) + if e_type != _ET_DYN or e_machine != _EM_X86_64: + return False + if e_phnum in (0, _PN_XNUM) or e_phentsize < _PHDR.size: + return False + + table = bytearray(_read_at(f, e_phoff, e_phnum * e_phentsize)) + phdrs = [ + list(_PHDR.unpack_from(table, i * e_phentsize)) for i in range(e_phnum) + ] + if any(p[_TYPE] == _PT_INTERP for p in phdrs): + return False + cover_index = next( + ( + i + for i, p in enumerate(phdrs) + if p[_TYPE] == _PT_LOAD + and p[_OFFSET] <= e_phoff < p[_OFFSET] + p[_FILESZ] + ), + None, + ) + if cover_index is None: + return False + cover = phdrs[cover_index] + if cover[_OFFSET] != e_phoff or cover[_VADDR] == cover[_OFFSET]: + return False + # The pinned table ends up above every other segment, so only the last + # PT_LOAD can take it without breaking their ascending p_vaddr order. + if any(p[_TYPE] == _PT_LOAD for p in phdrs[cover_index + 1 :]): + return False + if require_section is not None and require_section not in _section_names( + f, e_shoff, e_shentsize, e_shnum, e_shstrndx + ): + return False + + alignment = max(cover[_ALIGN], _PAGE_SIZE) + ceiling = _round_up( + max( + ( + p[_VADDR] + p[_MEMSZ] + for i, p in enumerate(phdrs) + if p[_TYPE] == _PT_LOAD and i != cover_index + ), + default=0, + ), + alignment, + ) + new_phoff = e_phoff + if e_phoff < ceiling: + # Pinning the table where it is would overlap another segment. + # Copy the whole covering segment, spare slots included, so the + # file holds every byte the PT_LOAD claims. + cover_bytes = _read_at(f, e_phoff, cover[_FILESZ]) + file_size = f.seek(0, os.SEEK_END) + new_phoff = _round_up(max(file_size, ceiling), alignment) + f.write(b"\0" * (new_phoff - file_size)) + f.write(cover_bytes) + + for i, p in enumerate(phdrs): + if i == cover_index or p[_TYPE] == _PT_PHDR: + p[_OFFSET] = p[_VADDR] = p[_PADDR] = new_phoff + _PHDR.pack_into(table, i * e_phentsize, *p) + f.seek(new_phoff) + f.write(table) + if new_phoff != e_phoff: + f.seek(_E_PHOFF_OFFSET) + f.write(struct.pack("`` + (``12345-coverage``), since both produce outputs with the same + names. platform: Platform name ('linux' or 'windows'). github_repository: Repository in 'owner/repo' format. If None, reads GITHUB_REPOSITORY env var (default: 'ROCm/TheRock'). @@ -318,7 +321,8 @@ def from_workflow_run( provided, uses it directly for fork detection and bucket cutover dating (no API call). lookup_workflow_run: If True and ``workflow_run`` is not provided, - fetches the workflow run from the GitHub API using ``run_id``. + fetches the workflow run from the GitHub API using ``run_id``, + less any ``-`` suffix the API would not recognize. Most callers running inside their own CI workflow do not need this — environment variables suffice. Set this when looking up another repository's workflow run (e.g. fetching artifacts). @@ -326,7 +330,9 @@ def from_workflow_run( None, falls back to the RELEASE_TYPE environment variable. """ workflow_run_id = ( - run_id if lookup_workflow_run and workflow_run is None else None + run_id.partition("-")[0] + if lookup_workflow_run and workflow_run is None + else None ) external_repo, bucket = _retrieve_bucket_info( github_repository=github_repository, diff --git a/build_tools/artifact_manager.py b/build_tools/artifact_manager.py index ddb5d23f0c1..121482653a7 100644 --- a/build_tools/artifact_manager.py +++ b/build_tools/artifact_manager.py @@ -236,11 +236,12 @@ def parse_excluded_components(raw_components: str) -> set[str]: return components -def parse_excluded_artifacts( +def parse_artifact_names( raw_artifacts: str, valid_artifacts: set[str], + option_name: str, ) -> set[str]: - """Parse and validate artifact names passed to --exclude-artifacts.""" + """Parse and validate a comma- or semicolon-separated artifact name list.""" artifacts = { artifact.strip() for artifact in raw_artifacts.replace(";", ",").split(",") @@ -249,7 +250,7 @@ def parse_excluded_artifacts( invalid_artifacts = artifacts - valid_artifacts if invalid_artifacts: raise ValueError( - "Invalid artifact name(s): " + f"Invalid artifact name(s) for {option_name}: " f"{', '.join(sorted(invalid_artifacts))}. " f"Valid artifacts are: {', '.join(sorted(valid_artifacts))}" ) @@ -469,9 +470,23 @@ def do_fetch(args: argparse.Namespace): ) target_families = parse_target_families(args) - excluded_artifacts = parse_excluded_artifacts( - args.exclude_artifacts, - set(topology.artifacts.keys()), + valid_artifacts = set(topology.artifacts.keys()) + + # An include list narrows the stage's inbound set; hybrid fetches use it to + # pull one instrumented artifact from a coverage run and everything else + # from a regular run. + included_artifacts = parse_artifact_names( + args.artifact_names, valid_artifacts, "--artifact-names" + ) + if included_artifacts: + log(f"Limiting to artifacts: {', '.join(sorted(included_artifacts))}") + inbound &= included_artifacts + if not inbound: + log("No requested artifacts are inbound for this stage") + return + + excluded_artifacts = parse_artifact_names( + args.exclude_artifacts, valid_artifacts, "--exclude-artifacts" ) if excluded_artifacts: log(f"Excluding artifacts: {', '.join(sorted(excluded_artifacts))}") @@ -1341,6 +1356,13 @@ def main(argv: Optional[List[str]] = None): help="Comma- or semicolon-separated artifact components to exclude " f"when fetching. Valid components: {', '.join(ARTIFACT_COMPONENTS)}", ) + fetch_parser.add_argument( + "--artifact-names", + type=str, + default="", + help="Comma- or semicolon-separated artifact names to fetch. Empty " + "(the default) fetches every artifact inbound to the stage.", + ) fetch_parser.add_argument( "--exclude-artifacts", type=str, diff --git a/build_tools/github_actions/configure_coverage_ci.py b/build_tools/github_actions/configure_coverage_ci.py new file mode 100644 index 00000000000..727d10a48cd --- /dev/null +++ b/build_tools/github_actions/configure_coverage_ci.py @@ -0,0 +1,907 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +"""Builds the job matrix for TheRock's nightly code coverage CI workflow. + +Coverage is opt-in per project: instrumenting a library slows it down +substantially and only pays off for projects whose test suites are good enough +to produce a meaningful signal. `COVERAGE_PROJECTS` below is the allowlist, and +everything the coverage workflows need to know about a project lives there. + +Environment variables: + - PROJECTS_TO_TEST: comma-separated project keys; empty selects every + project in the allowlist. + - AMDGPU_FAMILIES: comma-separated GPU families to build coverage for. + - COVERAGE_CONFIG_SOURCE: "/@" holding the per-project + coverage metadata files (currently ROCm/rocm-libraries). + +Outputs (GITHUB_OUTPUT): + - coverage_matrix: JSON array of per-project job configurations. + - dist_amdgpu_families: semicolon-separated families, as CMake expects them. + - families_matrix_json: JSON array of {amdgpu_family} objects for the + per-arch stages of multi_arch_build_portable_linux.yml. + - coverage_cmake_options: the coverage flags for the nightly full-stack + instrumented build, collapsed to a THEROCK_COVERAGE_*_ALL group option + whenever the selection covers a whole group. +""" + +import argparse +import json +import os +import sys +from dataclasses import dataclass, field +from pathlib import Path +from typing import Optional + +sys.path.insert(0, os.fspath(Path(__file__).resolve().parent)) +sys.path.insert(0, os.fspath(Path(__file__).resolve().parents[1])) + +from github_actions_api import gha_set_output + +DEFAULT_AMDGPU_FAMILIES = "gfx94X-dcgpu" +DEFAULT_COVERAGE_CONFIG_SOURCE = "ROCm/rocm-libraries@main" + +# Values for CoverageProject.source_repo: the repo a project ships from. Drives +# the group aliases (rocm_libraries_all / rocm_systems_all) below. +ROCM_LIBRARIES = "rocm-libraries" +ROCM_SYSTEMS = "rocm-systems" + +# Build stages that can carry an instrumented project. compiler-runtime is +# always built because every other stage takes its inbound artifacts from it. +STAGE_COMPILER_RUNTIME = "compiler-runtime" +STAGE_MATH_LIBS = "math-libs" +STAGE_COMM_LIBS = "comm-libs" +STAGE_PROFILER_APPS = "profiler-apps" + +# Stages a project may be registered against, validated below. +KNOWN_STAGES = frozenset( + { + STAGE_COMPILER_RUNTIME, + STAGE_MATH_LIBS, + STAGE_COMM_LIBS, + STAGE_PROFILER_APPS, + } +) + +# Only compiler-runtime and math-libs have build jobs. rccl and rocshmem +# (comm-libs) are blocked upstream, so selecting them is rejected here rather +# than failing hours later on a missing inbound artifact. +BUILDABLE_STAGES = frozenset({STAGE_COMPILER_RUNTIME, STAGE_MATH_LIBS}) + + +@dataclass(frozen=True) +class CoverageProject: + """Everything the coverage pipeline needs to know about one project. + + Attributes: + cmake_target: TheRock subproject name. Upper cased to form the + `_ENABLE_COVERAGE` flag, which is TheRock's own knob and + is not what the subproject sees; see coverage_option. + coverage_option: CMake option the upstream project implements to turn + its instrumentation on. These names are not standardised -- most + projects use `BUILD_CODE_COVERAGE` or `CODE_COVERAGE`, a few use + `_ENABLE_COVERAGE`, RCCL uses `ENABLE_CODE_COVERAGE` -- + so therock_subproject.cmake translates TheRock's flag into this + name when configuring the subproject. Being per-subproject is what + keeps the generic names from instrumenting the whole build. + unsupported_reason: Set instead of coverage_option when the project + cannot be measured by this pipeline. Such a project stays in the + registry so the gap is recorded, but is left out of the group + aliases and rejected if named explicitly. + blocked_reason: Set alongside coverage_option when the project is + measurable in principle but its instrumented build is currently + broken. Unlike unsupported_reason this is expected to be + temporary, so the project only drops out of the default selection + and the group aliases; naming it still works, which is how the + block gets retested once the fix lands. + artifact_names: BUILD_TOPOLOGY artifact(s) the instrumented project + ships in. These are grouped (`rand` holds both rocRAND and + hipRAND), which is why artifact_relpaths exists. + artifact_relpaths: Subproject stage directories inside those artifacts + that belong to this project alone. The nightly hybrid fetch copies + only these over the baseline install tree, so a run measuring + hipRAND does not also pick up an instrumented rocRAND. + stage: Build stage that produces the project. The nightly needs a build + job for this stage, so onboarding a project from a new stage means + adding one. + test_component: Key in fetch_test_configurations.py's test matrix. + coverage_config: Per-project coverage metadata file, relative to the + root of the repository named by COVERAGE_CONFIG_SOURCE. + object_globs: Globs, relative to the extracted artifact directory, + matching the instrumented binaries handed to `llvm-cov`. A glob + prefixed with `!` removes its matches instead, for test binaries + that share an install directory with a sibling project's. + fetch_artifact_args: Arguments to install_rocm_from_artifacts.py that + pull the instrumented libraries into the report generation job. + codecov_flag: Flag the report is filed under in Codecov. + source_repo: The repo this project ships from (ROCM_LIBRARIES or + ROCM_SYSTEMS); drives the group aliases. + extra_cmake_targets: Additional subprojects to instrument alongside + cmake_target. Header-only projects report against their own test + binaries, and some of those are built by a sibling subproject + (rocPRIM's live in rocPRIM_tests), which therefore has to be + instrumented too or the binaries carry no coverage mapping. + """ + + cmake_target: str + stage: str + test_component: str + coverage_config: str + coverage_option: str = "" + unsupported_reason: str = "" + blocked_reason: str = "" + artifact_names: list[str] = field(default_factory=list) + artifact_relpaths: list[str] = field(default_factory=list) + object_globs: list[str] = field(default_factory=list) + fetch_artifact_args: str = "" + codecov_flag: str = "" + source_repo: str = ROCM_LIBRARIES + extra_cmake_targets: list[str] = field(default_factory=list) + + +# Projects that participate in coverage CI. +# +# A run never instruments this whole list unless it is asked to. `PROJECTS_TO_TEST` +# selects a subset by name, or by one of the group aliases below, and only the +# stages that subset needs get built. Selecting everything is possible and +# expensive: instrumented builds are substantially slower than normal ones. +# +# A selection covering several projects instruments all of them in one build and +# still reports on each separately, because `object_globs` is what scopes a +# report: llvm-cov only reports functions found in the objects it is handed, so +# counters emitted by a sibling project are never looked up. +# +# Header-only projects (rocPRIM, hipCUB, rocThrust, rocWMMA) ship no library to +# report against, so their object_globs name their installed test binaries +# instead, and their fetch args include --tests to bring those binaries into the +# report job. +COVERAGE_PROJECTS: dict[str, CoverageProject] = { + # + # rocm-libraries -- math-libs stage + # + "rocrand": CoverageProject( + cmake_target="rocRAND", + artifact_names=["rand"], + artifact_relpaths=["math-libs/rocRAND/stage"], + coverage_option="CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="rocrand", + coverage_config="projects/rocrand/test_categories_coverage.yaml", + object_globs=["lib/librocrand.so*"], + fetch_artifact_args="--rand", + codecov_flag="rocRAND", + ), + "hiprand": CoverageProject( + cmake_target="hipRAND", + artifact_names=["rand"], + artifact_relpaths=["math-libs/hipRAND/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="hiprand", + coverage_config="projects/hiprand/test_categories_coverage.yaml", + object_globs=["lib/libhiprand.so*"], + fetch_artifact_args="--rand", + codecov_flag="hipRAND", + ), + "rocfft": CoverageProject( + cmake_target="rocFFT", + artifact_names=["fft"], + artifact_relpaths=["math-libs/rocFFT/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="rocfft", + coverage_config="projects/rocfft/test_categories_coverage.yaml", + object_globs=["lib/librocfft.so*"], + fetch_artifact_args="--fft", + codecov_flag="rocFFT", + ), + "hipfft": CoverageProject( + cmake_target="hipFFT", + artifact_names=["fft"], + artifact_relpaths=["math-libs/hipFFT/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="hipfft", + coverage_config="projects/hipfft/test_categories_coverage.yaml", + object_globs=["lib/libhipfft.so*"], + fetch_artifact_args="--fft", + codecov_flag="hipFFT", + ), + "rocblas": CoverageProject( + cmake_target="rocBLAS", + artifact_names=["blas"], + artifact_relpaths=["math-libs/BLAS/rocBLAS/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="rocblas", + coverage_config="projects/rocblas/test_categories_coverage.yaml", + object_globs=["lib/librocblas.so*"], + fetch_artifact_args="--blas", + codecov_flag="rocBLAS", + ), + "hipblas": CoverageProject( + cmake_target="hipBLAS", + artifact_names=["blas"], + artifact_relpaths=["math-libs/BLAS/hipBLAS/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="hipblas", + coverage_config="projects/hipblas/test_categories_coverage.yaml", + object_globs=["lib/libhipblas.so*"], + fetch_artifact_args="--blas", + codecov_flag="hipBLAS", + ), + "hipblaslt": CoverageProject( + cmake_target="hipBLASLt", + artifact_names=["blas"], + artifact_relpaths=["math-libs/BLAS/hipBLASLt/stage"], + coverage_option="HIPBLASLT_ENABLE_COVERAGE", + # The coverage-only branch of clients/CMakeLists.txt links hipblaslt-test + # against a bare `rocroller` while its neighbours use the imported + # targets. rocRoller exports as roc::rocroller, and under TheRock it is a + # separate subproject found via its package config, so the bare name is + # not a target here and reaches the linker as -lrocroller with no -L to + # resolve it. hipBLASLt's own build gets away with it by having rocRoller + # in-tree. Fix belongs upstream in ROCm/rocm-libraries. + blocked_reason=( + "hipblaslt-test fails to link with 'unable to find library " + "-lrocroller'; its coverage build needs roc::rocroller upstream" + ), + stage=STAGE_MATH_LIBS, + test_component="hipblaslt", + coverage_config="projects/hipblaslt/test_categories_coverage.yaml", + object_globs=["lib/libhipblaslt.so*"], + fetch_artifact_args="--blas", + codecov_flag="hipBLASLt", + ), + "rocsparse": CoverageProject( + cmake_target="rocSPARSE", + artifact_names=["sparse"], + artifact_relpaths=["math-libs/BLAS/rocSPARSE/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="rocsparse", + coverage_config="projects/rocsparse/test_categories_coverage.yaml", + object_globs=["lib/librocsparse.so*"], + fetch_artifact_args="--sparse", + codecov_flag="rocSPARSE", + ), + "hipsparse": CoverageProject( + cmake_target="hipSPARSE", + artifact_names=["sparse"], + artifact_relpaths=["math-libs/BLAS/hipSPARSE/stage"], + unsupported_reason=( + "HIPSPARSE_ENABLE_COVERAGE selects gcov instrumentation, which writes .gcda files rather than the .profraw this pipeline merges" + ), + stage=STAGE_MATH_LIBS, + test_component="hipsparse", + coverage_config="projects/hipsparse/test_categories_coverage.yaml", + object_globs=["lib/libhipsparse.so*"], + fetch_artifact_args="--sparse", + codecov_flag="hipSPARSE", + ), + "hipsparselt": CoverageProject( + cmake_target="hipSPARSELt", + artifact_names=["sparse"], + artifact_relpaths=["math-libs/BLAS/hipSPARSELt/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="hipsparselt", + coverage_config="projects/hipsparselt/test_categories_coverage.yaml", + object_globs=["lib/libhipsparselt.so*"], + fetch_artifact_args="--sparse", + codecov_flag="hipSPARSELt", + ), + "rocsolver": CoverageProject( + cmake_target="rocSOLVER", + artifact_names=["solver"], + artifact_relpaths=["math-libs/BLAS/rocSOLVER/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="rocsolver", + coverage_config="projects/rocsolver/test_categories_coverage.yaml", + object_globs=["lib/librocsolver.so*"], + fetch_artifact_args="--solver", + codecov_flag="rocSOLVER", + ), + "hipsolver": CoverageProject( + cmake_target="hipSOLVER", + artifact_names=["solver"], + artifact_relpaths=["math-libs/BLAS/hipSOLVER/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="hipsolver", + coverage_config="projects/hipsolver/test_categories_coverage.yaml", + object_globs=["lib/libhipsolver.so*"], + fetch_artifact_args="--solver", + codecov_flag="hipSOLVER", + ), + "rocalution": CoverageProject( + cmake_target="rocALUTION", + artifact_names=["rocalution"], + artifact_relpaths=["math-libs/rocALUTION/stage"], + unsupported_reason=( + "BUILD_CODE_COVERAGE selects gcov instrumentation, which writes .gcda files rather than the .profraw this pipeline merges" + ), + stage=STAGE_MATH_LIBS, + test_component="rocalution", + coverage_config="projects/rocalution/test_categories_coverage.yaml", + object_globs=["lib/librocalution.so*"], + fetch_artifact_args="--rocalution", + codecov_flag="rocALUTION", + ), + "hiptensor": CoverageProject( + cmake_target="hipTensor", + artifact_names=["hiptensor"], + artifact_relpaths=["math-libs/hipTensor/stage"], + # Prefixed, unlike rocALUTION's bare CODE_COVERAGE just above. Getting + # this wrong is silent: the option is simply never read, the library + # builds uninstrumented, and the tests pass having written no profile. + coverage_option="HIPTENSOR_CODE_COVERAGE", + # Instrumenting hipTensor pushes libhiptensor.so past the 2GB that the + # small code model allows, and the link dies on out-of-range + # R_X86_64_PC32 relocations from .rodata into .text. Upstream applies + # coverage as a blanket add_compile_options(-fprofile-instr-generate + # -fcoverage-mapping), and hipTensor is built on composable_kernel, so + # the coverage mapping and name data scale with an already enormous + # amount of instantiated template code. Building it with a larger code + # model is the likely fix and belongs upstream. Blocked meanwhile + # because the failure takes the whole math-libs stage with it, and + # every other project in the stage along with it. + blocked_reason=( + "libhiptensor.so exceeds the small code model when instrumented; " + "the link fails on out-of-range R_X86_64_PC32 relocations" + ), + stage=STAGE_MATH_LIBS, + test_component="hiptensor", + coverage_config="projects/hiptensor/test_categories_coverage.yaml", + object_globs=["lib/libhiptensor.so*"], + fetch_artifact_args="--hiptensor", + codecov_flag="hipTensor", + ), + "miopen": CoverageProject( + cmake_target="MIOpen", + artifact_names=["miopen"], + artifact_relpaths=["ml-libs/MIOpen/stage"], + unsupported_reason="no coverage option in its CMake", + stage=STAGE_MATH_LIBS, + test_component="miopen", + coverage_config="projects/miopen/test_categories_coverage.yaml", + object_globs=["lib/libMIOpen.so*"], + fetch_artifact_args="--miopen", + codecov_flag="MIOpen", + ), + "hipdnn": CoverageProject( + cmake_target="hipDNN", + artifact_names=["hipdnn"], + artifact_relpaths=["ml-libs/hipDNN/stage"], + coverage_option="HIPDNN_ENABLE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="hipdnn", + coverage_config="projects/hipdnn/test_categories_coverage.yaml", + # Not libhipdnn.so: hipDNN's shared library is the backend, which is + # also the name TheRock's own packaging metadata records for it. + object_globs=["lib/libhipdnn_backend.so*"], + fetch_artifact_args="--hipdnn", + codecov_flag="hipDNN", + ), + # + # rocm-libraries -- math-libs stage, header-only + # + "rocprim": CoverageProject( + cmake_target="rocPRIM", + artifact_names=["prim"], + artifact_relpaths=["math-libs/rocPRIM/stage", "math-libs/rocPRIM_tests/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="rocprim", + coverage_config="projects/rocprim/test_categories_coverage.yaml", + # hipCUB's tests install into the same bin/ as test_hipcub_*. + object_globs=["bin/test_*", "!bin/test_hipcub_*"], + fetch_artifact_args="--prim --tests", + codecov_flag="rocPRIM", + # rocPRIM's tests are a sibling subproject, so both stage dirs are + # overlaid and both are instrumented. + extra_cmake_targets=["rocPRIM_tests"], + ), + "hipcub": CoverageProject( + cmake_target="hipCUB", + artifact_names=["prim"], + artifact_relpaths=["math-libs/hipCUB/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="hipcub", + coverage_config="projects/hipcub/test_categories_coverage.yaml", + object_globs=["bin/test_hipcub_*"], + fetch_artifact_args="--prim --tests", + codecov_flag="hipCUB", + ), + "rocthrust": CoverageProject( + cmake_target="rocThrust", + artifact_names=["prim"], + artifact_relpaths=["math-libs/rocThrust/stage"], + coverage_option="CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="rocthrust", + coverage_config="projects/rocthrust/test_categories_coverage.yaml", + object_globs=["bin/*.hip"], + fetch_artifact_args="--prim --tests", + codecov_flag="rocThrust", + ), + "rocwmma": CoverageProject( + cmake_target="rocWMMA", + artifact_names=["rocwmma"], + artifact_relpaths=["math-libs/rocWMMA/stage"], + # Prefixed, like hipTensor's above; a bare CODE_COVERAGE is never read. + coverage_option="ROCWMMA_CODE_COVERAGE", + stage=STAGE_MATH_LIBS, + test_component="rocwmma", + coverage_config="projects/rocwmma/test_categories_coverage.yaml", + object_globs=["bin/*_test*", "bin/gemm_*-validate"], + fetch_artifact_args="--rocwmma --tests", + codecov_flag="rocWMMA", + ), + # + # rocm-systems -- comm-libs stage + # + "rccl": CoverageProject( + cmake_target="rccl", + artifact_names=["rccl"], + artifact_relpaths=["comm-libs/rccl/stage"], + coverage_option="ENABLE_CODE_COVERAGE", + stage=STAGE_COMM_LIBS, + test_component="rccl", + coverage_config="projects/rccl/test_categories_coverage.yaml", + object_globs=["lib/librccl.so*"], + fetch_artifact_args="--rccl", + codecov_flag="rccl", + source_repo=ROCM_SYSTEMS, + ), + "rocshmem": CoverageProject( + cmake_target="rocshmem", + artifact_names=["rocshmem"], + artifact_relpaths=["comm-libs/rocshmem/stage"], + coverage_option="BUILD_CODE_COVERAGE", + stage=STAGE_COMM_LIBS, + test_component="rocshmem", + coverage_config="projects/rocshmem/test_categories_coverage.yaml", + object_globs=["lib/librocshmem.so*"], + fetch_artifact_args="--rocshmem", + codecov_flag="rocSHMEM", + source_repo=ROCM_SYSTEMS, + ), + # + # rocm-systems -- compiler-runtime stage + # + "rocprofiler-sdk": CoverageProject( + cmake_target="rocprofiler-sdk", + artifact_names=["rocprofiler-sdk"], + artifact_relpaths=["profiler/rocprofiler-sdk/stage"], + unsupported_reason=( + "ROCPROFILER_BUILD_CODECOV selects gcov instrumentation, which writes .gcda files rather than the .profraw this pipeline merges" + ), + stage=STAGE_COMPILER_RUNTIME, + test_component="rocprofiler-sdk", + coverage_config="projects/rocprofiler-sdk/test_categories_coverage.yaml", + object_globs=["lib/librocprofiler-sdk.so*"], + fetch_artifact_args="--rocprofiler-sdk", + codecov_flag="rocprofiler-sdk", + source_repo=ROCM_SYSTEMS, + ), + "aqlprofile": CoverageProject( + cmake_target="aqlprofile", + artifact_names=["aqlprofile"], + artifact_relpaths=["profiler/aqlprofile/stage"], + unsupported_reason="no coverage option in its CMake", + stage=STAGE_COMPILER_RUNTIME, + test_component="aqlprofile", + coverage_config="projects/aqlprofile/test_categories_coverage.yaml", + object_globs=["lib/libhsa-amd-aqlprofile*.so*"], + fetch_artifact_args="--aqlprofile", + codecov_flag="aqlprofile", + source_repo=ROCM_SYSTEMS, + ), + "amdsmi": CoverageProject( + cmake_target="amdsmi", + artifact_names=["core-amdsmi"], + artifact_relpaths=["core/amdsmi/stage"], + unsupported_reason="no coverage option in its CMake", + stage=STAGE_COMPILER_RUNTIME, + test_component="amdsmi", + coverage_config="projects/amdsmi/test_categories_coverage.yaml", + object_globs=["lib/libamd_smi.so*"], + fetch_artifact_args="--base-only", + codecov_flag="amdsmi", + source_repo=ROCM_SYSTEMS, + ), + # + # rocm-systems -- profiler-apps stage + # + "rocprofiler-compute": CoverageProject( + cmake_target="rocprofiler-compute", + artifact_names=["rocprofiler-compute"], + artifact_relpaths=["profiler/rocprofiler-compute/stage"], + unsupported_reason="a Python tool with no native instrumentation option", + stage=STAGE_PROFILER_APPS, + test_component="rocprofiler-compute", + coverage_config="projects/rocprofiler-compute/test_categories_coverage.yaml", + object_globs=["lib/librocprofiler-compute*.so*"], + fetch_artifact_args="--rocprofiler-compute", + codecov_flag="rocprofiler-compute", + source_repo=ROCM_SYSTEMS, + ), + "rocprofiler-systems": CoverageProject( + cmake_target="rocprofiler-systems", + artifact_names=["rocprofiler-systems"], + artifact_relpaths=["profiler/rocprofiler-systems/stage"], + unsupported_reason=( + "no build-time coverage option; its coverage feature instruments other programs at runtime" + ), + stage=STAGE_PROFILER_APPS, + test_component="rocprofiler-systems", + coverage_config="projects/rocprofiler-systems/test_categories_coverage.yaml", + object_globs=["lib/librocprofiler-systems*.so*"], + fetch_artifact_args="--rocprofiler-systems", + codecov_flag="rocprofiler-systems", + source_repo=ROCM_SYSTEMS, + ), +} + +_VALID_SOURCE_REPOS = frozenset({ROCM_LIBRARIES, ROCM_SYSTEMS}) +for _key, _proj in COVERAGE_PROJECTS.items(): + assert ( + _proj.source_repo in _VALID_SOURCE_REPOS + ), f"{_key}: source_repo must be one of {sorted(_VALID_SOURCE_REPOS)}" + # A project in a stage the workflow cannot build would be selectable and + # then silently tested against uninstrumented binaries. + assert ( + _proj.stage in KNOWN_STAGES + ), f"{_key}: stage must be one of {sorted(KNOWN_STAGES)}" + # An entry with neither would be selectable and instrument nothing; one + # with both leaves it ambiguous whether the project can be measured. + assert bool(_proj.coverage_option) != bool( + _proj.unsupported_reason + ), f"{_key}: set exactly one of coverage_option and unsupported_reason" + # Without these the nightly test job cannot tell which files belong to the + # project, and would measure an entirely non-instrumented install. + if _proj.coverage_option: + assert _proj.artifact_names, f"{_key}: needs artifacts to overlay" + assert _proj.artifact_relpaths, f"{_key}: needs relpaths to overlay" + +SUPPORTED_PROJECTS: frozenset[str] = frozenset( + k for k, v in COVERAGE_PROJECTS.items() if v.coverage_option +) +# Measurable, but the instrumented build is broken today. Kept selectable by +# name so the block can be retested, and kept out of everything that picks +# projects on the caller's behalf. +BLOCKED_PROJECTS: frozenset[str] = frozenset( + k for k, v in COVERAGE_PROJECTS.items() if v.blocked_reason +) +# What an empty selection defaults to: everything measurable that this workflow +# also has a build job for. Naming a project outside BUILDABLE_STAGES stays an +# error, so the narrowing applies only to the default, never to an explicit ask. +DEFAULT_PROJECTS: frozenset[str] = frozenset( + k + for k in SUPPORTED_PROJECTS - BLOCKED_PROJECTS + if COVERAGE_PROJECTS[k].stage in BUILDABLE_STAGES +) +# Group membership is limited to what can actually be measured, so a group +# alias never schedules a job that is guaranteed to fail for want of profiles. +ROCM_LIBRARIES_PROJECTS: frozenset[str] = frozenset( + k + for k, v in COVERAGE_PROJECTS.items() + if v.source_repo == ROCM_LIBRARIES and v.coverage_option and not v.blocked_reason +) +ROCM_SYSTEMS_PROJECTS: frozenset[str] = frozenset( + k + for k, v in COVERAGE_PROJECTS.items() + if v.source_repo == ROCM_SYSTEMS and v.coverage_option and not v.blocked_reason +) +# Projects shipped in the prim artifact, built by the legacy rocPRIM/hipCUB/ +# rocThrust subprojects or, by default, the unified hipCCL superbuild. +PRIM_PROJECTS: frozenset[str] = frozenset( + k for k, v in COVERAGE_PROJECTS.items() if "prim" in v.artifact_names +) + +# Group-alias tokens accepted by PROJECTS_TO_TEST (and the projects_to_test CI +# input) that expand to a whole component group. Matched case-insensitively. +_GROUP_ALIASES: dict[str, frozenset[str]] = { + "rocm_libraries_all": ROCM_LIBRARIES_PROJECTS, + "rocm_systems_all": ROCM_SYSTEMS_PROJECTS, + "all": SUPPORTED_PROJECTS - BLOCKED_PROJECTS, +} + + +def parse_projects(raw_projects: str) -> list[str]: + """Resolves PROJECTS_TO_TEST into an ordered list of known project keys. + + Recognizes the group aliases 'rocm_libraries_all', 'rocm_systems_all', and + 'all', which expand to a whole component group and may be mixed with + explicit project names. Matching is case-insensitive. + + An empty selection means every project this workflow can build, which is + narrower than 'all' while comm-libs has no build job. Asking for those + projects by name or via a group alias still fails in + resolve_build_stages(), so nothing is dropped from an explicit request. + + Projects with a known-broken instrumented build are likewise left out of + the default and the aliases, but naming one is honoured with a warning + rather than an error: retesting a block is the only way to notice it has + been fixed. + """ + requested = [p.strip().lower() for p in raw_projects.split(",") if p.strip()] + if not requested: + return sorted(DEFAULT_PROJECTS) + + expanded: list[str] = [] + unknown: list[str] = [] + unsupported: list[str] = [] + for token in requested: + if token in _GROUP_ALIASES: + group = _GROUP_ALIASES[token] + if not group: + raise ValueError( + f"Group alias '{token}' selected no projects: no such " + "projects are onboarded to coverage yet." + ) + expanded.extend(sorted(group)) + elif token in SUPPORTED_PROJECTS: + expanded.append(token) + elif token in COVERAGE_PROJECTS: + unsupported.append(token) + else: + unknown.append(token) + + if unknown: + raise ValueError( + f"Unknown coverage project(s): {', '.join(sorted(unknown))}. " + f"Coverage-enabled projects are: {', '.join(sorted(SUPPORTED_PROJECTS))}. " + f"Group aliases: {', '.join(sorted(_GROUP_ALIASES))}." + ) + + # Failing here beats scheduling a build and several hours of tests that + # cannot produce a report. + if unsupported: + detail = "; ".join( + f"{name} ({COVERAGE_PROJECTS[name].unsupported_reason})" + for name in sorted(unsupported) + ) + raise ValueError( + f"Coverage is not supported for: {detail}. These projects are " + "registered so the gap is tracked, but cannot be measured until " + "they emit LLVM profraw data." + ) + # Preserve first-seen order, drop duplicates (explicit + alias overlap). + selection = list(dict.fromkeys(expanded)) + + # Not fatal: the caller went out of their way to name these, and the only + # way to find out a block has been lifted is to run into it again. + for name in selection: + blocked_reason = COVERAGE_PROJECTS[name].blocked_reason + if blocked_reason: + print( + f"warning: {name} is expected to fail: {blocked_reason}", + file=sys.stderr, + ) + return selection + + +def parse_amdgpu_families(raw_families: str) -> list[str]: + families = [f.strip() for f in raw_families.split(",") if f.strip()] + return families or [DEFAULT_AMDGPU_FAMILIES] + + +def parse_config_source(raw_source: str) -> tuple[str, str]: + """Splits "/@" into its repository and ref.""" + source = raw_source.strip() or DEFAULT_COVERAGE_CONFIG_SOURCE + repository, separator, ref = source.partition("@") + if not separator or not repository or not ref: + raise ValueError( + f"Invalid coverage config source '{raw_source}'. " + "Expected the form '/@'." + ) + return repository, ref + + +def build_coverage_matrix( + project_keys: list[str], + amdgpu_families: list[str], + config_repository: str, + config_ref: str, +) -> list[dict]: + """Produces one job configuration per (project, GPU family) pair.""" + matrix = [] + for project_key in project_keys: + project = COVERAGE_PROJECTS[project_key] + for family in amdgpu_families: + matrix.append( + { + "project_name": project_key, + "cmake_target": project.cmake_target, + "test_component": project.test_component, + "coverage_config": project.coverage_config, + "coverage_config_repository": config_repository, + "coverage_config_ref": config_ref, + "object_globs": ",".join(project.object_globs), + "fetch_artifact_args": project.fetch_artifact_args, + # The report job fetches this stage's sources so the HTML + # rendering has something to annotate. + "build_stage": project.stage, + "codecov_flag": project.codecov_flag or project_key, + "amdgpu_families": family, + "artifact_names": ",".join(project.artifact_names), + "artifact_relpaths": ",".join(project.artifact_relpaths), + } + ) + return matrix + + +def build_coverage_cmake_options(project_keys: list[str]) -> list[str]: + """Picks the CMake coverage flags that instrument exactly this selection. + + A nightly run instruments every onboarded project, which is what the + THEROCK_COVERAGE_*_ALL group options say directly; a narrowed selection + falls back to naming each project. Both spellings reach the same + _ENABLE_COVERAGE flags, so this only affects how the configure + line reads. + """ + selected = set(project_keys) + covers_libraries = bool(ROCM_LIBRARIES_PROJECTS) and ROCM_LIBRARIES_PROJECTS <= ( + selected + ) + covers_systems = bool(ROCM_SYSTEMS_PROJECTS) and ROCM_SYSTEMS_PROJECTS <= selected + + options: list[str] = [] + grouped: set[str] = set() + if covers_libraries and covers_systems: + options.append("-DTHEROCK_COVERAGE_ALL=ON") + grouped |= ROCM_LIBRARIES_PROJECTS | ROCM_SYSTEMS_PROJECTS + else: + if covers_libraries: + options.append("-DTHEROCK_COVERAGE_ROCM_LIBRARIES_ALL=ON") + grouped |= ROCM_LIBRARIES_PROJECTS + if covers_systems: + options.append("-DTHEROCK_COVERAGE_ROCM_SYSTEMS_ALL=ON") + grouped |= ROCM_SYSTEMS_PROJECTS + + # The group options already cover extra_cmake_targets, because emit_cmake + # writes them into the group lists CMakeLists.txt expands. + for key in project_keys: + if key in grouped: + continue + project = COVERAGE_PROJECTS[key] + for target in [project.cmake_target, *project.extra_cmake_targets]: + options.append(f"-D{target.upper()}_ENABLE_COVERAGE=ON") + + # The unified hipCCL superbuild has no rocPRIM/hipCUB/rocThrust subprojects + # for their coverage flags to reach, so it ships uninstrumented tests whose + # stage dirs (hipCCL_tests) the replacement does not recognize. Build the + # legacy subprojects these entries are registered against instead. + if selected & PRIM_PROJECTS: + options.append("-DTHEROCK_FLAG_BUILD_HIPCCL=OFF") + return options + + +def resolve_build_stages(project_keys: list[str]) -> set[str]: + """Returns the stages the selection needs, rejecting stages with no build job. + + Only compiler-runtime and math-libs have build jobs. A project from any + other stage would have nothing built for it, and that would not surface + until the test job found no instrumented files to overlay, hours in, so it + is an error here instead. + """ + stages = {COVERAGE_PROJECTS[key].stage for key in project_keys} + unbuildable = sorted(stages - BUILDABLE_STAGES) + if unbuildable: + blocked = sorted( + key for key in project_keys if COVERAGE_PROJECTS[key].stage in unbuildable + ) + raise ValueError( + f"cannot build coverage for {', '.join(blocked)}: this workflow has " + f"build jobs for {', '.join(sorted(BUILDABLE_STAGES))} only, not " + f"{', '.join(unbuildable)}. Restore the staged build jobs in " + "multi_arch_ci_coverage_linux.yml to select these projects." + ) + return stages + + +def emit_cmake(output_path: Path) -> None: + """Writes the coverage group lists and option-name map as a CMake include. + + Two things come out of here. The group lists drive the + THEROCK_COVERAGE_*_ALL options, and are limited to projects that can be + measured and are not currently blocked. The option map is not: a blocked + project stays in it so that naming it explicitly still instruments it. + The option map tells therock_subproject.cmake which flag each + subproject actually implements, since the names are not standardised + upstream. CMakeLists.txt reads both so it does not hardcode either. + """ + + def targets(repo: str) -> list[str]: + names: set[str] = set() + for p in COVERAGE_PROJECTS.values(): + if p.source_repo != repo or not p.coverage_option: + continue + # Kept in step with the Python group sets: a group flag that + # instrumented a project whose instrumented build is known to fail + # would take the whole stage down with it. + if p.blocked_reason: + continue + # extra_cmake_targets are part of the group: a group flag that + # instrumented rocPRIM but not rocPRIM_tests would leave rocPRIM's + # only reportable binaries without a coverage mapping. + names.update([p.cmake_target, *p.extra_cmake_targets]) + return sorted(names, key=str.lower) + + # A sibling subproject is built from the same upstream source as its parent, + # so it takes the same option. + option_lines = [] + for p in sorted(COVERAGE_PROJECTS.values(), key=lambda v: v.cmake_target.lower()): + if not p.coverage_option: + continue + for target in sorted([p.cmake_target, *p.extra_cmake_targets], key=str.lower): + option_lines.append( + f"set(THEROCK_COVERAGE_OPTION_{target.upper()} {p.coverage_option})\n" + ) + + output_path.write_text( + "# Generated by configure_coverage_ci.py --emit-cmake; do not edit.\n" + "# Edit COVERAGE_PROJECTS in\n" + "# build_tools/github_actions/configure_coverage_ci.py instead.\n" + f"set(THEROCK_COVERAGE_ROCM_LIBRARIES_PROJECTS {' '.join(targets(ROCM_LIBRARIES))})\n" + f"set(THEROCK_COVERAGE_ROCM_SYSTEMS_PROJECTS {' '.join(targets(ROCM_SYSTEMS))})\n" + + "".join(option_lines) + ) + + +def main(argv: Optional[list[str]] = None) -> int: + parser = argparse.ArgumentParser(description=__doc__) + parser.add_argument( + "--print", + action="store_true", + dest="print_matrix", + help="Print the matrix to stdout instead of writing GITHUB_OUTPUT", + ) + parser.add_argument( + "--emit-cmake", + type=Path, + default=None, + metavar="PATH", + help="Write the coverage group lists as a CMake include to PATH, then exit.", + ) + args = parser.parse_args(argv) + + if args.emit_cmake is not None: + emit_cmake(args.emit_cmake) + return 0 + + project_keys = parse_projects(os.getenv("PROJECTS_TO_TEST", "")) + amdgpu_families = parse_amdgpu_families(os.getenv("AMDGPU_FAMILIES", "")) + config_repository, config_ref = parse_config_source( + os.getenv("COVERAGE_CONFIG_SOURCE", "") + ) + + matrix = build_coverage_matrix( + project_keys, amdgpu_families, config_repository, config_ref + ) + coverage_flags = build_coverage_cmake_options(project_keys) + needs_math_libs = STAGE_MATH_LIBS in resolve_build_stages(project_keys) + outputs = { + "coverage_matrix": json.dumps(matrix), + "dist_amdgpu_families": ";".join(amdgpu_families), + "families_matrix_json": json.dumps( + [{"amdgpu_family": family} for family in amdgpu_families] + ), + "coverage_cmake_options": " ".join(coverage_flags), + "needs_math_libs": "true" if needs_math_libs else "false", + } + + if args.print_matrix: + print(json.dumps(outputs, indent=2)) + else: + gha_set_output(outputs) + return 0 + + +if __name__ == "__main__": + sys.exit(main()) diff --git a/build_tools/github_actions/fetch_test_configurations.py b/build_tools/github_actions/fetch_test_configurations.py index a7d3076d5ba..1cffd02a751 100644 --- a/build_tools/github_actions/fetch_test_configurations.py +++ b/build_tools/github_actions/fetch_test_configurations.py @@ -51,6 +51,24 @@ def _get_artifact_path(artifact_path: str) -> str: return (OUTPUT_ARTIFACTS_DIR / artifact_path).as_posix() +# BUILD_VARIANT=coverage runs tests against instrumented libraries, which take +# far longer than the release timeouts below allow for: every executed region +# bumps a counter, and some upstream coverage options also drop optimization +# (hipRAND adds -O0) or compile in extra CPU-only test suites (rocRAND). +# The cap keeps the step inside test_code_coverage_component.yml's job +# timeout-minutes (210), which would otherwise cut it off first. +COVERAGE_TIMEOUT_MULTIPLIER = 4 +COVERAGE_TIMEOUT_CAP_MINUTES = 180 + + +def _coverage_timeout_minutes(timeout_minutes: int) -> int: + """Scales a component's timeout for an instrumented run, never shortening it.""" + scaled = min( + timeout_minutes * COVERAGE_TIMEOUT_MULTIPLIER, COVERAGE_TIMEOUT_CAP_MINUTES + ) + return max(timeout_minutes, scaled) + + # Maps a group label (the part after "test:") to the individual test matrix # keys it expands to. Use this when a single label should select multiple # related jobs without relying on name-prefix inference. @@ -1628,6 +1646,11 @@ def run(): job_config_data["timeout_minutes"] + 15 ) + if build_variant == "coverage": + job_config_data["timeout_minutes"] = _coverage_timeout_minutes( + job_config_data["timeout_minutes"] + ) + # For CI testing, we construct a shard array based on "total_shards" from "fetch_test_configurations.py" # This way, the test jobs will be split up into X shards. (ex: [1, 2, 3, 4] = 4 test shards) # For display purposes, we add "i + 1" for the job name (ex: 1 of 4). During the actual test sharding in the test executable, this array will become 0th index diff --git a/build_tools/github_actions/merge_coverage_report.py b/build_tools/github_actions/merge_coverage_report.py new file mode 100644 index 00000000000..92d0f8e12f2 --- /dev/null +++ b/build_tools/github_actions/merge_coverage_report.py @@ -0,0 +1,460 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +"""Turns the profraw files produced by a coverage test run into an lcov report. + +Instrumented binaries emit one `.profraw` per process, so a sharded test job +produces many of them spread across downloaded artifact directories. This +merges them into a single `.profdata` index and exports lcov for the +instrumented objects. + +Three renderings come out of the same index. lcov is what Codecov consumes, a +text table is what a reader wants when the question is just "what is the +number", and the HTML tree is the annotated line-by-line view. The last one is +the only one that reads the source files: the others report from the coverage +mappings alone. + +The `llvm-profdata` and `llvm-cov` binaries must come from the same compiler +that built the instrumented objects; a version mismatch produces an +unhelpfully generic "malformed instrumentation profile data" error. Both live +under `lib/llvm/bin` of an installed ROCm distribution. + +A report that measured nothing fails rather than publishing 0%. Most ways the +pipeline can break upstream of here still leave profraw files behind: an +uninstrumented library writes profiles with no counters, and a process that +crashes before its exit handler writes empty ones. +""" + +import argparse +import logging +import re +import shutil +import subprocess +import sys +from pathlib import Path + +logging.basicConfig(level=logging.INFO, format="%(message)s") + +EXECUTABLE_SUFFIX = ".exe" if sys.platform == "win32" else "" + + +def find_llvm_tool(llvm_bin_dir: Path, tool_name: str) -> Path: + """Locates an LLVM tool, preferring the ROCm distribution's own copy.""" + candidate = llvm_bin_dir / f"{tool_name}{EXECUTABLE_SUFFIX}" + if candidate.is_file(): + return candidate + + fallback = shutil.which(tool_name) + if fallback: + logging.warning( + "%s not found in %s, falling back to %s. A version mismatch with " + "the compiler that produced the profiles may cause failures.", + tool_name, + llvm_bin_dir, + fallback, + ) + return Path(fallback) + + raise FileNotFoundError( + f"Could not find '{tool_name}' in '{llvm_bin_dir}' or on PATH." + ) + + +def find_profraw_files(profraw_dir: Path) -> list[Path]: + return sorted(profraw_dir.rglob("*.profraw")) + + +def resolve_objects(rocm_dir: Path, object_globs: list[str]) -> list[Path]: + """Expands the per-project object globs into concrete instrumented files. + + Globs typically match a versioned family of symlinks (libfoo.so, + libfoo.so.1, ...) that all resolve to one file, so results are deduplicated + by real path to avoid handing llvm-cov the same object several times. + + A glob starting with '!' takes its matches back out. Header-only projects + report against their test binaries, and those can share a directory with + a sibling project's (rocPRIM's bin/test_* also matches hipCUB's tests). + """ + excluded = { + match + for pattern in object_globs + if pattern.startswith("!") + for match in rocm_dir.glob(pattern[1:]) + } + objects: dict[Path, Path] = {} + for pattern in object_globs: + if pattern.startswith("!"): + continue + matches = sorted(m for m in rocm_dir.glob(pattern) if m not in excluded) + if not matches: + logging.warning( + "No files matched object glob '%s' under %s", pattern, rocm_dir + ) + for match in matches: + if match.is_dir(): + continue + objects.setdefault(match.resolve(), match) + return sorted(objects.values()) + + +def merge_profraw(llvm_profdata: Path, profraw_files: list[Path], output: Path) -> None: + output.parent.mkdir(parents=True, exist_ok=True) + command = [ + str(llvm_profdata), + "merge", + "-sparse", + "-o", + str(output), + *[str(f) for f in profraw_files], + ] + logging.info("Merging %d profraw file(s) into %s", len(profraw_files), output) + subprocess.run(command, check=True) + + +def count_profiled_functions(llvm_profdata: Path, profdata: Path) -> int | None: + """Returns how many functions the merged profile holds counters for. + + None if llvm-profdata's summary has no such line, so an unexpected format + skips the check rather than failing a good report. + """ + result = subprocess.run( + [str(llvm_profdata), "show", str(profdata)], + check=True, + capture_output=True, + text=True, + ) + match = re.search(r"^Total functions: (\d+)$", result.stdout, re.MULTILINE) + return int(match.group(1)) if match else None + + +def strip_lcov_prefix(lcov: Path, prefix: str) -> None: + """Rewrites the lcov's SF: paths to drop a build-time prefix. + + llvm-cov records absolute paths from the build container, but Codecov + matches a report against the repository tree and needs repo-relative paths. + Its -path-equivalence can't map a prefix to an empty string (it rejects a + missing 'to'), so the leading repo root is stripped here instead. + """ + lines = lcov.read_text().splitlines(keepends=True) + rewritten = [] + for line in lines: + if line.startswith("SF:") and line[3:].startswith(prefix): + line = "SF:" + line[3 + len(prefix) :] + rewritten.append(line) + lcov.write_text("".join(rewritten)) + + +def count_lcov_lines(lcov: Path) -> tuple[int, int]: + """Returns (lines found, lines hit) summed over every file in an lcov report.""" + found = hit = 0 + with open(lcov) as lcov_file: + for line in lcov_file: + if line.startswith("LF:"): + found += int(line[3:]) + elif line.startswith("LH:"): + hit += int(line[3:]) + return found, hit + + +def build_cov_command( + llvm_cov: Path, + subcommand: str, + profdata: Path, + objects: list[Path], + path_equivalence: str | None, +) -> list[str]: + """Assembles the part of an llvm-cov command line the three outputs share.""" + # llvm-cov takes the first object positionally and the rest via -object. + command = [str(llvm_cov), subcommand, str(objects[0])] + for obj in objects[1:]: + command.extend(["-object", str(obj)]) + command.append(f"-instr-profile={profdata}") + if path_equivalence: + command.append(f"-path-equivalence={path_equivalence}") + return command + + +def export_lcov( + llvm_cov: Path, + profdata: Path, + objects: list[Path], + output: Path, + path_equivalence: str | None = None, + ignore_filename_regex: str | None = None, +) -> None: + command = build_cov_command(llvm_cov, "export", profdata, objects, path_equivalence) + command.append("--format=lcov") + if ignore_filename_regex: + command.append(f"-ignore-filename-regex={ignore_filename_regex}") + + output.parent.mkdir(parents=True, exist_ok=True) + logging.info("Exporting lcov for %d object(s) to %s", len(objects), output) + with open(output, "w") as lcov_file: + subprocess.run(command, check=True, stdout=lcov_file) + + +def write_summary( + llvm_cov: Path, + profdata: Path, + objects: list[Path], + output: Path, + path_equivalence: str | None = None, +) -> None: + """Writes the per-file coverage table, and echoes the totals to the log. + + This reads nothing but the coverage mappings, so it is the one rendering + that is complete even when the sources are not on the machine. + """ + command = build_cov_command(llvm_cov, "report", profdata, objects, path_equivalence) + + output.parent.mkdir(parents=True, exist_ok=True) + logging.info("Writing coverage summary to %s", output) + result = subprocess.run(command, check=True, capture_output=True, text=True) + output.write_text(result.stdout) + + for line in result.stdout.splitlines(): + if line.startswith("TOTAL"): + logging.info("%s", line) + + +def write_html( + llvm_cov: Path, + profdata: Path, + objects: list[Path], + output_dir: Path, + path_equivalence: str | None = None, + project_title: str | None = None, + ignore_filename_regex: str | None = None, +) -> None: + """Renders the browsable report into output_dir, entry point index.html. + + Unlike the other two this opens the source files, so anything not on the + machine is reported by llvm-cov as uncovered rather than annotated. Files + generated into the build tree are never in a source checkout, so a few + misses here are expected; ignore_filename_regex drops them from this view so + they don't log an error per file. They stay in the lcov and text summary, + which count the mappings without opening the sources. + + Each file is shown once, its counts summed over every template + instantiation as in the lcov. llvm-cov's default also repeats a header's + source for each instantiation, which made rocWMMA's report 3.3GB with + single pages near 1GB. Without those views no function name is printed, + so the report skips demangling, which llvm-cov does for every name up + front and which cost rocWMMA 3GB of the 16GB a hosted runner has. + """ + command = build_cov_command(llvm_cov, "show", profdata, objects, path_equivalence) + command.extend( + [ + "--format=html", + f"-output-dir={output_dir}", + "-show-instantiations=false", + # Every file rendered in parallel holds its own coverage on top of + # the mappings of all the objects, which are most of the memory. + "-num-threads=1", + ] + ) + if project_title: + command.append(f"--project-title={project_title}") + if ignore_filename_regex: + command.append(f"-ignore-filename-regex={ignore_filename_regex}") + + output_dir.mkdir(parents=True, exist_ok=True) + logging.info("Writing HTML report to %s", output_dir / "index.html") + subprocess.run(command, check=True) + + +def main(argv: list[str] | None = None) -> int: + parser = argparse.ArgumentParser(description=__doc__) + parser.add_argument( + "--profraw-dir", + type=Path, + required=True, + help="Directory searched recursively for .profraw files", + ) + parser.add_argument( + "--rocm-dir", + type=Path, + required=True, + help="Directory holding the installed ROCm artifacts under test", + ) + parser.add_argument( + "--object-globs", + type=str, + required=True, + help="Comma-separated globs, relative to --rocm-dir, matching the " + "instrumented binaries to report on (e.g. 'lib/libhiprand.so*')", + ) + parser.add_argument( + "--llvm-bin-dir", + type=Path, + default=None, + help="Directory holding llvm-profdata and llvm-cov " + "(default: /lib/llvm/bin)", + ) + parser.add_argument( + "--profdata-output", + type=Path, + default=Path("coverage-report/coverage.profdata"), + help="Path for the merged profdata index", + ) + parser.add_argument( + "--lcov-output", + type=Path, + default=Path("coverage-report/coverage.info"), + help="Path for the generated lcov report", + ) + parser.add_argument( + "--summary-output", + type=Path, + default=None, + help="Path for the plain text per-file coverage table. Needs no " + "sources, so it is complete wherever it runs", + ) + parser.add_argument( + "--html-output", + type=Path, + default=None, + help="Directory for the browsable HTML report, entry point " + "index.html. Renders source, so the sources must be present and " + "--path-equivalence set if they moved since the build", + ) + parser.add_argument( + "--path-equivalence", + type=str, + default=None, + help="',' remapping for source paths recorded at build " + "time, e.g. '/__w/TheRock/TheRock,${GITHUB_WORKSPACE}'. Used by the " + "HTML and text renderings, which need to reach the sources on disk", + ) + parser.add_argument( + "--lcov-strip-prefix", + type=str, + default=None, + help="Prefix to drop from every SF: path in the lcov export so the " + "emitted paths are repo-relative and Codecov can match them against " + "the repository tree, e.g. " + "'/__w/rocm-libraries/rocm-libraries/rocm-libraries/'. llvm-cov's " + "-path-equivalence can't map a prefix to an empty string, so the strip " + "is a post-process step on the lcov instead", + ) + parser.add_argument( + "--project-title", + type=str, + default=None, + help="Heading for the HTML report's index page", + ) + parser.add_argument( + "--ignore-filename-regex", + type=str, + default=None, + help="Regex of source paths to drop from the lcov export and the HTML " + "report, e.g. '.*/build/.*' to skip build-tree generated files that " + "aren't on the report machine and that Codecov can't map to the " + "repository. The text summary still counts them", + ) + parser.add_argument( + "--allow-empty", + action="store_true", + help="Exit successfully when no profraw files were collected, or " + "none of them hold any counters", + ) + args = parser.parse_args(argv) + + profraw_files = find_profraw_files(args.profraw_dir) + if not profraw_files: + message = f"No .profraw files found under {args.profraw_dir}" + if args.allow_empty: + logging.warning("%s, skipping report generation", message) + return 0 + logging.error( + "%s. The tests either never ran or the instrumented libraries were " + "not the ones loaded at runtime.", + message, + ) + return 1 + + llvm_bin_dir = args.llvm_bin_dir or (args.rocm_dir / "lib" / "llvm" / "bin") + llvm_profdata = find_llvm_tool(llvm_bin_dir, "llvm-profdata") + llvm_cov = find_llvm_tool(llvm_bin_dir, "llvm-cov") + + object_globs = [g.strip() for g in args.object_globs.split(",") if g.strip()] + objects = resolve_objects(args.rocm_dir, object_globs) + if not objects: + logging.error( + "None of the object globs (%s) matched anything under %s", + ", ".join(object_globs), + args.rocm_dir, + ) + return 1 + + merge_profraw(llvm_profdata, profraw_files, args.profdata_output) + if count_profiled_functions(llvm_profdata, args.profdata_output) == 0: + message = f"None of the {len(profraw_files)} profraw file(s) hold any counters" + if args.allow_empty: + logging.warning("%s, skipping report generation", message) + return 0 + logging.error( + "%s. The processes that wrote them loaded no instrumented code: " + "either the coverage flags never reached the compiler, or every " + "process died before writing its profile.", + message, + ) + return 1 + + try: + export_lcov( + llvm_cov, + args.profdata_output, + objects, + args.lcov_output, + None, + args.ignore_filename_regex, + ) + except subprocess.CalledProcessError: + logging.error( + "llvm-cov could not export the %d object(s). If it found no " + "coverage data, they carry no coverage mapping: the project's " + "coverage option did not take effect in its build.", + len(objects), + ) + return 1 + if args.lcov_strip_prefix: + strip_lcov_prefix(args.lcov_output, args.lcov_strip_prefix) + logging.info("Wrote coverage report to %s", args.lcov_output) + lines_found, lines_hit = count_lcov_lines(args.lcov_output) + + if args.summary_output: + write_summary( + llvm_cov, + args.profdata_output, + objects, + args.summary_output, + args.path_equivalence, + ) + + if args.html_output: + write_html( + llvm_cov, + args.profdata_output, + objects, + args.html_output, + args.path_equivalence, + args.project_title, + args.ignore_filename_regex, + ) + + # Checked after the other renderings so the report still shows what was + # measured. + if lines_found and not lines_hit: + logging.error( + "None of the %d instrumented line(s) in the reported objects ran. " + "The counters came from other binaries: the tests loaded a " + "different copy of the project, or never reached it.", + lines_found, + ) + return 1 + return 0 + + +if __name__ == "__main__": + sys.exit(main()) diff --git a/build_tools/github_actions/tests/configure_coverage_ci_test.py b/build_tools/github_actions/tests/configure_coverage_ci_test.py new file mode 100644 index 00000000000..014010fbf23 --- /dev/null +++ b/build_tools/github_actions/tests/configure_coverage_ci_test.py @@ -0,0 +1,534 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +import contextlib +import io +import json +import os +import sys +import tempfile +import unittest +from pathlib import Path +from unittest import mock + +# Add repo root to PYTHONPATH +sys.path.insert(0, os.fspath(Path(__file__).parent.parent)) +sys.path.insert(0, os.fspath(Path(__file__).parents[2])) + +import configure_coverage_ci +import merge_coverage_report +from _therock_utils.build_topology import get_topology + + +class ParseProjectsTest(unittest.TestCase): + def test_empty_selects_every_buildable_project(self): + # Registered-but-unmeasurable projects are deliberately left out: an + # empty selection should not schedule a job that cannot report. Nor + # should it pick projects whose stage this workflow cannot build. + self.assertEqual( + configure_coverage_ci.parse_projects(""), + sorted(configure_coverage_ci.DEFAULT_PROJECTS), + ) + + def test_default_selection_is_one_the_workflow_can_actually_build(self): + configure_coverage_ci.resolve_build_stages( + configure_coverage_ci.parse_projects("") + ) + + def test_default_leaves_out_what_cannot_be_built_or_is_known_broken(self): + self.assertEqual( + sorted( + configure_coverage_ci.SUPPORTED_PROJECTS + - configure_coverage_ci.DEFAULT_PROJECTS + ), + # rccl and rocshmem for want of a comm-libs build job, hipblaslt + # and hiptensor because their instrumented builds do not link. + ["hipblaslt", "hiptensor", "rccl", "rocshmem"], + ) + + def test_blocked_project_is_left_out_of_every_alias(self): + # An alias is the caller delegating the choice, so it must not hand + # back a project whose build is known to fail. + for alias in ("all", "rocm_libraries_all", "rocm_systems_all"): + with self.subTest(alias=alias): + self.assertFalse( + configure_coverage_ci.BLOCKED_PROJECTS + & configure_coverage_ci._GROUP_ALIASES[alias] + ) + + def test_blocked_project_is_still_honoured_when_named(self): + # Retesting is the only way to notice the block has been lifted. + for name in sorted(configure_coverage_ci.BLOCKED_PROJECTS): + with self.subTest(project=name): + self.assertEqual(configure_coverage_ci.parse_projects(name), [name]) + + def test_naming_a_blocked_project_warns_with_the_reason(self): + for name in sorted(configure_coverage_ci.BLOCKED_PROJECTS): + with self.subTest(project=name): + stderr = io.StringIO() + with contextlib.redirect_stderr(stderr): + configure_coverage_ci.parse_projects(name) + self.assertIn( + configure_coverage_ci.COVERAGE_PROJECTS[name].blocked_reason, + stderr.getvalue(), + ) + + def test_whitespace_and_case_are_normalized(self): + self.assertEqual(configure_coverage_ci.parse_projects(" HipRand "), ["hiprand"]) + + def test_duplicates_are_dropped_but_order_is_kept(self): + self.assertEqual( + configure_coverage_ci.parse_projects("hiprand,hiprand"), ["hiprand"] + ) + + def test_unknown_project_is_rejected(self): + with self.assertRaises(ValueError) as context: + configure_coverage_ci.parse_projects("hiprand,not-a-project") + self.assertIn("not-a-project", str(context.exception)) + + def test_all_alias_expands_to_every_measurable_project(self): + self.assertEqual( + configure_coverage_ci.parse_projects("all"), + sorted( + configure_coverage_ci.SUPPORTED_PROJECTS + - configure_coverage_ci.BLOCKED_PROJECTS + ), + ) + + def test_rocm_libraries_all_expands_to_that_group(self): + self.assertEqual( + configure_coverage_ci.parse_projects("rocm_libraries_all"), + sorted(configure_coverage_ci.ROCM_LIBRARIES_PROJECTS), + ) + + def test_group_aliases_are_case_insensitive(self): + self.assertEqual( + configure_coverage_ci.parse_projects(" ALL "), + configure_coverage_ci.parse_projects("all"), + ) + self.assertEqual( + configure_coverage_ci.parse_projects("Rocm_Libraries_All"), + configure_coverage_ci.parse_projects("rocm_libraries_all"), + ) + + def test_alias_and_explicit_name_overlap_is_deduped(self): + self.assertEqual( + configure_coverage_ci.parse_projects("rocm_libraries_all,hiprand"), + configure_coverage_ci.parse_projects("rocm_libraries_all"), + ) + + def test_empty_group_alias_is_rejected(self): + # Both groups are populated, so the empty case needs an empty group to + # exist at all. Selecting nothing is a caller mistake worth reporting + # rather than silently producing no jobs. + with mock.patch.dict( + configure_coverage_ci._GROUP_ALIASES, + {"rocm_systems_all": frozenset()}, + ): + with self.assertRaises(ValueError): + configure_coverage_ci.parse_projects("rocm_systems_all") + + def test_unknown_alias_like_token_is_rejected(self): + with self.assertRaises(ValueError): + configure_coverage_ci.parse_projects("rocm_everything_all") + + +class SourceRepoPartitionTest(unittest.TestCase): + def test_groups_partition_the_measurable_projects(self): + libraries = set(configure_coverage_ci.ROCM_LIBRARIES_PROJECTS) + systems = set(configure_coverage_ci.ROCM_SYSTEMS_PROJECTS) + self.assertTrue(libraries.isdisjoint(systems)) + self.assertEqual( + libraries | systems, + set( + configure_coverage_ci.SUPPORTED_PROJECTS + - configure_coverage_ci.BLOCKED_PROJECTS + ), + ) + + +class ParseAmdgpuFamiliesTest(unittest.TestCase): + def test_empty_falls_back_to_the_default_family(self): + self.assertEqual( + configure_coverage_ci.parse_amdgpu_families(""), + [configure_coverage_ci.DEFAULT_AMDGPU_FAMILIES], + ) + + def test_comma_separated_families_are_split(self): + self.assertEqual( + configure_coverage_ci.parse_amdgpu_families("gfx94X-dcgpu, gfx110X-all"), + ["gfx94X-dcgpu", "gfx110X-all"], + ) + + +class ParseConfigSourceTest(unittest.TestCase): + def test_repository_and_ref_are_split(self): + self.assertEqual( + configure_coverage_ci.parse_config_source("ROCm/rocm-libraries@develop"), + ("ROCm/rocm-libraries", "develop"), + ) + + def test_empty_falls_back_to_the_default_source(self): + self.assertEqual( + configure_coverage_ci.parse_config_source(""), + ("ROCm/rocm-libraries", "main"), + ) + + def test_missing_ref_is_rejected(self): + with self.assertRaises(ValueError): + configure_coverage_ci.parse_config_source("ROCm/rocm-libraries") + + +class BuildCoverageMatrixTest(unittest.TestCase): + def test_one_entry_per_project_and_family(self): + matrix = configure_coverage_ci.build_coverage_matrix( + ["hiprand"], + ["gfx94X-dcgpu", "gfx110X-all"], + "ROCm/rocm-libraries", + "main", + ) + self.assertEqual(len(matrix), 2) + self.assertEqual( + [entry["amdgpu_families"] for entry in matrix], + ["gfx94X-dcgpu", "gfx110X-all"], + ) + + def test_entries_are_json_serializable(self): + matrix = configure_coverage_ci.build_coverage_matrix( + ["hiprand"], ["gfx94X-dcgpu"], "ROCm/rocm-libraries", "main" + ) + self.assertEqual(json.loads(json.dumps(matrix)), matrix) + + def test_every_registered_project_declares_report_inputs(self): + for name, project in configure_coverage_ci.COVERAGE_PROJECTS.items(): + with self.subTest(project=name): + self.assertTrue(project.object_globs, "needs objects for llvm-cov") + self.assertTrue(project.fetch_artifact_args, "needs artifacts to fetch") + self.assertTrue(project.stage) + self.assertTrue(project.test_component) + + def test_every_measurable_project_names_an_upstream_option(self): + # TheRock's _ENABLE_COVERAGE is only a knob; instrumentation + # happens when the subproject's own option is set. An entry without one + # builds and tests cleanly and then reports nothing. + for name in sorted(configure_coverage_ci.SUPPORTED_PROJECTS): + with self.subTest(project=name): + self.assertTrue( + configure_coverage_ci.COVERAGE_PROJECTS[name].coverage_option + ) + + def test_unsupported_project_is_rejected_with_the_reason(self): + with self.assertRaises(ValueError) as context: + configure_coverage_ci.parse_projects("miopen") + self.assertIn("not supported", str(context.exception)) + self.assertIn("no coverage option", str(context.exception)) + + def test_unsupported_projects_stay_out_of_the_group_aliases(self): + # They remain registered so the gap is tracked, but must not be + # scheduled by a group selection. + for name in ("miopen", "hipsparse", "rocalution", "rocprofiler-sdk"): + with self.subTest(project=name): + self.assertIn(name, configure_coverage_ci.COVERAGE_PROJECTS) + self.assertNotIn(name, configure_coverage_ci.parse_projects("all")) + + def test_object_globs_reach_the_matrix(self): + # object_globs is what scopes a report to one project, so it has to + # survive into the matrix in the spelling llvm-cov is handed. + (entry,) = configure_coverage_ci.build_coverage_matrix( + ["hiprand"], ["gfx94X-dcgpu"], "ROCm/rocm-libraries", "main" + ) + self.assertEqual(entry["object_globs"], "lib/libhiprand.so*") + self.assertEqual(entry["fetch_artifact_args"], "--rand") + + def test_header_only_projects_report_on_their_own_test_binaries(self): + # rocPRIM, hipCUB and rocThrust install their tests into one flat bin/, + # and --prim --tests brings all three into every one of their reports. + installed = { + "rocprim": ["bin/test_basic", "bin/test_rocprim_tuple"], + "hipcub": ["bin/test_hipcub_basic"], + "rocthrust": ["bin/vector.hip"], + } + with tempfile.TemporaryDirectory() as tmp: + root = Path(tmp) + (root / "bin").mkdir() + for names in installed.values(): + for name in names: + (root / name).touch() + for project, names in installed.items(): + with self.subTest(project=project): + objects = merge_coverage_report.resolve_objects( + root, + configure_coverage_ci.COVERAGE_PROJECTS[project].object_globs, + ) + self.assertEqual( + [o.relative_to(root).as_posix() for o in objects], sorted(names) + ) + + +class BuildCoverageCmakeOptionsTest(unittest.TestCase): + def test_full_selection_collapses_to_the_group_option(self): + # What the nightly does: instrument every onboarded project at once. + self.assertEqual( + configure_coverage_ci.build_coverage_cmake_options( + sorted(configure_coverage_ci.ROCM_LIBRARIES_PROJECTS) + ), + [ + "-DTHEROCK_COVERAGE_ROCM_LIBRARIES_ALL=ON", + "-DTHEROCK_FLAG_BUILD_HIPCCL=OFF", + ], + ) + + def test_both_groups_collapse_to_the_combined_option(self): + self.assertEqual( + configure_coverage_ci.build_coverage_cmake_options( + sorted( + configure_coverage_ci.SUPPORTED_PROJECTS + - configure_coverage_ci.BLOCKED_PROJECTS + ) + ), + ["-DTHEROCK_COVERAGE_ALL=ON", "-DTHEROCK_FLAG_BUILD_HIPCCL=OFF"], + ) + + def test_blocked_project_is_named_rather_than_folded_into_a_group(self): + # It is outside the group lists, so the group flag would not reach it + # and the explicit ask would silently build uninstrumented. + self.assertEqual( + configure_coverage_ci.build_coverage_cmake_options(["hipblaslt"]), + ["-DHIPBLASLT_ENABLE_COVERAGE=ON"], + ) + + def test_partial_selection_names_each_project_in_upper_case(self): + # therock_subproject.cmake only forwards the upper case spelling. + self.assertEqual( + configure_coverage_ci.build_coverage_cmake_options(["rocblas"]), + ["-DROCBLAS_ENABLE_COVERAGE=ON"], + ) + + def test_partial_selection_also_names_extra_targets(self): + # rocPRIM reports against binaries built by rocPRIM_tests, so naming + # rocPRIM alone would instrument nothing the report reads. + self.assertEqual( + configure_coverage_ci.build_coverage_cmake_options(["rocprim"]), + [ + "-DROCPRIM_ENABLE_COVERAGE=ON", + "-DROCPRIM_TESTS_ENABLE_COVERAGE=ON", + "-DTHEROCK_FLAG_BUILD_HIPCCL=OFF", + ], + ) + + def test_prim_projects_build_the_legacy_subprojects(self): + # hipCCL has no hipCUB subproject for HIPCUB_ENABLE_COVERAGE to reach. + for key in sorted(configure_coverage_ci.PRIM_PROJECTS): + with self.subTest(key=key): + self.assertIn( + "-DTHEROCK_FLAG_BUILD_HIPCCL=OFF", + configure_coverage_ci.build_coverage_cmake_options([key]), + ) + + def test_prim_projects_are_rocprim_hipcub_and_rocthrust(self): + self.assertEqual( + configure_coverage_ci.PRIM_PROJECTS, + {"rocprim", "hipcub", "rocthrust"}, + ) + + def test_one_group_does_not_claim_to_cover_the_other(self): + options = configure_coverage_ci.build_coverage_cmake_options( + sorted(configure_coverage_ci.ROCM_LIBRARIES_PROJECTS) + ) + self.assertNotIn("-DTHEROCK_COVERAGE_ALL=ON", options) + self.assertNotIn("-DTHEROCK_COVERAGE_ROCM_SYSTEMS_ALL=ON", options) + + def test_no_selection_produces_no_options(self): + self.assertEqual(configure_coverage_ci.build_coverage_cmake_options([]), []) + + +class ResolveBuildStagesTest(unittest.TestCase): + def test_math_libs_project_needs_the_math_libs_stage(self): + self.assertEqual( + configure_coverage_ci.resolve_build_stages(["hiprand"]), + {"math-libs"}, + ) + + def test_compiler_runtime_project_needs_no_math_libs_build(self): + # compiler-runtime is built unconditionally, so a project living there + # must not drag in the per-architecture math-libs job. + self.assertEqual( + configure_coverage_ci.resolve_build_stages(["amdsmi"]), + {"compiler-runtime"}, + ) + + def test_whole_rocm_libraries_group_stays_within_math_libs(self): + keys = configure_coverage_ci.parse_projects("rocm_libraries_all") + self.assertEqual( + configure_coverage_ci.resolve_build_stages(keys), {"math-libs"} + ) + + def test_comm_libs_project_is_rejected_with_a_actionable_message(self): + # comm-libs has no build job, so selecting rccl would leave the test + # job with nothing to overlay. It has to fail here instead. + with self.assertRaises(ValueError) as caught: + configure_coverage_ci.resolve_build_stages(["rccl"]) + message = str(caught.exception) + self.assertIn("rccl", message) + self.assertIn("comm-libs", message) + self.assertIn("multi_arch_ci_coverage_linux.yml", message) + + def test_rejection_names_every_unbuildable_project_not_just_the_first(self): + with self.assertRaises(ValueError) as caught: + configure_coverage_ci.resolve_build_stages(["hiprand", "rccl", "rocshmem"]) + message = str(caught.exception) + self.assertIn("rccl", message) + self.assertIn("rocshmem", message) + + def test_every_buildable_stage_is_one_a_project_can_register_against(self): + self.assertTrue( + configure_coverage_ci.BUILDABLE_STAGES <= configure_coverage_ci.KNOWN_STAGES + ) + + +class RegistryMatchesBuildTopologyTest(unittest.TestCase): + """The registry duplicates facts the build topology already knows. + + Nothing keeps the two in step at runtime, so a project moving between + stages upstream would otherwise show up as a stage that is never built and + an overlay that finds nothing, hours into a run. + """ + + def test_every_project_names_the_stage_that_builds_its_artifacts(self): + topology = get_topology() + for key, project in sorted(configure_coverage_ci.COVERAGE_PROJECTS.items()): + for artifact in project.artifact_names: + with self.subTest(project=key, artifact=artifact): + self.assertEqual( + topology.get_stage_for_artifact(artifact), + project.stage, + ) + + +class MainTest(unittest.TestCase): + def setUp(self): + self._orig_env = os.environ.copy() + + def tearDown(self): + os.environ.clear() + os.environ.update(self._orig_env) + + def test_writes_expected_outputs(self): + with tempfile.TemporaryDirectory() as tmp: + output_path = Path(tmp) / "github_output" + output_path.touch() + os.environ["GITHUB_OUTPUT"] = os.fspath(output_path) + os.environ["PROJECTS_TO_TEST"] = "hiprand" + os.environ["AMDGPU_FAMILIES"] = "gfx94X-dcgpu" + os.environ["COVERAGE_CONFIG_SOURCE"] = "ROCm/rocm-libraries@main" + + self.assertEqual(configure_coverage_ci.main([]), 0) + + written = output_path.read_text() + self.assertIn("coverage_matrix", written) + self.assertIn("dist_amdgpu_families", written) + self.assertIn("families_matrix_json", written) + self.assertIn("coverage_cmake_options", written) + # One project out of a populated group names that project alone. + self.assertIn("-DHIPRAND_ENABLE_COVERAGE=ON", written) + # The per-project test and overlay inputs the nightly reads back. + self.assertIn('"test_component": "hiprand"', written) + self.assertIn('"object_globs": "lib/libhiprand.so*"', written) + # hipRAND is a math-libs project, so that stage has to be built. + self.assertIn("needs_math_libs=true", written) + + +class EmitCmakeTest(unittest.TestCase): + def test_emits_group_lists_from_allowlist(self): + with tempfile.TemporaryDirectory() as tmp: + out = Path(tmp) / "therock_coverage_projects.cmake" + self.assertEqual(configure_coverage_ci.main(["--emit-cmake", str(out)]), 0) + text = out.read_text() + self.assertIn("set(THEROCK_COVERAGE_ROCM_LIBRARIES_PROJECTS", text) + self.assertIn("hipRAND", text) + self.assertIn("set(THEROCK_COVERAGE_ROCM_SYSTEMS_PROJECTS", text) + self.assertIn("rccl", text) + # A group flag has to reach the sibling subproject that builds a + # header-only project's tests, or the group instruments rocPRIM and + # leaves its only reportable binaries without a coverage mapping. + self.assertIn("rocPRIM_tests", text) + # Group membership follows source_repo, not the stage. + libraries_line = next( + line + for line in text.splitlines() + if line.startswith("set(THEROCK_COVERAGE_ROCM_LIBRARIES_PROJECTS") + ) + self.assertNotIn("rccl", libraries_line) + # Unmeasurable projects are excluded, or a group build would set a + # flag their CMake does not implement. + self.assertNotIn("MIOpen", text) + self.assertNotIn("hipSPARSE ", text) + # The CMake group lists and the Python group sets decide the same + # thing in two places, so a blocked project has to leave both or a + # group build sets the flag the Python side just declined to. + self.assertNotIn("hipBLASLt", libraries_line) + + def test_blocked_project_keeps_its_option_mapping(self): + # Excluded from the group lists but still nameable, which needs the + # translation to its upstream option name to survive. + with tempfile.TemporaryDirectory() as tmp: + out = Path(tmp) / "therock_coverage_projects.cmake" + self.assertEqual(configure_coverage_ci.main(["--emit-cmake", str(out)]), 0) + self.assertIn( + "set(THEROCK_COVERAGE_OPTION_HIPBLASLT HIPBLASLT_ENABLE_COVERAGE)", + out.read_text(), + ) + + def test_emits_the_upstream_option_name_per_subproject(self): + # The generic names are why this is a per-subproject map: setting + # BUILD_CODE_COVERAGE globally would instrument everything that happens + # to understand it. + with tempfile.TemporaryDirectory() as tmp: + out = Path(tmp) / "therock_coverage_projects.cmake" + self.assertEqual(configure_coverage_ci.main(["--emit-cmake", str(out)]), 0) + text = out.read_text() + self.assertIn( + "set(THEROCK_COVERAGE_OPTION_HIPRAND BUILD_CODE_COVERAGE)", text + ) + self.assertIn("set(THEROCK_COVERAGE_OPTION_ROCRAND CODE_COVERAGE)", text) + self.assertIn( + "set(THEROCK_COVERAGE_OPTION_RCCL ENABLE_CODE_COVERAGE)", text + ) + self.assertIn( + "set(THEROCK_COVERAGE_OPTION_HIPBLASLT HIPBLASLT_ENABLE_COVERAGE)", text + ) + # A sibling subproject builds from the same source, so it takes the + # same option as its parent. + self.assertIn( + "set(THEROCK_COVERAGE_OPTION_ROCPRIM_TESTS BUILD_CODE_COVERAGE)", text + ) + self.assertNotIn("THEROCK_COVERAGE_OPTION_MIOPEN", text) + + def test_emit_cmake_needs_no_env(self): + # Must work without PROJECTS_TO_TEST / AMDGPU_FAMILIES / GITHUB_OUTPUT set. + import os + + saved = { + k: os.environ.pop(k, None) + for k in ( + "PROJECTS_TO_TEST", + "AMDGPU_FAMILIES", + "COVERAGE_CONFIG_SOURCE", + "GITHUB_OUTPUT", + ) + } + try: + with tempfile.TemporaryDirectory() as tmp: + out = Path(tmp) / "c.cmake" + self.assertEqual( + configure_coverage_ci.main(["--emit-cmake", str(out)]), 0 + ) + self.assertTrue(out.exists()) + finally: + for k, v in saved.items(): + if v is not None: + os.environ[k] = v + + +if __name__ == "__main__": + unittest.main() diff --git a/build_tools/github_actions/tests/fetch_test_configurations_test.py b/build_tools/github_actions/tests/fetch_test_configurations_test.py index 5d1772d3efc..73b9601d966 100644 --- a/build_tools/github_actions/tests/fetch_test_configurations_test.py +++ b/build_tools/github_actions/tests/fetch_test_configurations_test.py @@ -284,6 +284,42 @@ def test_tensilelite_quick_omits_ctest_stage(self): ) self.assertEqual(tensilelite["timeout_minutes"], 15) + # ----------------------- + # Coverage variant timeouts + # ----------------------- + + def test_coverage_variant_scales_timeouts(self): + os.environ["PROJECTS_TO_TEST"] = "rocrand" + os.environ["BUILD_VARIANT"] = "coverage" + release = fetch_test_configurations.test_matrix["rocrand"]["timeout_minutes"] + + fetch_test_configurations.run() + + self.assertEqual( + self._get_components()[0]["timeout_minutes"], + fetch_test_configurations._coverage_timeout_minutes(release), + ) + + def test_release_variant_keeps_timeouts(self): + os.environ["PROJECTS_TO_TEST"] = "rocrand" + os.environ["BUILD_VARIANT"] = "release" + + fetch_test_configurations.run() + + self.assertEqual( + self._get_components()[0]["timeout_minutes"], + fetch_test_configurations.test_matrix["rocrand"]["timeout_minutes"], + ) + + def test_coverage_timeout_is_multiplied(self): + self.assertEqual(fetch_test_configurations._coverage_timeout_minutes(15), 60) + + def test_coverage_timeout_stays_inside_the_job_timeout(self): + self.assertEqual(fetch_test_configurations._coverage_timeout_minutes(90), 180) + + def test_coverage_timeout_never_shortens_a_longer_one(self): + self.assertEqual(fetch_test_configurations._coverage_timeout_minutes(288), 288) + # ----------------------- # tensilelite-common (Tensile/Tests/common on real hardware) # ----------------------- diff --git a/build_tools/github_actions/tests/merge_coverage_report_test.py b/build_tools/github_actions/tests/merge_coverage_report_test.py new file mode 100644 index 00000000000..a94284f818d --- /dev/null +++ b/build_tools/github_actions/tests/merge_coverage_report_test.py @@ -0,0 +1,433 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +import os +import platform +import subprocess +import sys +import tempfile +import unittest +from pathlib import Path +from unittest import mock + +# Add repo root to PYTHONPATH +sys.path.insert(0, os.fspath(Path(__file__).parent.parent)) + +import merge_coverage_report + + +def is_windows() -> bool: + return platform.system() == "Windows" + + +class TempDirTestBase(unittest.TestCase): + def setUp(self): + self._temp_dir = tempfile.TemporaryDirectory() + self.addCleanup(self._temp_dir.cleanup) + self.root = Path(self._temp_dir.name) + + def touch(self, relative_path: str) -> Path: + path = self.root / relative_path + path.parent.mkdir(parents=True, exist_ok=True) + path.touch() + return path + + +class FindLlvmToolTest(TempDirTestBase): + def test_prefers_the_rocm_distribution_copy(self): + bin_dir = self.root / "lib" / "llvm" / "bin" + bin_dir.mkdir(parents=True) + expected = bin_dir / f"llvm-cov{merge_coverage_report.EXECUTABLE_SUFFIX}" + expected.touch() + + with mock.patch("shutil.which", return_value="/usr/bin/llvm-cov"): + self.assertEqual( + merge_coverage_report.find_llvm_tool(bin_dir, "llvm-cov"), expected + ) + + def test_falls_back_to_path(self): + bin_dir = self.root / "empty" + bin_dir.mkdir() + + with mock.patch("shutil.which", return_value="/usr/bin/llvm-cov"): + self.assertEqual( + merge_coverage_report.find_llvm_tool(bin_dir, "llvm-cov"), + Path("/usr/bin/llvm-cov"), + ) + + def test_raises_when_the_tool_is_missing_everywhere(self): + bin_dir = self.root / "empty" + bin_dir.mkdir() + + with mock.patch("shutil.which", return_value=None): + with self.assertRaises(FileNotFoundError): + merge_coverage_report.find_llvm_tool(bin_dir, "llvm-cov") + + +class FindProfrawFilesTest(TempDirTestBase): + def test_searches_recursively_and_ignores_other_files(self): + self.touch("shard0/a-1234.profraw") + self.touch("shard1/nested/b-5678.profraw") + self.touch("shard1/test-output.xml") + + found = merge_coverage_report.find_profraw_files(self.root) + + self.assertEqual([f.name for f in found], ["a-1234.profraw", "b-5678.profraw"]) + + def test_empty_directory_yields_nothing(self): + self.assertEqual(merge_coverage_report.find_profraw_files(self.root), []) + + +class ResolveObjectsTest(TempDirTestBase): + @unittest.skipIf(is_windows(), "symlinks require elevated privileges on Windows") + def test_versioned_symlinks_collapse_to_one_object(self): + real = self.touch("lib/libhiprand.so.1.0") + (self.root / "lib" / "libhiprand.so").symlink_to(real) + (self.root / "lib" / "libhiprand.so.1").symlink_to(real) + + objects = merge_coverage_report.resolve_objects( + self.root, ["lib/libhiprand.so*"] + ) + + self.assertEqual(len(objects), 1) + self.assertEqual(objects[0].resolve(), real.resolve()) + + def test_multiple_globs_are_combined(self): + self.touch("lib/libhiprand.so") + self.touch("bin/hiprand_tool") + + objects = merge_coverage_report.resolve_objects( + self.root, ["lib/libhiprand.so*", "bin/hiprand_*"] + ) + + self.assertEqual( + sorted(o.name for o in objects), ["hiprand_tool", "libhiprand.so"] + ) + + def test_directories_are_not_treated_as_objects(self): + (self.root / "lib" / "libhiprand.so.d").mkdir(parents=True) + + self.assertEqual( + merge_coverage_report.resolve_objects(self.root, ["lib/libhiprand.so*"]), [] + ) + + def test_unmatched_glob_yields_nothing(self): + self.assertEqual( + merge_coverage_report.resolve_objects(self.root, ["lib/nothing*"]), [] + ) + + def test_bang_glob_takes_matches_back_out(self): + self.touch("bin/test_basic") + self.touch("bin/test_hipcub_basic") + + objects = merge_coverage_report.resolve_objects( + self.root, ["bin/test_*", "!bin/test_hipcub_*"] + ) + + self.assertEqual([o.name for o in objects], ["test_basic"]) + + +class CountProfiledFunctionsTest(unittest.TestCase): + def _count(self, stdout: str): + with mock.patch("subprocess.run", return_value=mock.Mock(stdout=stdout)) as run: + count = merge_coverage_report.count_profiled_functions( + Path("/llvm/llvm-profdata"), Path("/tmp/c.profdata") + ) + self.assertEqual(run.call_args.args[0][1], "show") + return count + + def test_reads_the_summary_line(self): + self.assertEqual( + self._count("Instrumentation level: Front-end\nTotal functions: 1305\n"), + 1305, + ) + + def test_profile_without_counters_reads_as_zero(self): + self.assertEqual(self._count("Total functions: 0\n"), 0) + + def test_unrecognized_output_skips_the_check(self): + self.assertIsNone(self._count("something else entirely\n")) + + +class CountLcovLinesTest(TempDirTestBase): + def test_sums_over_every_file(self): + lcov = self.root / "coverage.info" + lcov.write_text( + "SF:a.cpp\nDA:1,1\nLF:10\nLH:4\nend_of_record\n" + "SF:b.cpp\nDA:1,0\nLF:5\nLH:0\nend_of_record\n" + ) + + self.assertEqual(merge_coverage_report.count_lcov_lines(lcov), (15, 4)) + + +class CommandConstructionTest(TempDirTestBase): + def test_merge_passes_every_profraw_file(self): + profraw_files = [self.touch("a.profraw"), self.touch("b.profraw")] + output = self.root / "out" / "coverage.profdata" + llvm_profdata = Path("/llvm/llvm-profdata") + + with mock.patch("subprocess.run") as run: + merge_coverage_report.merge_profraw(llvm_profdata, profraw_files, output) + + command = run.call_args.args[0] + self.assertEqual(command[:3], [str(llvm_profdata), "merge", "-sparse"]) + self.assertEqual(command[-2:], [str(profraw_files[0]), str(profraw_files[1])]) + self.assertTrue(output.parent.is_dir()) + + def test_export_passes_the_first_object_positionally(self): + objects = [self.touch("lib/a.so"), self.touch("lib/b.so")] + output = self.root / "out" / "coverage.info" + llvm_cov = Path("/llvm/llvm-cov") + profdata = Path("/tmp/coverage.profdata") + + with mock.patch("subprocess.run") as run: + merge_coverage_report.export_lcov(llvm_cov, profdata, objects, output) + + command = run.call_args.args[0] + self.assertEqual(command[:3], [str(llvm_cov), "export", str(objects[0])]) + self.assertIn("-object", command) + self.assertEqual(command[-2:], [f"-instr-profile={profdata}", "--format=lcov"]) + + def test_summary_reports_and_captures_the_table(self): + objects = [self.touch("lib/a.so")] + output = self.root / "out" / "coverage_summary.txt" + table = "Filename Cover\nTOTAL 77.78%\n" + + with mock.patch("subprocess.run", return_value=mock.Mock(stdout=table)) as run: + merge_coverage_report.write_summary( + Path("/llvm/llvm-cov"), Path("/tmp/c.profdata"), objects, output + ) + + self.assertEqual(run.call_args.args[0][1], "report") + self.assertEqual(output.read_text(), table) + + def test_html_renders_into_a_directory(self): + objects = [self.touch("lib/a.so")] + output_dir = self.root / "out" / "html" + + with mock.patch("subprocess.run") as run: + merge_coverage_report.write_html( + Path("/llvm/llvm-cov"), + Path("/tmp/c.profdata"), + objects, + output_dir, + project_title="hiprand", + ) + + command = run.call_args.args[0] + self.assertEqual(command[1], "show") + self.assertIn("--format=html", command) + self.assertIn(f"-output-dir={output_dir}", command) + self.assertIn("--project-title=hiprand", command) + + def test_html_shows_each_file_once_and_one_at_a_time(self): + # Per-instantiation views, demangling, and files rendered in parallel + # made rocWMMA's report 3.3GB and its rendering outgrow a 16GB runner. + with mock.patch("subprocess.run") as run: + merge_coverage_report.write_html( + Path("/llvm/llvm-cov"), + Path("/tmp/c.profdata"), + [self.touch("bin/gemm_xdl-validate")], + self.root / "out" / "html", + ) + + command = run.call_args.args[0] + self.assertIn("-show-instantiations=false", command) + self.assertIn("-num-threads=1", command) + self.assertEqual([arg for arg in command if arg.startswith("-Xdemangler")], []) + + def test_path_equivalence_reaches_every_rendering(self): + objects = [self.touch("lib/a.so")] + remap = "/__w/TheRock/TheRock,/home/runner/work" + + for build in ( + lambda: merge_coverage_report.export_lcov( + Path("/llvm/llvm-cov"), + Path("/tmp/c.profdata"), + objects, + self.root / "out" / "coverage.info", + remap, + ), + lambda: merge_coverage_report.write_html( + Path("/llvm/llvm-cov"), + Path("/tmp/c.profdata"), + objects, + self.root / "out" / "html", + remap, + ), + ): + with mock.patch("subprocess.run") as run: + build() + self.assertIn(f"-path-equivalence={remap}", run.call_args.args[0]) + + +class MainTest(TempDirTestBase): + def _argv(self, *extra: str) -> list[str]: + return [ + "--profraw-dir", + os.fspath(self.root / "profraw"), + "--rocm-dir", + os.fspath(self.root / "rocm"), + "--object-globs", + "lib/libhiprand.so*", + *extra, + ] + + def test_missing_profiles_fail_by_default(self): + (self.root / "profraw").mkdir() + + self.assertEqual(merge_coverage_report.main(self._argv()), 1) + + def test_missing_profiles_are_tolerated_with_allow_empty(self): + (self.root / "profraw").mkdir() + + self.assertEqual(merge_coverage_report.main(self._argv("--allow-empty")), 0) + + def test_unmatched_object_globs_fail(self): + self.touch("profraw/a.profraw") + llvm_bin_dir = self.root / "rocm" / "lib" / "llvm" / "bin" + llvm_bin_dir.mkdir(parents=True) + for tool in ("llvm-profdata", "llvm-cov"): + (llvm_bin_dir / f"{tool}{merge_coverage_report.EXECUTABLE_SUFFIX}").touch() + + self.assertEqual(merge_coverage_report.main(self._argv()), 1) + + def test_happy_path_invokes_both_llvm_tools(self): + self.touch("profraw/shard0/a.profraw") + self.touch("rocm/lib/libhiprand.so") + llvm_bin_dir = self.root / "rocm" / "lib" / "llvm" / "bin" + llvm_bin_dir.mkdir(parents=True) + for tool in ("llvm-profdata", "llvm-cov"): + (llvm_bin_dir / f"{tool}{merge_coverage_report.EXECUTABLE_SUFFIX}").touch() + + with mock.patch( + "subprocess.run", return_value=mock.Mock(stdout="Total functions: 3\n") + ) as run: + exit_code = merge_coverage_report.main( + self._argv( + "--profdata-output", + os.fspath(self.root / "out" / "coverage.profdata"), + "--lcov-output", + os.fspath(self.root / "out" / "coverage.info"), + ) + ) + + self.assertEqual(exit_code, 0) + self.assertEqual(run.call_count, 3) + invoked = [Path(call.args[0][0]).name for call in run.call_args_list] + self.assertEqual( + invoked, + [ + f"llvm-profdata{merge_coverage_report.EXECUTABLE_SUFFIX}", + f"llvm-profdata{merge_coverage_report.EXECUTABLE_SUFFIX}", + f"llvm-cov{merge_coverage_report.EXECUTABLE_SUFFIX}", + ], + ) + + def test_summary_and_html_are_opt_in(self): + self.touch("profraw/shard0/a.profraw") + self.touch("rocm/lib/libhiprand.so") + llvm_bin_dir = self.root / "rocm" / "lib" / "llvm" / "bin" + llvm_bin_dir.mkdir(parents=True) + for tool in ("llvm-profdata", "llvm-cov"): + (llvm_bin_dir / f"{tool}{merge_coverage_report.EXECUTABLE_SUFFIX}").touch() + + with mock.patch("subprocess.run", return_value=mock.Mock(stdout="")) as run: + exit_code = merge_coverage_report.main( + self._argv( + "--profdata-output", + os.fspath(self.root / "out" / "coverage.profdata"), + "--lcov-output", + os.fspath(self.root / "out" / "coverage.info"), + "--summary-output", + os.fspath(self.root / "out" / "coverage_summary.txt"), + "--html-output", + os.fspath(self.root / "out" / "html"), + ) + ) + + self.assertEqual(exit_code, 0) + # merge and check the index, then export, report and show off it. + subcommands = [call.args[0][1] for call in run.call_args_list] + self.assertEqual(subcommands, ["merge", "show", "export", "report", "show"]) + + +class MainRejectsEmptyReportsTest(TempDirTestBase): + """Every way the pipeline breaks upstream still leaves profraw files behind.""" + + def setUp(self): + super().setUp() + self.touch("profraw/shard0/a.profraw") + self.touch("rocm/lib/libhiprand.so") + llvm_bin_dir = self.root / "rocm" / "lib" / "llvm" / "bin" + llvm_bin_dir.mkdir(parents=True) + for tool in ("llvm-profdata", "llvm-cov"): + (llvm_bin_dir / f"{tool}{merge_coverage_report.EXECUTABLE_SUFFIX}").touch() + + def _main( + self, *extra: str, functions: int = 3, lcov: str = "", export_fails=False + ): + def fake_run(command, **kwargs): + tool, subcommand = Path(command[0]).stem, command[1] + if (tool, subcommand) == ("llvm-profdata", "show"): + return mock.Mock(stdout=f"Total functions: {functions}\n") + if subcommand == "export": + if export_fails: + raise subprocess.CalledProcessError(1, command) + kwargs["stdout"].write(lcov) + return mock.Mock(stdout="") + + with mock.patch("subprocess.run", side_effect=fake_run) as run: + exit_code = merge_coverage_report.main( + [ + "--profraw-dir", + os.fspath(self.root / "profraw"), + "--rocm-dir", + os.fspath(self.root / "rocm"), + "--object-globs", + "lib/libhiprand.so*", + "--profdata-output", + os.fspath(self.root / "out" / "coverage.profdata"), + "--lcov-output", + os.fspath(self.root / "out" / "coverage.info"), + *extra, + ] + ) + return exit_code, [call.args[0][1] for call in run.call_args_list] + + def test_profiles_without_counters_fail_before_export(self): + # What an uninstrumented library writes: a header and nothing else. + exit_code, subcommands = self._main(functions=0) + + self.assertEqual(exit_code, 1) + self.assertEqual(subcommands, ["merge", "show"]) + + def test_profiles_without_counters_are_tolerated_with_allow_empty(self): + exit_code, _ = self._main("--allow-empty", functions=0) + + self.assertEqual(exit_code, 0) + + def test_objects_without_a_coverage_mapping_fail(self): + exit_code, _ = self._main(export_fails=True) + + self.assertEqual(exit_code, 1) + + def test_report_that_never_reaches_the_objects_fails(self): + exit_code, subcommands = self._main( + "--summary-output", + os.fspath(self.root / "out" / "coverage_summary.txt"), + lcov="SF:a.cpp\nLF:10\nLH:0\nend_of_record\n", + ) + + self.assertEqual(exit_code, 1) + # The summary is still written, so the job page shows the zero. + self.assertEqual(subcommands, ["merge", "show", "export", "report"]) + + def test_report_with_any_line_hit_passes(self): + exit_code, _ = self._main(lcov="SF:a.cpp\nLF:10\nLH:1\nend_of_record\n") + + self.assertEqual(exit_code, 0) + + +if __name__ == "__main__": + unittest.main() diff --git a/build_tools/github_actions/tests/wait_for_baseline_stage_test.py b/build_tools/github_actions/tests/wait_for_baseline_stage_test.py new file mode 100644 index 00000000000..83190c317d2 --- /dev/null +++ b/build_tools/github_actions/tests/wait_for_baseline_stage_test.py @@ -0,0 +1,358 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +import os +import sys +from pathlib import Path +from unittest.mock import call, patch + +import pytest + +# Add github_actions to path so wait_for_baseline_stage is importable. +sys.path.insert(0, os.fspath(Path(__file__).parent.parent)) + +from github_actions_api import GitHubAPIError +from wait_for_baseline_stage import ( + MAX_CONSECUTIVE_API_ERRORS, + MAX_PAGES, + STAGE_JOB_NAMES, + State, + evaluate, + expected_job_names, + list_run_jobs, + main, + parse_families, + wait_for_stage, +) +from workflow_utils import WORKFLOWS_DIR, load_workflow + +LINUX_STAGES = "Linux::release / Build Multi-Arch Stages" +WINDOWS_STAGES = "Windows::release / Build Multi-Arch Stages" +COVERAGE = "Linux::CodeCoverage" +GFX94X = "Stage - Math Libs (gfx94X-dcgpu)" +GFX950 = "Stage - Math Libs (gfx950-dcgpu)" + + +def job(name, status="completed", conclusion="success", job_id=1): + return { + "name": name, + "status": status, + "conclusion": conclusion if status == "completed" else None, + "html_url": f"https://github.com/ROCm/rocm-libraries/actions/runs/1/job/{job_id}", + } + + +def math_libs(family, prefix=LINUX_STAGES, **kwargs): + return job( + f"{prefix} / math-libs ({family}, linux-gfx942-1gpu) / Stage - Math Libs ({family})", + **kwargs, + ) + + +SETUP = job("setup / setup", job_id=100) +COMPILER_RUNTIME = job( + f"{LINUX_STAGES} / compiler-runtime / Stage - Compiler Runtime", job_id=101 +) + + +# --------------------------------------------------------------------------- +# Inputs +# --------------------------------------------------------------------------- + + +class TestParseFamilies: + def test_accepts_semicolons_and_commas(self): + assert parse_families("gfx94X-dcgpu; gfx950-dcgpu,gfx1151") == [ + "gfx94X-dcgpu", + "gfx950-dcgpu", + "gfx1151", + ] + + def test_empty(self): + assert parse_families(" ; ") == [] + + +class TestExpectedJobNames: + def test_math_libs_names_a_job_per_family(self): + assert expected_job_names("math-libs", ["gfx94X-dcgpu", "gfx950-dcgpu"]) == [ + GFX94X, + GFX950, + ] + + def test_compiler_runtime_is_one_job_for_every_family(self): + assert expected_job_names("compiler-runtime", ["gfx94X-dcgpu"]) == [ + "Stage - Compiler Runtime" + ] + + +# --------------------------------------------------------------------------- +# evaluate() +# --------------------------------------------------------------------------- + + +class TestEvaluate: + def test_ready_once_the_job_succeeds(self): + state, lines = evaluate( + [SETUP, COMPILER_RUNTIME, math_libs("gfx94X-dcgpu")], "math-libs", [GFX94X] + ) + assert state is State.READY + assert lines == [ + f"{GFX94X}: success https://github.com/ROCm/rocm-libraries/actions/runs/1/job/1" + ] + + def test_waits_while_the_job_runs(self): + jobs = [SETUP, math_libs("gfx94X-dcgpu", status="in_progress")] + state, lines = evaluate(jobs, "math-libs", [GFX94X]) + assert state is State.WAITING + assert lines[0].startswith(f"{GFX94X}: in_progress ") + + def test_waits_before_the_stage_starts(self): + state, lines = evaluate( + [SETUP, job(COMPILER_RUNTIME["name"], status="in_progress")], + "math-libs", + [GFX94X], + ) + assert state is State.WAITING + assert lines == [f"{GFX94X}: not started"] + + @pytest.mark.parametrize("conclusion", ["failure", "cancelled", "skipped"]) + def test_fails_when_the_job_does_not_succeed(self, conclusion): + jobs = [math_libs("gfx94X-dcgpu", conclusion=conclusion)] + state, lines = evaluate(jobs, "math-libs", [GFX94X]) + assert state is State.FAILED + assert lines[0].startswith(f"{GFX94X}: {conclusion} ") + + def test_ignores_windows_jobs_with_the_same_name(self): + windows_done = math_libs("gfx94X-dcgpu", prefix=WINDOWS_STAGES, job_id=2) + linux_running = math_libs("gfx94X-dcgpu", status="in_progress") + state, _ = evaluate([windows_done, linux_running], "math-libs", [GFX94X]) + assert state is State.WAITING + + windows_failed = math_libs( + "gfx94X-dcgpu", prefix=WINDOWS_STAGES, conclusion="failure", job_id=2 + ) + state, _ = evaluate( + [windows_failed, math_libs("gfx94X-dcgpu")], "math-libs", [GFX94X] + ) + assert state is State.READY + + @pytest.mark.parametrize( + "caller", + ["math-libs", "math-libs (gfx94X-dcgpu, linux-gfx942-1gpu)"], + ) + def test_fails_when_the_stage_did_not_run(self, caller): + skipped_call = job(f"{LINUX_STAGES} / {caller}", conclusion="skipped") + state, lines = evaluate( + [SETUP, job(COMPILER_RUNTIME["name"], conclusion="failure"), skipped_call], + "math-libs", + [GFX94X], + ) + assert state is State.FAILED + assert lines[0].startswith(f"{GFX94X}: will not run; the math-libs stage was") + + def test_fails_when_the_run_builds_only_other_families(self): + state, lines = evaluate( + [math_libs("gfx950-dcgpu", status="in_progress")], "math-libs", [GFX94X] + ) + assert state is State.FAILED + assert lines == [f"{GFX94X}: not built by this run, which builds {GFX950}"] + + def test_coverage_stage_jobs_are_not_the_baseline(self): + coverage_job = job( + f"{COVERAGE} / Build Instrumented Math Libs (gfx94X-dcgpu)" + " / Stage - Coverage Math Libs (gfx94X-dcgpu)", + job_id=3, + ) + state, lines = evaluate([coverage_job], "math-libs", [GFX94X]) + assert state is State.WAITING + assert lines == [f"{GFX94X}: not started"] + + def test_waits_for_every_family(self): + names = [GFX94X, GFX950] + partial = [ + math_libs("gfx94X-dcgpu"), + math_libs("gfx950-dcgpu", status="queued", job_id=2), + ] + assert evaluate(partial, "math-libs", names)[0] is State.WAITING + + done = [math_libs("gfx94X-dcgpu"), math_libs("gfx950-dcgpu", job_id=2)] + assert evaluate(done, "math-libs", names)[0] is State.READY + + def test_compiler_runtime_stage(self): + windows_running = job( + f"{WINDOWS_STAGES} / compiler-runtime / Stage - Compiler Runtime", + status="in_progress", + job_id=2, + ) + state, _ = evaluate( + [COMPILER_RUNTIME, windows_running], + "compiler-runtime", + ["Stage - Compiler Runtime"], + ) + assert state is State.READY + + +# --------------------------------------------------------------------------- +# list_run_jobs() +# --------------------------------------------------------------------------- + + +class TestListRunJobs: + def test_paginates(self): + page1 = {"jobs": [job(f"job {i}", job_id=i) for i in range(100)]} + page2 = {"jobs": [job("last", job_id=100)]} + with patch( + "wait_for_baseline_stage.gha_send_request", side_effect=[page1, page2] + ) as send: + jobs = list_run_jobs("ROCm/rocm-libraries", "123") + assert len(jobs) == 101 + urls = [c.args[0] for c in send.call_args_list] + assert urls == [ + "https://api.github.com/repos/ROCm/rocm-libraries/actions/runs/123/jobs" + f"?filter=latest&per_page=100&page={page}" + for page in (1, 2) + ] + + def test_raises_past_the_page_limit(self): + full_page = {"jobs": [job(f"job {i}", job_id=i) for i in range(100)]} + with patch( + "wait_for_baseline_stage.gha_send_request", return_value=full_page + ) as send: + with pytest.raises(GitHubAPIError): + list_run_jobs("ROCm/rocm-libraries", "123") + assert send.call_count == MAX_PAGES + + +# --------------------------------------------------------------------------- +# wait_for_stage() +# --------------------------------------------------------------------------- + +WAITING = [math_libs("gfx94X-dcgpu", status="in_progress")] +READY = [math_libs("gfx94X-dcgpu")] +FAILED = [math_libs("gfx94X-dcgpu", conclusion="failure")] + + +def run_wait(snapshots, monotonic=None, timeout_minutes=60): + clock = {"side_effect": monotonic} if monotonic else {"return_value": 0} + with ( + patch("wait_for_baseline_stage.list_run_jobs", side_effect=snapshots), + patch("wait_for_baseline_stage.time.monotonic", **clock), + patch("wait_for_baseline_stage.time.sleep") as sleep, + ): + result = wait_for_stage( + "ROCm/rocm-libraries", + "123", + "math-libs", + [GFX94X], + poll_seconds=180, + timeout_minutes=timeout_minutes, + ) + return result, sleep + + +class TestWaitForStage: + def test_returns_zero_once_ready(self, capsys): + result, sleep = run_wait([WAITING, WAITING, READY]) + assert result == 0 + assert sleep.call_args_list == [call(180), call(180)] + # Only status changes are logged. + assert capsys.readouterr().out.count("Baseline math-libs") == 2 + + def test_fails_once_the_next_poll_confirms_a_failure(self, capsys): + result, sleep = run_wait([FAILED, FAILED]) + assert result == 1 + assert sleep.call_count == 1 + assert "::error::The baseline math-libs build did not succeed" in ( + capsys.readouterr().out + ) + + def test_unconfirmed_failure_keeps_waiting(self): + other_family_only = [math_libs("gfx950-dcgpu", status="in_progress")] + result, _ = run_wait([other_family_only, WAITING, READY]) + assert result == 0 + + def test_times_out(self, capsys): + # Deadline at 600s; the first check passes, the second is past it. + result, sleep = run_wait( + [WAITING, WAITING], monotonic=[0, 0, 601], timeout_minutes=10 + ) + assert result == 1 + assert sleep.call_count == 1 + assert "::error::Timed out after 10 minutes" in capsys.readouterr().out + + def test_gives_up_after_consecutive_api_errors(self): + errors = [GitHubAPIError("boom")] * MAX_CONSECUTIVE_API_ERRORS + result, sleep = run_wait(errors) + assert result == 1 + assert sleep.call_count == MAX_CONSECUTIVE_API_ERRORS - 1 + + def test_recovers_from_api_errors(self): + result, _ = run_wait([GitHubAPIError("boom"), READY]) + assert result == 0 + + +# --------------------------------------------------------------------------- +# main() +# --------------------------------------------------------------------------- + + +class TestMain: + def test_math_libs_needs_families(self): + with pytest.raises(SystemExit) as exc: + main( + ["--stage=math-libs", "--repository=ROCm/rocm-libraries", "--run-id=1"] + ) + assert exc.value.code == 2 + + def test_needs_repository_and_run_id(self, monkeypatch): + monkeypatch.delenv("GITHUB_REPOSITORY", raising=False) + monkeypatch.delenv("GITHUB_RUN_ID", raising=False) + with pytest.raises(SystemExit) as exc: + main(["--stage=compiler-runtime"]) + assert exc.value.code == 2 + + def test_waits_for_the_named_jobs(self, monkeypatch): + monkeypatch.setenv("GITHUB_REPOSITORY", "ROCm/rocm-libraries") + monkeypatch.setenv("GITHUB_RUN_ID", "123") + with patch( + "wait_for_baseline_stage.wait_for_stage", return_value=0 + ) as wait_mock: + assert ( + main( + ["--stage=math-libs", "--amdgpu-families=gfx94X-dcgpu;gfx950-dcgpu"] + ) + == 0 + ) + wait_mock.assert_called_once_with( + "ROCm/rocm-libraries", "123", "math-libs", [GFX94X, GFX950], 180, 300 + ) + + +# --------------------------------------------------------------------------- +# Workflow drift +# --------------------------------------------------------------------------- + + +class TestStageJobNames: + def test_match_the_regular_linux_build(self): + jobs = load_workflow(WORKFLOWS_DIR / "multi_arch_build_portable_linux.yml")[ + "jobs" + ] + for stage, template in STAGE_JOB_NAMES.items(): + stage_inputs = jobs[stage]["with"] + assert stage_inputs["stage_name"] == stage + assert stage_inputs["stage_display_name"] == template.format( + amdgpu_family="${{ matrix.family_info.amdgpu_family }}" + ) + + def test_coverage_stage_jobs_are_named_apart(self): + jobs = load_workflow(WORKFLOWS_DIR / "multi_arch_ci_coverage_linux.yml")["jobs"] + prefixes = [template.split("{")[0] for template in STAGE_JOB_NAMES.values()] + coverage_names = [ + job_def["with"]["stage_display_name"] + for job_def in jobs.values() + if "stage_display_name" in job_def.get("with", {}) + ] + assert coverage_names + for name in coverage_names: + assert not any(name.startswith(prefix) for prefix in prefixes), name diff --git a/build_tools/github_actions/wait_for_baseline_stage.py b/build_tools/github_actions/wait_for_baseline_stage.py new file mode 100644 index 00000000000..88fa3f0c7cc --- /dev/null +++ b/build_tools/github_actions/wait_for_baseline_stage.py @@ -0,0 +1,264 @@ +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +"""Waits for this run's regular Linux build to finish a stage. + +multi_arch_ci_coverage_linux.yml tests each instrumented project against a +non-instrumented baseline that it installs from S3. When the caller builds that +baseline in the same run (the rocm-libraries nightly does), coverage can start +alongside the regular build instead of after it, as long as its tests wait for +the baseline. A reusable workflow cannot `needs:` a job inside another one, so +this polls the run's jobs and exits 0 once the regular build's stage job has +succeeded for every requested family. + +It exits 1 once one of those jobs fails, is cancelled or is skipped, or once +the run shows the job will never start: the regular build skipped the stage +(an earlier stage failed, or the caller marked it prebuilt or skipped), or +built it only for other families. + +On a public repository, listing the jobs needs no `actions` permission. + +Usage: + + python build_tools/github_actions/wait_for_baseline_stage.py \\ + --stage=math-libs --amdgpu-families="gfx94X-dcgpu" +""" + +import argparse +from datetime import datetime, timezone +from enum import Enum +import os +import re +import sys +import time + +from github_actions_api import GitHubAPIError, gha_send_request + +# stage_display_name of each stage's job in multi_arch_build_portable_linux.yml, +# which keys every stage's job by the stage name. +STAGE_JOB_NAMES = { + "compiler-runtime": "Stage - Compiler Runtime", + "math-libs": "Stage - Math Libs ({amdgpu_family})", +} + +JOBS_PER_PAGE = 100 +MAX_PAGES = 20 +REQUEST_TIMEOUT_SECONDS = 60 +MAX_CONSECUTIVE_API_ERRORS = 5 + + +class State(Enum): + READY = "ready" + WAITING = "waiting" + FAILED = "failed" + + +def log(*args): + timestamp = datetime.now(timezone.utc).strftime("%H:%M:%SZ") + print(f"[{timestamp}]", *args, flush=True) + + +def parse_families(value: str) -> list[str]: + """Splits a comma- or semicolon-separated family list.""" + return [family.strip() for family in re.split(r"[,;]", value) if family.strip()] + + +def expected_job_names(stage: str, amdgpu_families: list[str]) -> list[str]: + template = STAGE_JOB_NAMES[stage] + if "{amdgpu_family}" not in template: + return [template] + return [template.format(amdgpu_family=family) for family in amdgpu_families] + + +def list_run_jobs(repository: str, run_id: str) -> list[dict]: + """Returns the latest attempt of every job in the run.""" + jobs: list[dict] = [] + for page in range(1, MAX_PAGES + 1): + url = ( + f"https://api.github.com/repos/{repository}/actions/runs/{run_id}/jobs" + f"?filter=latest&per_page={JOBS_PER_PAGE}&page={page}" + ) + response = gha_send_request(url, timeout_seconds=REQUEST_TIMEOUT_SECONDS) + page_jobs = response.get("jobs", []) + jobs.extend(page_jobs) + if len(page_jobs) < JOBS_PER_PAGE: + return jobs + raise GitHubAPIError(f"Run {run_id} has more than {len(jobs)} jobs") + + +def _name_parts(job: dict) -> list[str]: + # Jobs inside reusable workflows are named " / ... / ". + return job.get("name", "").split(" / ") + + +def _is_windows(job: dict) -> bool: + # Windows builds reuse the stage job names. multi_arch_ci.yml and the + # rocm-libraries nightly name the top-level job Windows::. + return _name_parts(job)[0].lower().startswith("windows") + + +def _describe(job: dict) -> str: + return f"{job.get('conclusion') or job.get('status')} {job.get('html_url', '')}" + + +def _did_not_succeed(job: dict) -> bool: + return job.get("status") == "completed" and job.get("conclusion") != "success" + + +def evaluate( + jobs: list[dict], stage: str, job_names: list[str] +) -> tuple[State, list[str]]: + """Returns whether every job in job_names succeeded, with a line per job.""" + jobs_by_leaf: dict[str, list[dict]] = {} + for job in jobs: + if not _is_windows(job): + jobs_by_leaf.setdefault(_name_parts(job)[-1], []).append(job) + + # A reusable-workflow call that did not run is listed as a single job named + # after the caller: the stage name, plus the matrix values once expanded. + stage_not_run = [ + job + for leaf, leaf_jobs in jobs_by_leaf.items() + if leaf == stage or leaf.startswith(f"{stage} (") + for job in leaf_jobs + if _did_not_succeed(job) + ] + template = STAGE_JOB_NAMES[stage] + family_prefix = template.split("{")[0] if "{" in template else None + built_names = sorted( + leaf + for leaf in jobs_by_leaf + if family_prefix and leaf.startswith(family_prefix) + ) + + lines = [] + failed = waiting = False + for name in job_names: + matches = jobs_by_leaf.get(name, []) + unsuccessful = [job for job in matches if _did_not_succeed(job)] + unfinished = [job for job in matches if job.get("status") != "completed"] + if unsuccessful: + failed = True + lines.extend(f"{name}: {_describe(job)}" for job in unsuccessful) + elif unfinished: + waiting = True + lines.extend(f"{name}: {_describe(job)}" for job in unfinished) + elif matches: + lines.extend(f"{name}: {_describe(job)}" for job in matches) + elif stage_not_run: + failed = True + lines.append( + f"{name}: will not run; the {stage} stage was " + f"{_describe(stage_not_run[0])}" + ) + elif built_names: + failed = True + lines.append( + f"{name}: not built by this run, which builds {', '.join(built_names)}" + ) + else: + waiting = True + lines.append(f"{name}: not started") + + if failed: + return State.FAILED, lines + if waiting: + return State.WAITING, lines + return State.READY, lines + + +def wait_for_stage( + repository: str, + run_id: str, + stage: str, + job_names: list[str], + poll_seconds: int, + timeout_minutes: int, +) -> int: + deadline = time.monotonic() + timeout_minutes * 60 + api_errors = 0 + last_lines: list[str] | None = None + unconfirmed_failure: list[str] | None = None + while True: + try: + jobs = list_run_jobs(repository, run_id) + except GitHubAPIError as e: + api_errors += 1 + log(f"Could not list jobs ({api_errors}/{MAX_CONSECUTIVE_API_ERRORS}): {e}") + if api_errors >= MAX_CONSECUTIVE_API_ERRORS: + print("::error::Gave up listing this run's jobs") + return 1 + else: + api_errors = 0 + state, lines = evaluate(jobs, stage, job_names) + if lines != last_lines: + log(f"Baseline {stage} {state.value}:") + for line in lines: + print(f" {line}", flush=True) + last_lines = lines + if state is State.READY: + return 0 + if state is State.FAILED: + # A snapshot spans several API pages, so act on a failure only + # once the next poll shows it too. + if lines == unconfirmed_failure: + print( + f"::error::The baseline {stage} build did not succeed;" + " coverage tests would run against missing artifacts" + ) + return 1 + unconfirmed_failure = lines + else: + unconfirmed_failure = None + + if time.monotonic() >= deadline: + print( + f"::error::Timed out after {timeout_minutes} minutes waiting for" + f" the baseline {stage} build" + ) + return 1 + time.sleep(poll_seconds) + + +def main(argv: list[str] | None = None) -> int: + parser = argparse.ArgumentParser( + description="Wait for this run's regular Linux build to finish a stage." + ) + parser.add_argument("--stage", required=True, choices=sorted(STAGE_JOB_NAMES)) + parser.add_argument( + "--amdgpu-families", + default="", + help="Comma- or semicolon-separated families; required for math-libs.", + ) + parser.add_argument( + "--repository", + default=os.environ.get("GITHUB_REPOSITORY", ""), + help="Repository in owner/repo format (default: $GITHUB_REPOSITORY).", + ) + parser.add_argument( + "--run-id", + default=os.environ.get("GITHUB_RUN_ID", ""), + help="Workflow run ID (default: $GITHUB_RUN_ID).", + ) + parser.add_argument("--poll-seconds", type=int, default=180) + parser.add_argument("--timeout-minutes", type=int, default=300) + args = parser.parse_args(argv) + + if not args.repository or not args.run_id: + parser.error("--repository and --run-id are required outside GitHub Actions") + job_names = expected_job_names(args.stage, parse_families(args.amdgpu_families)) + if not job_names: + parser.error(f"--stage={args.stage} needs --amdgpu-families") + + return wait_for_stage( + args.repository, + args.run_id, + args.stage, + job_names, + args.poll_seconds, + args.timeout_minutes, + ) + + +if __name__ == "__main__": + sys.exit(main()) diff --git a/build_tools/install_rocm_code_coverage_build.py b/build_tools/install_rocm_code_coverage_build.py index 2a34bec63b3..5931695fb17 100644 --- a/build_tools/install_rocm_code_coverage_build.py +++ b/build_tools/install_rocm_code_coverage_build.py @@ -16,11 +16,13 @@ [** all supported options of install_rocm_from_artifacts.py] """ + import os import re import sys import argparse import platform +from collections.abc import Iterable from pathlib import Path, PurePosixPath from install_rocm_from_artifacts import main as install_from_artifacts_main @@ -34,10 +36,10 @@ from _therock_utils.archive_util import open_archive_for_read from _therock_utils.artifacts import ArtifactName from _therock_utils.cmake_amdgpu_targets import amdgpu_family_map, expand_families +from _therock_utils.elf_phdr import pin_phdr_table -# Maps each --replace- flag (argparse dest) to the TheRock artifact name -# that ships the instrumented library. rocBLAS is packaged in the 'blas' -# artifact and rocSOLVER in the 'solver' artifact (see BUILD_TOPOLOGY.toml). +# Maps each --replace- flag to its TheRock artifact and library folder. +# e.g. rocBLAS ships in the 'blas' artifact (see BUILD_TOPOLOGY.toml). COMPONENT_MAP = { # component: (artifact_name, library_folder), "rocfft": ("fft", "rocFFT"), @@ -65,12 +67,67 @@ "composablekernel": ("composable-kernel", "composable_kernel"), } +# Backend library folders a hip* wrapper calls at runtime that ship in the same +# artifact. They are swapped along with the wrapper so both come from one +# build: hipRAND's linkage test compares hiprandGetVersion() (answered by +# librocrand) against the ROCRAND_VERSION it was compiled with, and fails when +# the baseline's rocRAND is a different version than the coverage build's. +SAME_ARTIFACT_BACKENDS = { + "hipblas": ("rocBLAS",), + "hipfft": ("rocFFT",), + "hiprand": ("rocRAND",), + "hipsolver": ("rocSOLVER",), + "hipsparse": ("rocSPARSE",), +} + + +def replacement_folders(components: Iterable[str]) -> dict[str, list[str]]: + """Map each artifact to swap to the library folders to extract from it. + + A selected component brings its own folder plus its SAME_ARTIFACT_BACKENDS. + """ + artifacts: dict[str, list[str]] = {} + for comp in components: + name, folder = COMPONENT_MAP[comp] + folders = artifacts.setdefault(name, []) + for f in (folder, *SAME_ARTIFACT_BACKENDS.get(comp, ())): + if f not in folders: + folders.append(f) + return artifacts + + +# Host libraries a component links that ship in another artifact, as +# (artifact, library folder, file name prefix under lib/). Only these files are +# swapped, from the coverage run's generic lib archive: the baseline's copies +# come from a different source commit, and their C++ interfaces need not match +# the instrumented component's. hipSPARSELt picks and launches its kernels +# through hipBLASLt's tensilelite-host and origami; against the baseline's it +# aborts with heap corruption or "No Tensile solution found". +OTHER_ARTIFACT_HOST_LIBS = { + "hipsparselt": ( + ("blas", "hipBLASLt", "libtensilelite-host"), + ("blas", "origami", "liborigami"), + ), +} + + +def host_library_deps( + components: Iterable[str], +) -> dict[str, list[tuple[str, str]]]: + """Map each artifact to the (library folder, file prefix) pairs to take from it.""" + deps: dict[str, list[tuple[str, str]]] = {} + for comp in components: + for artifact, folder, lib_prefix in OTHER_ARTIFACT_HOST_LIBS.get(comp, ()): + entries = deps.setdefault(artifact, []) + if (folder, lib_prefix) not in entries: + entries.append((folder, lib_prefix)) + return deps + def _read_passthrough_options(passthrough_argv): - """Read (without consuming) options shared with install_rocm_from_artifacts. + """Read options shared with install_rocm_from_artifacts without consuming them. - These are parsed non-destructively so the same argv can still be forwarded - to install_rocm_from_artifacts.py unchanged. + Parsed non-destructively so the same argv still forwards unchanged. """ reader = argparse.ArgumentParser(add_help=False) # --amdgpu-family and --artifact-group share a dest, mirroring install_rocm_from_artifacts. @@ -86,8 +143,8 @@ def _read_passthrough_options(passthrough_argv): def _target_families(family, amdgpu_targets): """Build the family match set: generic + family + expanded gfx targets. - blas/solver are target-specific artifacts named per family (mono-arch) or - per gfx target (kpack-split), so both spellings must be matched. + Target-specific artifacts (blas/solver) are named per family (mono-arch) or + per gfx target (kpack-split), so match both spellings. """ families = ["generic"] if family: @@ -97,16 +154,28 @@ def _target_families(family, amdgpu_targets): return families -def download_replacement_artifacts(code_coverage_run_id, artifact_names, opts): - """Download the instrumented replacement artifacts from the code-coverage run. - - Uses the code coverage run ID as the run-id for the S3 backend, then fetches - every component tar matching the requested artifact names and target family. +def download_replacement_artifacts( + code_coverage_run_id, + artifact_names, + opts, + code_coverage_run_github_repo, + code_coverage_release_type, + host_lib_artifacts=(), +): + """Download instrumented replacement artifacts from the code-coverage run. + + Fetches every component tar matching the requested artifact names and family, + plus only the generic lib tar of each host_lib_artifacts entry not already + fetched whole. + The coverage run lives in its own repo (--code-coverage-run-github-repo), + separate from the generic baseline's --run-github-repo. Pass it explicitly: + the S3 bucket lookup 404s if the run id is queried against the wrong repo. """ backend = create_backend_from_env( run_id=code_coverage_run_id, - github_repository=opts.run_github_repo, + github_repository=code_coverage_run_github_repo, platform=platform.system().lower(), + release_type=code_coverage_release_type, ) log(f"Fetching replacement artifacts from {backend.base_uri}") @@ -121,6 +190,22 @@ def download_replacement_artifacts(code_coverage_run_id, artifact_names, opts): f"{backend.base_uri} for families {target_families}" ) + host_lib_names = set(host_lib_artifacts) - set(artifact_names) + if host_lib_names: + log(f"Matching host library artifacts {sorted(host_lib_names)} (generic lib)") + host_matched = find_available_artifacts( + host_lib_names, + ["generic"], + available, + excluded_components={"dbg", "dev", "doc", "run", "test"}, + ) + if not host_matched: + raise IOError( + f"ERROR: No host library artifacts {sorted(host_lib_names)} found " + f"in {backend.base_uri}" + ) + matched += host_matched + dest_dir = opts.output_dir / "code-coverage-replacements" dest_dir.mkdir(parents=True, exist_ok=True) @@ -142,9 +227,8 @@ def download_replacement_artifacts(code_coverage_run_id, artifact_names, opts): def _replace_scoped_member(tf, member, dest_path, output_dir, relpaths): """Write a single archive member into the flattened install tree. - Mirrors the file/symlink/dir/hardlink handling used when TheRock flattens - an artifact archive, so replaced files keep their exec bits and link - structure. Any existing file/symlink at dest_path is removed first. + Mirrors TheRock's file/symlink/dir/hardlink flattening so replaced files + keep their exec bits and link structure. Removes any existing file first. """ if dest_path.is_symlink() or (dest_path.exists() and not dest_path.is_dir()): os.unlink(dest_path) @@ -163,8 +247,8 @@ def _replace_scoped_member(tf, member, dest_path, output_dir, relpaths): elif member.issym(): dest_path.symlink_to(member.linkname) elif member.islnk(): - # Hardlink target is archive-relative; strip its manifest prefix so it - # resolves to the already-written file in the flattened output tree. + # Hardlink target is archive-relative; strip the manifest prefix to + # resolve it against the already-written flattened file. for prefix in relpaths: prefix_slash = prefix + "/" if member.linkname.startswith(prefix_slash): @@ -179,13 +263,50 @@ def _replace_scoped_member(tf, member, dest_path, output_dir, relpaths): raise IOError(f"Unhandled tar member: {member}") +def _names_folder(folder, path): + """Whether `path` has a component naming `folder` (e.g. rocBLAS/, rocblas.so).""" + return re.search(rf"{folder}[^a-zA-Z]", path, flags=re.IGNORECASE) + + +def _is_own_stage(folder, prefix): + """Whether manifest relpath `prefix` is one of `folder`'s stage directories.""" + return bool( + re.search(rf"(^|/){folder}[^a-zA-Z]", prefix + "/", flags=re.IGNORECASE) + ) + + +def _is_component_path(folder, prefix, scoped_path, owns_archive): + """Whether a member under manifest relpath `prefix` belongs to `folder`. + + Everything under the component's own stage directories does, named after it + or not: the test binaries a header-only project's report measures are named + after what they test (rocThrust's bin/merge.hip, rocPRIM's + bin/test_block_scan). Elsewhere only a path naming the component does -- + except the device code under .kpack/. Every component in the artifact shares + one kpack archive, stored under whichever stage directory comes first + (prim_test's sits in hipCUB's), and host binaries look their kernels up in + it by name. Swapping the host binaries without it leaves them loading the + baseline's kernels, which fail once the two builds differ (a bumped + HIPCUB__NS mangles to symbols the baseline kpack does not have). + So it follows the host binaries: it is replaced whenever `owns_archive`, + i.e. the component has a stage directory in this artifact. + """ + if _names_folder(folder, scoped_path): + return True + if scoped_path == ".kpack" or scoped_path.startswith(".kpack/"): + return owns_archive + return _is_own_stage(folder, prefix) + + def replace_instrumented_libraries(artifacts, dest_dir, output_dir): """Extract instrumented libs from downloaded archives into the install tree. - For every replacement archive under dest_dir, read its artifact_manifest.txt - to learn the relpath prefixes, then flatten (strip prefix) each member into - output_dir -- but only members whose scoped path matches the artifact's - library folder (rocBLAS/rocSOLVER), so unrelated files are left in place. + Reads each archive's artifact_manifest.txt for relpath prefixes, then flattens + members into output_dir -- but only those belonging to the artifact's library + folder (e.g. rocBLAS), leaving unrelated files in place. Instrumented shared + libraries whose program header table kpack moved get it pinned back (see + _therock_utils/elf_phdr.py); otherwise their profile writer can segfault + at exit before writing any counters. """ archives = sorted( p for p in dest_dir.iterdir() if p.name.endswith((".tar.zst", ".tar.xz")) @@ -194,6 +315,7 @@ def replace_instrumented_libraries(artifacts, dest_dir, output_dir): log(f"No replacement archives found in {dest_dir}") return + written_files = [] for archive in archives: an = ArtifactName.from_filename(archive.name) folders = artifacts.get(an.name) if an else None @@ -204,17 +326,8 @@ def replace_instrumented_libraries(artifacts, dest_dir, output_dir): log(f"Replacing '{folder}' paths from {archive.name} into {output_dir}") replaced = 0 with open_archive_for_read(archive) as tf: - manifest_member = tf.next() - if ( - manifest_member is None - or manifest_member.name != "artifact_manifest.txt" - ): - raise IOError( - f"Artifact archive {archive} must have artifact_manifest.txt " - "as its first member" - ) - with tf.extractfile(manifest_member) as mf_file: - relpaths = [r for r in mf_file.read().decode().splitlines() if r] + relpaths = _read_manifest(tf, archive) + owns_archive = any(_is_own_stage(folder, r) for r in relpaths) while member := tf.next(): for prefix in relpaths: @@ -222,18 +335,89 @@ def replace_instrumented_libraries(artifacts, dest_dir, output_dir): if not member.name.startswith(prefix_slash): continue scoped_path = member.name[len(prefix_slash) :] - if not re.search( - rf"{folder}[^a-zA-Z]", scoped_path, flags=re.IGNORECASE + if not _is_component_path( + folder, prefix, scoped_path, owns_archive ): break dest_path = output_dir / PurePosixPath(scoped_path) _replace_scoped_member( tf, member, dest_path, output_dir, relpaths ) + if member.isfile(): + written_files.append(dest_path) replaced += 1 break log(f" Replaced {replaced} '{folder}' path(s) from {archive.name}") + _pin_instrumented(written_files) + + +def replace_host_libraries(host_libs, dest_dir, output_dir): + """Extract only the named host libraries from other artifacts' lib archives. + + host_libs maps an artifact to (library folder, file prefix) pairs, as built + by host_library_deps. A member is taken when it sits in that folder's stage + directory under lib/ and its file name starts with the prefix, so + libtensilelite-host.so* comes over without the rest of hipBLASLt. + """ + written_files = [] + for archive in sorted( + p for p in dest_dir.iterdir() if p.name.endswith((".tar.zst", ".tar.xz")) + ): + an = ArtifactName.from_filename(archive.name) + if not an or an.name not in host_libs: + continue + if an.component != "lib" or an.target_family != "generic": + continue + for folder, lib_prefix in host_libs[an.name]: + log(f"Replacing '{folder}' {lib_prefix}* from {archive.name}") + replaced = 0 + with open_archive_for_read(archive) as tf: + relpaths = _read_manifest(tf, archive) + while member := tf.next(): + for prefix in relpaths: + if not _is_own_stage(folder, prefix): + continue + prefix_slash = prefix + "/" + if not member.name.startswith(prefix_slash): + continue + scoped_path = member.name[len(prefix_slash) :] + if not scoped_path.startswith(f"lib/{lib_prefix}"): + break + dest_path = output_dir / PurePosixPath(scoped_path) + _replace_scoped_member( + tf, member, dest_path, output_dir, relpaths + ) + if member.isfile(): + written_files.append(dest_path) + replaced += 1 + break + if not replaced: + raise IOError( + f"ERROR: No {lib_prefix}* found under '{folder}' in {archive.name}" + ) + log(f" Replaced {replaced} path(s)") + + _pin_instrumented(written_files) + + +def _read_manifest(tf, archive): + """Return the relpath prefixes listed by the archive's leading manifest.""" + manifest_member = tf.next() + if manifest_member is None or manifest_member.name != "artifact_manifest.txt": + raise IOError( + f"Artifact archive {archive} must have artifact_manifest.txt " + "as its first member" + ) + with tf.extractfile(manifest_member) as mf_file: + return [r for r in mf_file.read().decode().splitlines() if r] + + +def _pin_instrumented(written_files): + for path in written_files: + if pin_phdr_table(path, require_section="__llvm_prf_cnts"): + log(f" Pinned the program header table kpack moved in {path}") + def main(argv): parser = argparse.ArgumentParser(prog="code-coverage-installer") @@ -242,6 +426,32 @@ def main(argv): type=str, help="run id of the build from which instrumental components needs to be replaced", ) + # Repo owning --code-coverage-run-id (the instrumented run), separate from + # --run-github-repo (the generic baseline). Defaults to $GITHUB_REPOSITORY. + # parse_known_args consumes it so it does not leak into the generic install. + parser.add_argument( + "--code-coverage-run-github-repo", + type=str, + default=os.environ.get("GITHUB_REPOSITORY"), + help=( + "GitHub repository (owner/name) that owns --code-coverage-run-id, " + "used to resolve the instrumented artifact backend. Defaults to " + "$GITHUB_REPOSITORY (the coverage run's own repo). This is separate " + "from --run-github-repo, which owns the generic --run-id build." + ), + ) + parser.add_argument( + "--code-coverage-release-type", + type=str, + default="ci", + help=( + "Release type of the instrumented build identified by " + "--code-coverage-run-id (e.g. 'ci', 'nightly'). Controls which S3 " + "bucket is used for the replacement artifact lookup. Defaults to " + "'ci' because instrumented builds always run as CI jobs. This is " + "separate from the baseline's RELEASE_TYPE env var." + ), + ) artifacts_group = parser.add_argument_group("replace_comps") for comp in COMPONENT_MAP.keys(): artifacts_group.add_argument( @@ -257,11 +467,9 @@ def main(argv): install_from_artifacts_main(extra_args) # collect artifacts details as per the user options - artifacts = {} - for comp, (name, folder) in COMPONENT_MAP.items(): - if not getattr(args, f"replace_{comp}"): - continue # skip for non-selected components - artifacts.setdefault(name, []).append(folder) + selected = [comp for comp in COMPONENT_MAP if getattr(args, f"replace_{comp}")] + artifacts = replacement_folders(selected) + host_libs = host_library_deps(selected) if not artifacts: log("No --replace-* components specified; nothing to download.") @@ -272,13 +480,20 @@ def main(argv): "--code-coverage-run-id is required when using --replace-* options" ) - # download selected component artifacts + # download selected component artifacts (instrumented, from the coverage repo) dest_dir = download_replacement_artifacts( - args.code_coverage_run_id, artifacts.keys(), opts + args.code_coverage_run_id, + artifacts.keys(), + opts, + args.code_coverage_run_github_repo, + args.code_coverage_release_type, + host_lib_artifacts=host_libs.keys(), ) # replace selected library folder paths in selected component artifacts replace_instrumented_libraries(artifacts, dest_dir, opts.output_dir) + # and the host libraries they link from other artifacts + replace_host_libraries(host_libs, dest_dir, opts.output_dir) if __name__ == "__main__": diff --git a/build_tools/tests/artifact_manager_tool_test.py b/build_tools/tests/artifact_manager_tool_test.py index 06477ca6969..c6ec389686b 100644 --- a/build_tools/tests/artifact_manager_tool_test.py +++ b/build_tools/tests/artifact_manager_tool_test.py @@ -720,6 +720,68 @@ def mock_extract(request: artifact_manager.ExtractRequest) -> Path: self.assertIn("test-artifact_lib_generic.tar.zst", fetched_keys) self.assertNotIn("second-artifact_lib_generic.tar.zst", fetched_keys) + def test_fetch_artifact_names_limits_to_named_artifacts(self) -> None: + """Test that --artifact-names narrows the fetch to an include list.""" + import artifact_manager + + self._create_staged_artifact("test-artifact", "lib", "generic") + self._create_staged_artifact("second-artifact", "lib", "generic") + + extract_calls: list[artifact_manager.ExtractRequest] = [] + + def mock_extract(request: artifact_manager.ExtractRequest) -> Path: + extract_calls.append(request) + return request.output_dir + + with mock.patch("artifact_manager.extract_artifact", mock_extract): + argv = [ + "fetch", + "--stage", + "all", + "--output-dir", + str(self.output_dir), + "--topology", + str(self.topology_path), + "--local-staging-dir", + str(self.staging_dir), + "--platform", + TEST_PLATFORM, + "--run-id", + "local", + "--artifact-names", + "second-artifact", + ] + + artifact_manager.main(argv) + + fetched_keys = [c.archive_path.name for c in extract_calls] + self.assertEqual(fetched_keys, ["second-artifact_lib_generic.tar.zst"]) + + def test_fetch_artifact_names_rejects_unknown_artifact(self) -> None: + """Test that --artifact-names validates artifact names.""" + import artifact_manager + + argv = [ + "fetch", + "--stage", + "all", + "--output-dir", + str(self.output_dir), + "--topology", + str(self.topology_path), + "--local-staging-dir", + str(self.staging_dir), + "--platform", + TEST_PLATFORM, + "--run-id", + "local", + "--artifact-names", + "unknown-artifact", + ] + + with self.assertRaises(ValueError): + artifact_manager.main(argv) + def test_fetch_exclude_artifacts_rejects_unknown_artifact(self) -> None: """Test that --exclude-artifacts validates artifact names.""" argv = [ diff --git a/build_tools/tests/elf_phdr_test.py b/build_tools/tests/elf_phdr_test.py new file mode 100644 index 00000000000..1726120f7df --- /dev/null +++ b/build_tools/tests/elf_phdr_test.py @@ -0,0 +1,247 @@ +#!/usr/bin/env python +# Copyright Advanced Micro Devices, Inc. +# SPDX-License-Identifier: MIT + +import collections +import struct +import tempfile +import unittest +from pathlib import Path + +from _therock_utils.elf_phdr import pin_phdr_table + +_EHDR = struct.Struct("<16sHHIQQQIHHHHHH") +_PHDR = struct.Struct(" replace step is a no-op. + mock.patch.object(mod, "replace_instrumented_libraries"), + mock.patch.dict(os.environ, patched_env, clear=False), + ): + mod.main(argv) + return captured + + def test_replacement_uses_code_coverage_repo_not_generic_repo(self): + captured = self._run_main(["--code-coverage-run-github-repo", "ROCm/TheRock"]) + self.assertEqual(captured["run_id"], "222") + self.assertEqual(captured["github_repository"], "ROCm/TheRock") + self.assertNotEqual(captured["github_repository"], "ROCm/rocm-libraries") + + def test_replacement_repo_defaults_to_github_repository_env(self): + # main() must be re-imported so the argparse default picks up the env, + # since the default is evaluated at parse time inside main(). + captured = self._run_main([], env={"GITHUB_REPOSITORY": "ROCm/TheRock"}) + self.assertEqual(captured["github_repository"], "ROCm/TheRock") + + +def _write_archive(path, members): + """Write an artifact archive of `members`: name -> content, None for a dir.""" + with tarfile.open(path, "w:xz") as tf: + for name, content in members.items(): + info = tarfile.TarInfo(name) + if content is None: + info.type = tarfile.DIRTYPE + tf.addfile(info) + else: + info.size = len(content) + tf.addfile(info, io.BytesIO(content)) + + +class TestReplaceInstrumentedLibraries(unittest.TestCase): + def _replaced_files(self, archive_name, artifacts, members): + """Run the replacement into an empty tree; return the files it wrote.""" + with tempfile.TemporaryDirectory() as tmp: + tmp = Path(tmp) + downloads = tmp / "downloads" + downloads.mkdir() + _write_archive(downloads / archive_name, members) + output_dir = tmp / "install" + with mock.patch.object(mod, "pin_phdr_table", return_value=False): + mod.replace_instrumented_libraries(artifacts, downloads, output_dir) + return sorted( + p.relative_to(output_dir).as_posix() + for p in output_dir.rglob("*") + if p.is_file() + ) + + def test_pins_program_headers_of_every_replaced_file(self): + with tempfile.TemporaryDirectory() as tmp: + tmp = Path(tmp) + downloads = tmp / "downloads" + downloads.mkdir() + prefix = "math-libs/BLAS/rocBLAS/stage" + other_prefix = "math-libs/BLAS/hipBLAS/stage" + members = { + "artifact_manifest.txt": f"{prefix}\n{other_prefix}\n".encode(), + f"{prefix}/lib/librocblas.so.5.8": b"instrumented library", + f"{prefix}/lib/rocblas/library": None, + f"{prefix}/lib/rocblas/library/TensileLibrary.dat": b"data", + f"{other_prefix}/lib/libhipblas.so.3.8": b"some other library", + } + _write_archive(downloads / "blas_lib_generic.tar.xz", members) + output_dir = tmp / "install" + + with mock.patch.object(mod, "pin_phdr_table", return_value=False) as pin: + mod.replace_instrumented_libraries( + {"blas": ["rocBLAS"]}, downloads, output_dir + ) + + library = output_dir / "lib" / "librocblas.so.5.8" + data = output_dir / "lib" / "rocblas" / "library" / "TensileLibrary.dat" + self.assertEqual(library.read_bytes(), b"instrumented library") + self.assertFalse((output_dir / "lib" / "libhipblas.so.3.8").exists()) + self.assertEqual( + pin.call_args_list, + [ + mock.call(library, require_section="__llvm_prf_cnts"), + mock.call(data, require_section="__llvm_prf_cnts"), + ], + ) + + def test_replaces_every_file_in_the_component_stage_dirs(self): + # Laid out like prim_test: tests named after what they test, and the + # device code of all three projects in hipCUB's stage directory. + stages = { + "hipCUB": "math-libs/hipCUB/stage", + "rocPRIM": "math-libs/rocPRIM_tests/stage", + "rocThrust": "math-libs/rocThrust/stage", + } + files = { + "hipCUB": ["bin/test_hipcub_block_scan"], + "rocPRIM": ["bin/test_block_scan"], + "rocThrust": ["bin/merge.hip", "bin/rocthrust/CTestTestfile.cmake"], + } + members = { + "artifact_manifest.txt": "".join( + f"{s}\n" for s in stages.values() + ).encode(), + f"{stages['hipCUB']}/.kpack/prim_test_gfx942.kpack": b"device code", + } + for folder, paths in files.items(): + for path in paths: + members[f"{stages[folder]}/{path}"] = b"instrumented test" + + # The shared kpack follows whichever project's host binaries are + # swapped in, or they look up kernels the baseline's kpack lacks. + for folder, paths in files.items(): + with self.subTest(folder=folder): + self.assertEqual( + self._replaced_files( + "prim_test_generic.tar.xz", {"prim": [folder]}, members + ), + sorted([".kpack/prim_test_gfx942.kpack", *paths]), + ) + + def test_keeps_device_code_of_artifacts_without_the_component_stage(self): + prefix = "math-libs/BLAS/hipBLASLt/stage" + members = { + "artifact_manifest.txt": f"{prefix}\n".encode(), + f"{prefix}/.kpack/blas_lib_gfx942.kpack": b"device code", + f"{prefix}/lib/libhipblaslt.so": b"some other library", + } + self.assertEqual( + self._replaced_files( + "blas_lib_gfx942.tar.xz", {"blas": ["rocBLAS"]}, members + ), + [], + ) + + def test_replaces_device_code_named_after_the_component(self): + prefix = "math-libs/rocWMMA/stage" + members = { + "artifact_manifest.txt": f"{prefix}\n".encode(), + f"{prefix}/.kpack/rocwmma_test_gfx942.kpack": b"device code", + f"{prefix}/bin/gemm_xdl-validate": b"instrumented test", + } + self.assertEqual( + self._replaced_files( + "rocwmma_test_gfx942.tar.xz", {"rocwmma": ["rocWMMA"]}, members + ), + [".kpack/rocwmma_test_gfx942.kpack", "bin/gemm_xdl-validate"], + ) + + def test_wrapper_brings_its_backend_from_the_same_build(self): + # hipRAND's linkage test compares the librocrand version against the + # rocRAND headers it was built with, so rocRAND must be swapped too. + hiprand = "math-libs/hipRAND/stage" + rocrand = "math-libs/rocRAND/stage" + members = { + "artifact_manifest.txt": f"{hiprand}\n{rocrand}\n".encode(), + f"{hiprand}/lib/libhiprand.so.1.1": b"instrumented hiprand", + f"{rocrand}/lib/librocrand.so.1.1": b"instrumented rocrand", + } + self.assertEqual( + self._replaced_files( + "rand_lib_generic.tar.xz", + mod.replacement_folders(["hiprand"]), + members, + ), + ["lib/libhiprand.so.1.1", "lib/librocrand.so.1.1"], + ) + + +class TestReplacementFolders(unittest.TestCase): + def test_wrapper_adds_its_same_artifact_backend(self): + self.assertEqual( + mod.replacement_folders(["hiprand"]), {"rand": ["hipRAND", "rocRAND"]} + ) + + def test_backend_is_not_listed_twice(self): + self.assertEqual( + mod.replacement_folders(["hipfft", "rocfft"]), + {"fft": ["hipFFT", "rocFFT"]}, + ) + + def test_backend_alone_does_not_bring_its_wrapper(self): + self.assertEqual(mod.replacement_folders(["rocrand"]), {"rand": ["rocRAND"]}) + + def test_backends_ship_in_their_wrappers_artifact(self): + folders = {folder for _, folder in mod.COMPONENT_MAP.values()} + for comp, backends in mod.SAME_ARTIFACT_BACKENDS.items(): + artifact = mod.COMPONENT_MAP[comp][0] + for backend in backends: + with self.subTest(comp=comp, backend=backend): + self.assertIn(backend, folders) + owner = next( + name + for name, folder in mod.COMPONENT_MAP.values() + if folder == backend + ) + self.assertEqual(owner, artifact) + + +class TestHostLibraries(unittest.TestCase): + def test_hipsparselt_takes_tensilelite_host_and_origami_from_blas(self): + self.assertEqual( + mod.host_library_deps(["hipsparselt"]), + { + "blas": [ + ("hipBLASLt", "libtensilelite-host"), + ("origami", "liborigami"), + ] + }, + ) + + def test_component_without_host_libraries_needs_none(self): + self.assertEqual(mod.host_library_deps(["rocblas"]), {}) + + def test_downloads_only_the_generic_lib_archive_of_host_artifacts(self): + class _Backend: + base_uri = "s3://fake-bucket/fake" + + def list_artifacts(self, name_filter=None): + return [ + "sparse_lib_generic.tar.zst", + "sparse_test_gfx942.tar.zst", + "blas_lib_generic.tar.zst", + "blas_lib_gfx942.tar.zst", + "blas_dev_generic.tar.zst", + "blas_test_generic.tar.zst", + ] + + with tempfile.TemporaryDirectory() as tmp: + opts = mock.Mock( + artifact_group="gfx94X-dcgpu", amdgpu_targets="", output_dir=Path(tmp) + ) + with ( + mock.patch.object( + mod, "create_backend_from_env", return_value=_Backend() + ), + mock.patch.object( + mod, "download_artifact", return_value=Path(tmp) / "x" + ) as download, + ): + mod.download_replacement_artifacts( + "222", {"sparse"}, opts, "ROCm/TheRock", "ci", {"blas"} + ) + fetched = sorted(c.args[0].artifact_key for c in download.call_args_list) + self.assertEqual( + fetched, + [ + "blas_lib_generic.tar.zst", + "sparse_lib_generic.tar.zst", + "sparse_test_gfx942.tar.zst", + ], + ) + + def test_replaces_only_the_named_host_libraries(self): + hipblaslt = "math-libs/BLAS/hipBLASLt/stage" + origami = "math-libs/BLAS/origami/stage" + rocblas = "math-libs/BLAS/rocBLAS/stage" + members = { + "artifact_manifest.txt": f"{hipblaslt}\n{origami}\n{rocblas}\n".encode(), + f"{hipblaslt}/lib/libhipblaslt.so.1.0": b"hipblaslt", + f"{hipblaslt}/lib/libtensilelite-host.so.1.0": b"tensilelite-host", + f"{origami}/lib/liborigami.so.1.0": b"origami", + f"{rocblas}/lib/librocblas.so.5": b"rocblas", + } + with tempfile.TemporaryDirectory() as tmp: + tmp = Path(tmp) + downloads = tmp / "downloads" + downloads.mkdir() + _write_archive(downloads / "blas_lib_generic.tar.xz", members) + output_dir = tmp / "install" + with mock.patch.object(mod, "pin_phdr_table", return_value=False): + mod.replace_host_libraries( + mod.host_library_deps(["hipsparselt"]), downloads, output_dir + ) + written = sorted( + p.relative_to(output_dir).as_posix() + for p in output_dir.rglob("*") + if p.is_file() + ) + self.assertEqual( + written, ["lib/liborigami.so.1.0", "lib/libtensilelite-host.so.1.0"] + ) + + def test_missing_host_library_is_an_error(self): + prefix = "math-libs/BLAS/hipBLASLt/stage" + members = { + "artifact_manifest.txt": f"{prefix}\n".encode(), + f"{prefix}/lib/libhipblaslt.so.1.0": b"hipblaslt", + } + with tempfile.TemporaryDirectory() as tmp: + tmp = Path(tmp) + downloads = tmp / "downloads" + downloads.mkdir() + _write_archive(downloads / "blas_lib_generic.tar.xz", members) + with self.assertRaises(IOError): + mod.replace_host_libraries( + {"blas": [("hipBLASLt", "libtensilelite-host")]}, + downloads, + tmp / "install", + ) + + +if __name__ == "__main__": + unittest.main() diff --git a/build_tools/tests/workflow_outputs_test.py b/build_tools/tests/workflow_outputs_test.py index 9e9514da99d..adfe443d3e3 100644 --- a/build_tools/tests/workflow_outputs_test.py +++ b/build_tools/tests/workflow_outputs_test.py @@ -332,6 +332,27 @@ def test_lookup_workflow_run_triggers_api(self, mock_retrieve): release_type=None, ) + @mock.patch("_therock_utils.workflow_outputs._retrieve_bucket_info") + def test_lookup_finds_a_variant_run_by_its_workflow_run_id(self, mock_retrieve): + """A - suffix stays in the prefix but not in the API lookup.""" + mock_retrieve.return_value = ( + "ROCm-rocm-libraries/", + "therock-ci-artifacts-external", + ) + root = WorkflowOutputRoot.from_workflow_run( + run_id="99999-coverage", + platform="linux", + github_repository="ROCm/rocm-libraries", + lookup_workflow_run=True, + ) + self.assertEqual(root.prefix, "ROCm-rocm-libraries/99999-coverage-linux") + mock_retrieve.assert_called_once_with( + github_repository="ROCm/rocm-libraries", + workflow_run_id="99999", + workflow_run=None, + release_type=None, + ) + @mock.patch("_therock_utils.workflow_outputs._retrieve_bucket_info") def test_with_workflow_run_dict(self, mock_retrieve): """When workflow_run is provided, it's passed through (no API call).""" diff --git a/cmake/therock_coverage_profile_stub.c b/cmake/therock_coverage_profile_stub.c new file mode 100644 index 00000000000..3b4069f016c --- /dev/null +++ b/cmake/therock_coverage_profile_stub.c @@ -0,0 +1,29 @@ +// Copyright Advanced Micro Devices, Inc. +// SPDX-License-Identifier: MIT + +// No-op definitions of every entry point into the GPU half of the LLVM profile +// runtime (InstrProfilingPlatformROCm.cpp, in libclang_rt.profile.a and +// libclang_rt.profile_rocm.a). +// therock_subproject.cmake links this into each coverage-instrumented +// executable and shared library so that the real ones are never linked; see the +// coverage comments there. + +#pragma GCC visibility push(hidden) + +void __llvm_profile_offload_register_shadow_variable(void *Ptr) { (void)Ptr; } + +void __llvm_profile_offload_register_section_shadow_variable(void *Ptr) { + (void)Ptr; +} + +void __llvm_profile_offload_register_dynamic_module(int ModuleLoadRc, + void **Ptr, + const void *Image) { + (void)ModuleLoadRc; + (void)Ptr; + (void)Image; +} + +void __llvm_profile_offload_unregister_dynamic_module(void *Ptr) { (void)Ptr; } + +int __llvm_profile_hip_collect_device_data(void) { return 0; } diff --git a/cmake/therock_subproject.cmake b/cmake/therock_subproject.cmake index 9733f4a4476..eb00bfea1de 100644 --- a/cmake/therock_subproject.cmake +++ b/cmake/therock_subproject.cmake @@ -528,6 +528,7 @@ function(therock_cmake_subproject_declare target_name) THEROCK_PREFIX_DIR "${_prefix_dir}" THEROCK_CMAKE_SOURCE_DIR "${_cmake_source_dir}" THEROCK_CMAKE_PROJECT_INIT_FILE "${ARG_BINARY_DIR}/${ARG_BUILD_DIR}_init.cmake" + THEROCK_CMAKE_PROJECT_COVERAGE_FILE "${ARG_BINARY_DIR}/${ARG_BUILD_DIR}_coverage.cmake" THEROCK_CMAKE_PROJECT_TOOLCHAIN_FILE "${ARG_BINARY_DIR}/${ARG_BUILD_DIR}_toolchain.cmake" THEROCK_CMAKE_ARGS "${ARG_CMAKE_ARGS}" THEROCK_CMAKE_INCLUDES "${ARG_CMAKE_INCLUDES}" @@ -649,6 +650,7 @@ function(therock_cmake_subproject_activate target_name) get_target_property(_cmake_args "${target_name}" THEROCK_CMAKE_ARGS) get_target_property(_cmake_includes "${target_name}" THEROCK_CMAKE_INCLUDES) get_target_property(_cmake_project_init_file "${target_name}" THEROCK_CMAKE_PROJECT_INIT_FILE) + get_target_property(_cmake_project_coverage_file "${target_name}" THEROCK_CMAKE_PROJECT_COVERAGE_FILE) get_target_property(_cmake_project_toolchain_file "${target_name}" THEROCK_CMAKE_PROJECT_TOOLCHAIN_FILE) get_target_property(_cmake_source_dir "${target_name}" THEROCK_CMAKE_SOURCE_DIR) get_target_property(_exclude_from_all "${target_name}" THEROCK_EXCLUDE_FROM_ALL) @@ -918,6 +920,154 @@ function(therock_cmake_subproject_activate target_name) _therock_get_flag_init_contents(_flag_init_contents "${target_name}") string(APPEND _init_contents "${_flag_init_contents}") + # Give the LLVM profile runtime the libraries it depends on. + # + # -fprofile-instr-generate links libclang_rt.profile_rocm.a, whose + # InstrProfilingPlatformROCm.cpp calls dlsym, dladdr and pthread_once. The + # driver does not put -ldl or -lpthread on the link line, and on the + # manylinux base those symbols are still in libdl and libpthread rather than + # in libc, so an instrumented shared library is left with undefined + # references. The library itself links -- shared objects tolerate that -- but + # every executable linking against it then fails + # --no-allow-shlib-undefined, which is how this shows up: hipRAND's tests + # cannot link against an instrumented libhiprand.so. + # + # push-state/pop-state keeps --no-as-needed scoped to these two, so the + # DT_NEEDED entries are recorded no matter where the flags land relative to + # the objects that need them, without forcing a DT_NEEDED for every other + # library on the line. + string(TOUPPER "${_logical_target_name}" _coverage_project_name) + set(_coverage_var_name "${_coverage_project_name}_ENABLE_COVERAGE") + if(NOT MSVC AND ${_coverage_var_name}) + set(_coverage_link_flags + " -Wl,--push-state,--no-as-needed -ldl -lpthread -Wl,--pop-state") + string(APPEND _init_contents + "string(APPEND CMAKE_EXE_LINKER_FLAGS \"${_coverage_link_flags}\")\n") + string(APPEND _init_contents + "string(APPEND CMAKE_SHARED_LINKER_FLAGS \"${_coverage_link_flags}\")\n") + string(APPEND _init_contents + "string(APPEND CMAKE_MODULE_LINKER_FLAGS \"${_coverage_link_flags}\")\n") + endif() + + # Keep the instrumentation off the GPU. + # + # The upstream coverage options pass -fprofile-instr-generate and + # -fcoverage-mapping unqualified, and the HIP driver forwards both to the + # device compilation, so kernels are instrumented too. That is not what this + # pipeline measures, and it does not survive contact with the runtime: the + # device image gains a per-TU __llvm_profile_sections_ symbol that + # libclang_rt.profile_rocm.a looks up with hipModuleGetGlobal at exit, and a + # __profd_ for every kernel. Symbols the runtime cannot resolve abort the + # process after the tests have already passed (rocSPARSE, on a rocPRIM + # trampoline kernel), and where the readback does succeed it writes a second + # profile named after the device -- gfx942:sramecc+:xnack-.0. -- + # which cannot be uploaded as an artifact because of the colons. + # + # -Xarch_device confines the negation to the device compilation; the host + # keeps -fprofile-instrument=clang and its coverage mapping. Order decides + # this, since the last of a -f/-fno- pair wins and the project's own flags + # come from COMPILE_OPTIONS, which land after CMAKE__FLAGS. Appending + # to the compile rule after is the one position that is reliably + # last. The rule does not exist yet in CMAKE_PROJECT_TOP_LEVEL_INCLUDES, + # which runs before any language is enabled, hence the separate file below. + # + # This reaches the languages a project() call enables, which is where the + # HIP sources in these projects are compiled -- as CXX, through the amd-hip + # toolchain. A language brought up later by a bare enable_language() is not + # covered, and would need the same rewrite from wherever that happens. + # + # The same file takes these projects off ccache. ccache (4.11 at least) + # removes every -Xarch_ pair unless also arrived as an + # Apple-style -arch, from the cache key and from the command it runs alike. + # That drops the negation below, so kernels end up instrumented after all, + # and drops the -Xarch_host spelling some upstream coverage options use + # (rocRAND, hipRAND), so those libraries end up not instrumented at all. + # Nothing fails at build time: either way the damage only shows up once the + # tests run. + set(_coverage_project_include_arg) + set(_coverage_configure_depends) + if(NOT MSVC AND ${_coverage_var_name}) + set(_coverage_host_only_contents + "# Generated by therock_subproject.cmake; see the coverage comment there.\n") + foreach(_coverage_lang IN ITEMS C CXX HIP) + string(APPEND _coverage_host_only_contents + "set(CMAKE_${_coverage_lang}_COMPILER_LAUNCHER \"\")\n") + endforeach() + # Runs on every project() call, so the rewrite has to be idempotent. + # -Wno-unused-command-line-argument because -Xarch_device is unused for + # the plain C++ sources that make up most of these projects, and several + # of them build with -Werror. + foreach(_coverage_lang IN ITEMS C CXX HIP) + string(APPEND _coverage_host_only_contents + "if(CMAKE_${_coverage_lang}_COMPILE_OBJECT AND NOT CMAKE_${_coverage_lang}_COMPILE_OBJECT MATCHES \"fno-profile-instr-generate\")\n" + " string(REPLACE \"\" \" -Xarch_device -fno-profile-instr-generate -Xarch_device -fno-coverage-mapping -Wno-unused-command-line-argument\"\n" + " CMAKE_${_coverage_lang}_COMPILE_OBJECT \"\${CMAKE_${_coverage_lang}_COMPILE_OBJECT}\")\n" + "endif()\n") + endforeach() + # Uninstrumented kernels don't stop clang from giving every HIP translation + # unit a host-side shadow of its device counters, registered from a + # constructor, or the driver from force-linking the GPU half of the profile + # runtime (InstrProfilingPlatformROCm.cpp, in both libclang_rt.profile.a + # and libclang_rt.profile_rocm.a) to take the registrations. At exit that + # half asks the HIP runtime for the device side of each shadow, which only + # exists when the device code has profile data. The HIP runtime aborts + # ("Cannot create GlobalVar Obj") right after the first module writes its + # profile, so every other module's is lost. Linking a stub that defines + # each of its entry points as a no-op into every executable and shared + # library takes the registrations instead. -fuse-cuid=none would stop the + # shadows instead, but then every translation unit defines the same + # __hip_cuid_ and any library with two HIP sources fails to link. Should a + # compiler update add an entry point the stub lacks, links fail on + # duplicate symbols rather than bringing the GPU half back. + # + # The GPU half also wraps hipLaunchKernel, hipModuleLoad and the like, as + # weak definitions of those names, so it is extracted for them too by a + # link that reaches the profile runtime archives while they are still + # unresolved. The driver adds the archives after the HIP runtime, but + # rocSPARSE names them itself, ahead of it. Putting the HIP runtime in + # front of every object resolves those functions first; --as-needed keeps + # it from becoming a dependency of anything that does not call into it. + set(_coverage_profile_stub_source + "${THEROCK_SOURCE_DIR}/cmake/therock_coverage_profile_stub.c") + string(APPEND _coverage_host_only_contents + "set(_therock_coverage_stub \"\${CMAKE_BINARY_DIR}/therock_coverage_profile_stub.o\")\n" + "if(\"${_coverage_profile_stub_source}\" IS_NEWER_THAN \"\${_therock_coverage_stub}\")\n" + " if(CMAKE_C_COMPILER)\n" + " set(_therock_coverage_cc \"\${CMAKE_C_COMPILER}\")\n" + " else()\n" + " set(_therock_coverage_cc \"\${CMAKE_CXX_COMPILER}\")\n" + " endif()\n" + " execute_process(\n" + " COMMAND \"\${_therock_coverage_cc}\" -x c -O2 -fPIC -c\n" + " \"${_coverage_profile_stub_source}\" -o \"\${_therock_coverage_stub}\"\n" + " RESULT_VARIABLE _therock_coverage_result)\n" + " if(NOT _therock_coverage_result EQUAL 0)\n" + " message(FATAL_ERROR \"Could not compile ${_coverage_profile_stub_source}\")\n" + " endif()\n" + "endif()\n" + "set(_therock_coverage_hip \"\")\n" + "if(THEROCK_TOOLCHAIN_ROOT)\n" + " file(GLOB _therock_coverage_hip_libs \"\${THEROCK_TOOLCHAIN_ROOT}/lib/libamdhip64.so*\")\n" + " if(_therock_coverage_hip_libs)\n" + " list(GET _therock_coverage_hip_libs 0 _therock_coverage_hip_lib)\n" + " set(_therock_coverage_hip \" -Wl,--push-state,--as-needed \${_therock_coverage_hip_lib} -Wl,--pop-state\")\n" + " endif()\n" + "endif()\n" + "foreach(_therock_coverage_kind IN ITEMS EXE SHARED MODULE)\n" + " if(NOT CMAKE_\${_therock_coverage_kind}_LINKER_FLAGS MATCHES \"therock_coverage_profile_stub\")\n" + " string(APPEND CMAKE_\${_therock_coverage_kind}_LINKER_FLAGS \" \${_therock_coverage_stub}\${_therock_coverage_hip}\")\n" + " endif()\n" + "endforeach()\n") + file(CONFIGURE OUTPUT "${_cmake_project_coverage_file}" + CONTENT "${_coverage_host_only_contents}" @ONLY ESCAPE_QUOTES) + list(APPEND _fprint_files + "${_cmake_project_coverage_file}" "${_coverage_profile_stub_source}") + set(_coverage_project_include_arg + "-DCMAKE_PROJECT_INCLUDE=${_cmake_project_coverage_file}") + set(_coverage_configure_depends + "${_cmake_project_coverage_file}" "${_coverage_profile_stub_source}") + endif() + file(CONFIGURE OUTPUT "${_cmake_project_init_file}" CONTENT "${_init_contents}" @ONLY ESCAPE_QUOTES) list(APPEND _fprint_files "${_cmake_project_init_file}") list(APPEND _fprint_files "${ROCM_BUILD_FLAGS_STATE_FILE}") @@ -933,6 +1083,32 @@ function(therock_cmake_subproject_activate target_name) list(APPEND _cmake_args "${${target_name}_CMAKE_ARGS}") + # Translate TheRock's -D_ENABLE_COVERAGE into whichever option the + # subproject actually implements. CMake definitions are conventionally upper + # case, so only the upper case spelling of TheRock's flag is honored (i.e. + # HIPDNN_ENABLE_COVERAGE, not hipDNN_ENABLE_COVERAGE). + # + # The names upstream are not standardised: most projects take + # BUILD_CODE_COVERAGE or CODE_COVERAGE, a few take _ENABLE_COVERAGE, + # RCCL takes ENABLE_CODE_COVERAGE. The generic ones are the reason this is + # done per subproject rather than by setting a global: defining + # BUILD_CODE_COVERAGE for the whole build would instrument every project that + # happens to understand it, which is exactly what per-project coverage is + # trying to avoid. + # + # _coverage_project_name and _coverage_var_name are set further up, where the + # profile runtime's link libraries are added to the init file. + if(DEFINED ${_coverage_var_name}) + set(_coverage_option "${THEROCK_COVERAGE_OPTION_${_coverage_project_name}}") + if(NOT _coverage_option) + message(FATAL_ERROR + "${_logical_target_name}: ${_coverage_var_name} was set but the project " + "has no coverage option registered. Add one to COVERAGE_PROJECTS in " + "build_tools/github_actions/configure_coverage_ci.py.") + endif() + list(APPEND _cmake_args "-D${_coverage_option}=${${_coverage_var_name}}") + endif() + # Derive the CMAKE_BUILD_TYPE from either {project}_BUILD_TYPE or the global # CMAKE_BUILD_TYPE. set(_cmake_build_type "${${target_name}_BUILD_TYPE}") @@ -1068,6 +1244,7 @@ function(therock_cmake_subproject_activate target_name) "-DTHEROCK_STAGE_INSTALL_ROOT=${_stage_dir}" "-DCMAKE_TOOLCHAIN_FILE=${_cmake_project_toolchain_file}" "-DCMAKE_PROJECT_TOP_LEVEL_INCLUDES=${_cmake_project_init_file}" + ${_coverage_project_include_arg} ${_cmake_args} # CMake doesn't always generate a compile_commands.json so touch one to keep # the build graph sane. @@ -1085,6 +1262,7 @@ function(therock_cmake_subproject_activate target_name) "${_cmake_source_dir}/CMakeLists.txt" "${_cmake_project_toolchain_file}" "${_cmake_project_init_file}" + ${_coverage_configure_depends} "${_global_post_include}" "${_hardening_module}" "${ROCM_BUILD_FLAGS_STATE_FILE}" diff --git a/docs/development/README.md b/docs/development/README.md index 2a576b0fff6..4e75f94ab38 100644 --- a/docs/development/README.md +++ b/docs/development/README.md @@ -10,6 +10,7 @@ - [Artifacts](artifacts.md) - [Build System](build_system.md) +- [Code Coverage](code_coverage.md) - [Dependencies](dependencies.md) - [Development Guide](development_guide.md) - [Installing Artifacts](installing_artifacts.md) @@ -30,6 +31,7 @@ - [S3 Buckets](s3_buckets.md) - [GitHub Actions Debugging](github_actions_debugging.md) - [CI Behavior Manipulation](ci_behavior_manipulation.md) +- [Coverage CI](code_coverage.md#coverage-ci) - [Stage Reuse](stage_reuse.md) - [CCache Troubleshooting](ccache_troubleshooting.md) - [WSL ROCDXG CI Stage](wsl_rocdxg.md) diff --git a/docs/development/ci_overview.md b/docs/development/ci_overview.md index f8753850c64..9d4a9ba1e8a 100644 --- a/docs/development/ci_overview.md +++ b/docs/development/ci_overview.md @@ -50,6 +50,16 @@ the job starts on a Windows runner, then runs artifact fetch, configure, build, and upload steps inside a WSL Ubuntu shell. See [WSL ROCDXG CI Stage](wsl_rocdxg.md) for details. +Code coverage has its own entry point, because the projects it measures have to +be rebuilt with instrumentation before their tests can be run: + +- [`.github/workflows/multi_arch_ci_coverage_linux.yml`](/.github/workflows/multi_arch_ci_coverage_linux.yml) - the entry point (`setup_coverage_matrix`, the compiler-runtime and math-libs builds, `coverage_report`, `ci_summary`). Dispatched by hand, or called from another repository through `workflow_call`: pass the run id of a recent nightly as `baseline_run_id` to take the non-instrumented dependencies from it, or leave it empty and set `baseline_release_type` to `ci` to test this run's own instrumented stack. A future `multi_arch_ci_coverage_windows.yml` will mirror it + - [`.github/workflows/multi_arch_ci_coverage_report.yml`](/.github/workflows/multi_arch_ci_coverage_report.yml) - runs one project's tests on one GPU family, then turns the collected profiles into an lcov file, a text summary and a browsable HTML report + +See [Code Coverage](code_coverage.md) for the CMake options and how to onboard a +project, and [Coverage CI](code_coverage.md#coverage-ci) for the end-to-end +walkthrough of a run. + ## Build Phase TheRock builds ROCm components from source and produces **artifacts** - archive slices of key components. @@ -180,6 +190,7 @@ to identify affected job groups. - [artifacts.md](artifacts.md) - Artifact organization and packaging - [build_system.md](build_system.md) - CMake build architecture +- [code_coverage.md](code_coverage.md) - Instrumented builds and coverage reports - [dependencies.md](dependencies.md) - Dependency management - [installing_artifacts.md](installing_artifacts.md) - Installing ROCm from artifacts @@ -196,6 +207,7 @@ to identify affected job groups. - [workflow_outputs.md](workflow_outputs.md) - CI output directory structure - [github_actions_debugging.md](github_actions_debugging.md) - Debugging GitHub Actions - [ci_behavior_manipulation.md](ci_behavior_manipulation.md) - Controlling CI behavior with labels and inputs +- [code_coverage.md#coverage-ci](code_coverage.md#coverage-ci) - How a coverage run is dispatched, built, tested and reported - [stage_reuse.md](stage_reuse.md) - Reusing unaffected build stages from a baseline run - [manifest_diff.md](manifest_diff.md) - Manifest diff report (submodule SHA changes between two commits) diff --git a/docs/development/code_coverage.md b/docs/development/code_coverage.md new file mode 100644 index 00000000000..30d3dc9ad0e --- /dev/null +++ b/docs/development/code_coverage.md @@ -0,0 +1,434 @@ +# Code Coverage + +TheRock builds ROCm libraries with LLVM source-based coverage and reports to +[Codecov](https://app.codecov.io/gh/ROCm/TheRock). Coverage is opt-in per +project — instrumentation slows a library substantially, and a report is only +worth producing when the test suite gives a meaningful signal. + +## How it works + +- **Instrumented** — compiled with `-fprofile-instr-generate`; every line bumps + a counter. Slower and bigger; never shipped. +- **profraw** — the counter file an instrumented process writes on exit. +- **lcov** — the final report: which lines ran and which did not. + +Pipeline: build instrumented → run tests → collect profraw → convert to lcov. + +A project's tests load its dependencies too. If those were instrumented, their +counters would pollute the report. So exactly one library is instrumented; the +rest come from the regular nightly build (the **baseline**): + +| | Source | What it provides | +| ------------ | ------------------- | ---------------------------------- | +| **Baseline** | Regular nightly run | Whole ROCm stack, non-instrumented | +| **Coverage** | This run | Only the one instrumented project | + +The test runner installs the baseline, then replaces only the selected +project's libraries and tests with instrumented ones. See +[Coverage CI](#coverage-ci). + +## Enabling coverage for a local build + +```bash +cmake -B build -GNinja . \ + -DTHEROCK_AMDGPU_FAMILIES=gfx94X-dcgpu \ + -DHIPRAND_ENABLE_COVERAGE=ON +``` + +`_ENABLE_COVERAGE` is TheRock's own flag; upstream option names are +not standardised (hipRAND: `BUILD_CODE_COVERAGE`, rocRAND: `CODE_COVERAGE`, +RCCL: `ENABLE_CODE_COVERAGE`), so `therock_subproject.cmake` translates it to +whichever name `COVERAGE_PROJECTS` registers for that project. Passing the flag +for an unregistered project is a configure error, not a silent no-op. + +### Enabling a whole group + +| Option | Instruments | +| ------------------------------------- | -------------------------------------------- | +| `THEROCK_COVERAGE_ROCM_LIBRARIES_ALL` | all coverage-enabled rocm-libraries projects | +| `THEROCK_COVERAGE_ROCM_SYSTEMS_ALL` | all coverage-enabled rocm-systems projects | +| `THEROCK_COVERAGE_ALL` | both of the above | + +Each expands to individual `_ENABLE_COVERAGE` flags. An explicit +`-D_ENABLE_COVERAGE=OFF` wins over a group flag. Membership is +generated from `COVERAGE_PROJECTS`; these options default `OFF` and have no +effect on a normal build. + +### Host code only + +Upstream coverage flags are unqualified and the HIP driver forwards them to +device compilation. Two failures result: the runtime aborts looking up missing +profiling symbols in the device image, and when readback succeeds it writes a +profile named after the device target (colons included, breaking CI artifact +upload). `therock_subproject.cmake` appends the negation scoped to device +compilation: + +``` +-Xarch_device -fno-profile-instr-generate -Xarch_device -fno-coverage-mapping +``` + +The later of a `-f`/`-fno-` pair wins; the project's own flags arrive after +`CMAKE__FLAGS`, so the negation is appended to +`CMAKE__COMPILE_OBJECT` via a generated file passed as +`CMAKE_PROJECT_INCLUDE`. + +The same file clears `CMAKE__COMPILER_LAUNCHER`, so instrumented +projects never compile through ccache. ccache (4.11 at least) removes every +`-Xarch_ ` pair whose arch did not also arrive as an Apple-style +`-arch`, from both the cache key and the command it runs. That drops the +negation above, so kernels are instrumented after all, and drops the +`-Xarch_host` spelling rocRAND's and hipRAND's own coverage options use, so +those libraries are not instrumented at all. Neither fails the build; both +only show up once the tests run. Projects that are not instrumented keep using +the cache. + +### The profile runtime's link dependencies + +`libclang_rt.profile_rocm.a` needs `-ldl` and `-lpthread`, which the driver +does not supply. On the manylinux base these are not in libc, so an instrumented +shared library gets undefined references and any executable linking it fails +`--no-allow-shlib-undefined`. `therock_subproject.cmake` adds both libraries +automatically. + +### The profile runtime's GPU half + +Keeping kernels uninstrumented does not stop the abort described above. Clang +still gives every HIP translation unit a host-side shadow of its device +counters, registered from a constructor. The driver still force-links +`InstrProfilingPlatformROCm.cpp`, which both `libclang_rt.profile.a` and +`libclang_rt.profile_rocm.a` carry, to take those registrations. At exit, once +a module has written its host profile, that file asks the HIP runtime for the +device side of each shadow. The device side only exists when the device code +has profile data, so the HIP runtime aborts +(`Cannot create GlobalVar Obj for symbol: __llvm_profile_sections_`). Every +module whose profile writer has not run yet loses its profile, which is how an +instrumented test executable costs the library under test its coverage. The same +file also wraps `hipLaunchKernel`, `hipModuleLoad` and the like, as weak +definitions of those names. + +`therock_subproject.cmake` therefore compiles +`cmake/therock_coverage_profile_stub.c`, which defines that file's five entry +points as hidden no-ops. It adds the object to `CMAKE_EXE_LINKER_FLAGS`, +`CMAKE_SHARED_LINKER_FLAGS` and `CMAKE_MODULE_LINKER_FLAGS`, followed by the +HIP runtime under `--as-needed`, so nothing pulls in the real one. The HIP +runtime has to come first because a link that reaches the profile runtime +archives while `hipLaunchKernel` is still unresolved extracts the real file for +its wrappers. The driver adds the archives after the HIP runtime, but rocSPARSE +names them itself, ahead of it. If a compiler update adds an entry point the +stub lacks, links fail on duplicate symbols. `-fuse-cuid=none` would remove the +shadows instead, but then every translation unit defines the same +`__hip_cuid_`, and any library with two HIP sources fails to link. + +## Producing a report locally + +Set `LLVM_PROFILE_FILE` with `%p`/`%m` substitutions to keep per-process files +distinct, run the tests, then call `merge_coverage_report.py`: + +```bash +export LLVM_PROFILE_FILE="$PWD/coverage-report/profraw/%p-%m.profraw" +ctest --test-dir build/math-libs/hipRAND/build + +python build_tools/github_actions/merge_coverage_report.py \ + --profraw-dir coverage-report/profraw \ + --rocm-dir build/dist/rocm \ + --object-globs "lib/libhiprand.so*" \ + --summary-output coverage-report/coverage_summary.txt \ + --html-output coverage-report/html +``` + +`llvm-profdata` and `llvm-cov` must match the compiler that built the objects; +the script prefers the copies under `/lib/llvm/bin`. Only the HTML +reads source files; lcov and the text table work from coverage mappings alone. +Use `--path-equivalence ,` if the source tree has moved since the +build. + +## Coverage CI + +```mermaid +graph TD + dispatch[Dispatch with projects_to_test] --> matrix[setup_coverage_matrix] + matrix --> compilerRuntime[Build compiler-runtime] + compilerRuntime --> mathLibs[Build instrumented math-libs] + matrix --> gate[Wait for a same-run baseline] + mathLibs --> report[Per project: configure, test, report] + gate --> report + report --> codecov[Codecov] +``` + +`multi_arch_ci_coverage_linux.yml` is the entry point: it computes the matrix, +builds the instrumented stack, then fans out to +`multi_arch_ci_coverage_report.yml` per project per GPU family. +`configure_coverage_ci.py` is the registry — CMake target, build stage, test +component, object globs, and Codecov flag for every onboarded project. + +### Dispatching a run + +Start from the Actions tab or `gh` CLI, or let the rocm-libraries nightly +call it via a cross-repo reusable `uses:` workflow call. There is no cron trigger. + +Pass a recent nightly run id as `baseline_run_id` (keep `baseline_release_type` +at `nightly`) and set `projects_to_test`; other inputs use sensible defaults. +Running coverage separately keeps failures off the nightly's status. Automatic +nightly dispatch is a follow-up; `baseline_run_id` is already the right input. + +### The instrumented build + +`setup_coverage_matrix` runs `configure_coverage_ci.py`, emitting the matrix, +`coverage_cmake_options`, and `needs_math_libs`. Selections naming a project +whose stage has no build job fail immediately. + +`build_compiler_runtime` always runs first — every other stage pulls its +inbound artifacts from it. It is a regular, non-instrumented build: no +compiler-runtime project has a supported coverage option. +`build_instrumented_math_libs` fans out over GPU families and is skipped when +`needs_math_libs` is false; 17 of the 19 measurable projects live there. + +`coverage_cmake_options` is appended to the math-libs configure line only. A +selection covering a whole group collapses to that group's option (the default +selection becomes `-DTHEROCK_COVERAGE_ROCM_LIBRARIES_ALL=ON`); a narrow one +names each project. `CMakeLists.txt` expands group options to +`_ENABLE_COVERAGE` flags; `therock_subproject.cmake` translates each +to the project's upstream name. Everything not selected builds normally. + +Build jobs use `multi_arch_build_portable_linux_artifacts.yml` directly. +Coverage sets three of its inputs: + +- `extra_cmake_options` carries `coverage_cmake_options`. +- `artifact_run_id` is `-coverage`. Artifacts and logs publish there + under `release_type: ci`, not under the run id itself. A caller that builds a + regular stack in the same run (the rocm-libraries nightly does, and uses it + as the baseline) publishes artifacts with the same names, which a shared + namespace would overwrite. +- `external_repo_config` is forwarded from the caller, so a rocm-libraries + nightly measures the commit it is testing. Without it the build uses + TheRock's pinned submodules. + +### Waiting for a same-run baseline + +Only the tests need the baseline, so a caller that builds it in the same run +(the rocm-libraries nightly) can start coverage alongside that build. When +`baseline_run_id` is this run's id, `wait_for_baseline` runs +`wait_for_baseline_stage.py`, which polls the run's jobs until the regular +Linux build's `Stage - Math Libs ()` job has succeeded for every family +(`Stage - Compiler Runtime` when no math-libs project is selected). +`coverage_report` needs it, so tests start once the baseline is in S3. A +baseline from an earlier run skips the wait. + +The wait fails, skipping the tests, when that job fails, when the stage never +runs (an earlier stage failed, or the caller marked it prebuilt or skipped), +when the run builds the stage only for other families, or after 300 minutes. +Jobs are matched by display name, ignoring those under the `Windows::` job, +which reuse the names; `wait_for_baseline_stage_test.py` checks the names +against `multi_arch_build_portable_linux.yml`. On a public repository, listing +the jobs needs no `actions` permission. + +### Test execution + +Per shard, `test_code_coverage_component.yml` does three things: + +1. **Install baseline + swap project.** `install_rocm_code_coverage_build.py` + installs from `--run-id` (the nightly) and replaces only the + project-under-test from `--code-coverage-run-id` (`-coverage`). It + then moves each replaced library's program header table back to where the + profile writer looks for it (see + [Program headers kpack moved](#program-headers-kpack-moved)). +1. **Set `LLVM_PROFILE_FILE`** to a per-shard path (`%p`/`%m`) so concurrent + processes don't overwrite each other. +1. **Run tests and upload profraw** under `always()` — a failing shard still + exercised code. Device-side profiles (named after the GPU target, colons + included) are dropped with a warning first: one of them makes + `upload-artifact` reject the whole upload. + +With `BUILD_VARIANT=coverage`, `fetch_test_configurations.py` multiplies each +component's `timeout_minutes` by 4, capped at 180 so the step ends inside the +job's 210-minute limit, and never lowered below the release value. +Instrumented tests run slower, and some upstream coverage options do more than +instrument: hipRAND's `BUILD_CODE_COVERAGE` adds `-O0 -g`, and rocRAND's +`CODE_COVERAGE` compiles in extra CPU-only test suites. + +#### Why there are two run ids + +The whole stack is built instrumented once to avoid rebuilding per-project +dependencies. But reporting must be per project: an instrumented rocRAND writes +its own profiles alongside hipRAND's, shifting hipRAND's numbers on unrelated +changes. The baseline-install-then-overlay isolates exactly one project. + +The baseline is read from `baseline_release_type` (normally `nightly`), not the +coverage run's `ci` channel — artifacts are bucketed per channel. The swap +downloads the project's whole artifact (`rand`, which also contains rocRAND) +but extracts only the project's files: everything under its own stage +directories (`math-libs/hipRAND/stage`), and elsewhere only paths naming its +library folder in `COMPONENT_MAP` (`hipRAND`). The stage directories are what +bring in a header-only project's tests, which are named after what they test +(rocThrust's `bin/merge.hip`). The device code under `.kpack/` is shared by +every project in the artifact, but host binaries look their kernels up in it by +name, so it is replaced along with them whenever the project has a stage +directory in the artifact; keeping the baseline's would fail with `Cannot find Symbol` once the two builds' versions differ (`HIPCUB_400700_NS` vs +`HIPCUB_400701_NS`). Other projects' baseline binaries in that artifact then +load the instrumented kpack, so only the swapped project's tests are expected +to pass. + +The prim projects (rocPRIM, hipCUB, rocThrust) are built in the coverage run +with `-DTHEROCK_FLAG_BUILD_HIPCCL=OFF`. The default unified hipCCL superbuild +has no rocPRIM/hipCUB/rocThrust subprojects for their `_ENABLE_COVERAGE` +flags to reach, so it would ship uninstrumented tests, all under one +`hipCCL_tests` stage directory. With no stage directory of its own, hipCUB would +not take the shared kpack, and its host binaries would fail with `device kernel image is invalid` against the baseline's kernels. + +A hip\* wrapper also brings the backend it calls at runtime from the same +artifact (`SAME_ARTIFACT_BACKENDS`: hipRAND brings rocRAND, hipFFT rocFFT, and +so on). Otherwise the instrumented wrapper runs against the baseline's backend, +which may be a different version than the one it was built with: hipRAND's +linkage test then fails comparing `hiprandGetVersion()` (answered by +librocrand) with the `ROCRAND_VERSION` it was compiled against. The backend's +objects are not in the wrapper's `--object-globs`, so its own code stays out of +the report, but header templates both instantiate share profile counters. +Dependencies in other artifacts (fftw3, rocBLAS for hipSOLVER) still come from +the baseline, except host libraries listed in `OTHER_ARTIFACT_HOST_LIBS`: +hipSPARSELt takes `libtensilelite-host` (hipBLASLt) and `liborigami` from the +coverage run's `blas_lib_generic`, because it calls their C++ interfaces and +aborts against the baseline's copies. Only those files are swapped. The +instrumented stack lives under +`-coverage`, so even a baseline from the same workflow run is never +overwritten by it. + +#### Program headers kpack moved + +kpack (`KPACK_SPLIT_ARTIFACTS`, on by default) moves the program header table +of each library it rewrites, the ones carrying device code, to the end of the +file. A `PT_LOAD` of its own maps the table there, at an address that is only +page-congruent with its file offset. ld.so copes. compiler-rt's profile writer +does not: to record binary IDs it reads the table at `__ehdr_start + e_phoff`. +For such a library that address is either unmapped, so rocBLAS and rocSPARSE +segfault at exit before writing a counter, or inside another segment, so +rocRAND reads unrelated bytes. kpack already pins `p_vaddr == p_offset` for +executables (`normalize_phdr_vaddr` in `rocm_kpack/elf/phdr_manager.py`). The +installer applies the same edit to every replaced shared library that has a +`__llvm_prf_cnts` section (`_therock_utils/elf_phdr.py`), growing the file when +the table has to move above the library's address range. + +A missing instrumented artifact fails the install step. A swap that matches no +files does not: the installer logs `Replaced 0` and the tests run against the +baseline's uninstrumented binaries. The failure then surfaces in the report +job, which finds no profraw files. A swap that misses just the binaries the +report measures leaves it counters from other programs: none of its +instrumented lines ran, or only the few those programs share. + +#### What scopes a report to one project + +`object_globs` limits what `llvm-cov` reports on. The overlay limits what is +instrumented: inline and template code is `linkonce_odr` and lands in every +binary that uses it — an instrumented sibling writes counters that +`llvm-profdata` merges in, unreachable by any `-object` filter. This matters +most for header-only projects (rocPRIM, hipCUB, rocThrust, rocWMMA). + +### Artifacts + +| Artifact | Where | Purpose | +| ------------ | ----------------------------------- | ----------------------------- | +| Instrumented | CI bucket, `-coverage` | Stack under test + LLVM tools | +| Profraw | GitHub Actions artifacts, per shard | Raw profiles for aggregation | + +Profraw names include the project, GPU family, and shard index. + +### Report generation + +`aggregate_coverage` runs under `if: !cancelled()` so partial reports survive +shard failures. It downloads profraw artifacts, reinstalls the run's artifacts +(LLVM tools must match the compiler that produced the profiles), then runs +`merge_coverage_report.py`. A report that measured nothing fails rather than +publishing 0%: + +- No profraw files: the tests didn't run, or the wrong library was loaded. +- No matching objects: `object_globs` doesn't match the project's install + layout. +- Profiles with no counters: the processes loaded no instrumented code. Either + the coverage flags never reached the compiler, or every process crashed + before writing its profile. +- Objects with no coverage mapping (`llvm-cov`: "no coverage data found"): the + project's coverage option did not take effect. +- No line in the reported objects ran: the tests loaded a different copy of the + project, or never reached it. + +The `coverage-report--` artifact: + +| File | What it is | +| ---------------------- | -------------------------- | +| `coverage.info` | lcov for Codecov | +| `coverage_summary.txt` | per-file table with totals | +| `html/index.html` | browsable annotated report | + +Totals are also written to the job summary page. The HTML step runs +`fetch_sources.py` and remaps paths with `--path-equivalence`; it is +`continue-on-error` (losing it costs only the annotated view). Codecov is +skipped, not failed, when `CODECOV_TOKEN` is absent. + +The annotated view shows each file once, its counts summed over every template +instantiation as in the lcov. With llvm-cov's per-instantiation views, demangled, +rocWMMA's report was 3.3GB and rendering it outgrew the 16GB hosted runner the +report job runs on. + +## Selecting projects to run + +`projects_to_test` takes comma-separated project names or group aliases +(`rocm_libraries_all`, `rocm_systems_all`, `all`). Empty selects every project +with a build job (currently `rocm_libraries_all`). + +`rccl` and `rocshmem` live in `comm-libs`, which has no build job; they are +excluded from the default and rejected (not just dropped) when named or reached +via an alias. See [Adding a stage](#adding-a-stage). + +`hipblaslt` and `hiptensor` are excluded from the default and aliases because +their instrumented builds fail. Naming one still works and prints the reason — +retesting is how the block gets noticed as fixed. See +[Blocked projects](#blocked-projects). + +Aliases resolve to concrete names before the matrix is built; nothing downstream +sees them. + +## Adding a project + +Add an entry to `COVERAGE_PROJECTS` in `configure_coverage_ci.py`. + +`coverage_option` must name the CMake option the project implements, and it must +select LLVM instrumentation (not gcov, which writes `.gcda` files this pipeline +can't read). Names vary: most use `BUILD_CODE_COVERAGE` or `CODE_COVERAGE`; +RCCL uses `ENABLE_CODE_COVERAGE`; rocWMMA and hipTensor prefix the project +(`ROCWMMA_CODE_COVERAGE`). A wrong name still configures and builds cleanly; +it only shows up as a report with no coverage data. + +Verify a local instrumented build produces a non-empty report before committing +the entry. Set `source_repo` (`ROCM_LIBRARIES` or `ROCM_SYSTEMS`) to route the +project into the correct group alias; membership lists are generated at configure +time, so no CMake edit is needed. `object_globs` must match what the project +actually installs — a miss fails the report job rather than publishing zero. A +glob prefixed with `!` removes its matches, for test binaries that share an +install directory with a sibling project's (rocPRIM's `bin/test_*` would +otherwise include hipCUB's `bin/test_hipcub_*`). + +## Blocked projects + +A `blocked_reason` alongside `coverage_option` means the instrumented build is +broken. The project is dropped from the default and group aliases (including +`--emit-cmake` output) but can still be named explicitly, printing the reason. +A stage is built as a unit, so one blocked project takes the whole stage down. +Drop `blocked_reason` once the upstream fix lands. + +`hipblaslt` — its coverage branch links against bare `rocroller` while the rest +of its CMakeLists uses `roc::rocroller`. Under TheRock rocRoller is a separate +subproject; the bare name reaches the linker as `-lrocroller` with no `-L`. + +`hiptensor` — upstream applies coverage as a blanket `add_compile_options` on a +composable_kernel-based library. The coverage mapping data exceeds 2GB, failing +the link with out-of-range `R_X86_64_PC32` relocations. A larger code model is +the likely fix. + +## Adding a stage + +`configure_coverage_ci.py` rejects selections reaching beyond `compiler-runtime` +and `math-libs` (`BUILDABLE_STAGES`). To add a stage: add a build job in +`multi_arch_ci_coverage_linux.yml` and add the stage name to `BUILDABLE_STAGES`. +Order matters — a stage needs its inbound artifacts built first. `comm-libs` +(needed by `rccl`/`rocshmem`) depends on `emulation`; that's two jobs, not one. +Consult `build_topology`; a missing inbound artifact won't surface until deep in +the build. diff --git a/docs/development/development_guide.md b/docs/development/development_guide.md index 0d0f6343d2e..0d355ef0385 100644 --- a/docs/development/development_guide.md +++ b/docs/development/development_guide.md @@ -235,6 +235,25 @@ These CMake flags allow overriding the global options for individual sub-project > Only the on/off switch described here is implemented today; the name may change once that RFC is > finalized and approved. +- `{PROJECT}_ENABLE_COVERAGE` + +Builds a subproject with source based code coverage instrumentation, provided +that the subproject itself implements the flag. Unlike the flags above, the +project name must be upper case, matching the usual CMake convention for +definitions: `HIPDNN_ENABLE_COVERAGE`, not `hipDNN_ENABLE_COVERAGE`. + +```bash + # Instrument a single project + -DHIPRAND_ENABLE_COVERAGE=ON + + # Instrument several projects + -DHIPRAND_ENABLE_COVERAGE=ON -DROCBLAS_ENABLE_COVERAGE=ON +``` + +Instrumented device code additionally needs the ROCm profiling runtime, which +comes from the compiler build rather than from a top level TheRock option. See +[Code coverage](code_coverage.md) for the full workflow. + ### Additional CMake developer ergonomic flags We add developer ergonomic flags as needed in order to support project-wide development activities. This currently includes: diff --git a/docs/development/workflow_outputs.md b/docs/development/workflow_outputs.md index a901ef5420f..78bec71ab29 100644 --- a/docs/development/workflow_outputs.md +++ b/docs/development/workflow_outputs.md @@ -37,9 +37,20 @@ s3://{bucket}/{external_repo}{run_id}-{platform}/ | --------------- | ------------------------------- | --------------------------------------------------------------------- | | `bucket` | `therock-ci-artifacts` | Selected based on repo, fork status, release type | | `external_repo` | `""` or `"githubuser-TheRock/"` | Non-empty for forks and non-TheRock repos (format: `{owner}-{repo}/`) | -| `run_id` | `12345678901` | GitHub Actions workflow run ID | +| `run_id` | `12345678901` | GitHub Actions workflow run ID; see [Variant builds](#variant-builds) | | `platform` | `linux` or `windows` | Build platform | +### Variant builds + +A variant build that shares its workflow run with a regular build publishes +under `{run_id}-{variant}` instead, because both produce artifacts with the same +names. A caller that builds a regular stack and a coverage-instrumented one in +the same run gets them under `12345678901-linux/` and +`12345678901-coverage-linux/`. Pass the suffixed id as `artifact_run_id` to +`multi_arch_build_portable_linux_artifacts.yml`, and as `--run-id` to anything +that reads the variant's outputs. A lookup through the GitHub API uses the id +without its suffix. + ### Directory structure There are two CI pipeline architectures with different log layouts: @@ -269,7 +280,8 @@ the GitHub API to fetch workflow run metadata (for fork detection and bucket cutover dating). Most callers running inside their own CI workflow do not need this — environment variables (`GITHUB_REPOSITORY`, `IS_PR_FROM_FORK`) suffice. Set `lookup_workflow_run=True` when looking up another repository's workflow -run, e.g. when fetching artifacts. +run, e.g. when fetching artifacts. A `-{variant}` suffix on `run_id` is +dropped for the lookup. ### StorageBackend