diff --git a/projects/hip/.azuredevops/rocm-ci.yml b/projects/hip/.azuredevops/rocm-ci.yml new file mode 100644 index 0000000000..c7cc075ff0 --- /dev/null +++ b/projects/hip/.azuredevops/rocm-ci.yml @@ -0,0 +1,75 @@ +resources: + repositories: + - repository: pipelines_repo + type: github + endpoint: ROCm + name: ROCm/ROCm + - repository: matching_repo + type: github + endpoint: ROCm + name: ROCm/clr + ref: $(Build.SourceBranch) + - repository: hipother_repo + type: github + endpoint: ROCm + name: ROCm/hipother + ref: $(Build.SourceBranch) + pipelines: + - pipeline: rocr-runtime_pipeline + source: \ROCR-Runtime + trigger: + branches: + include: + - amd-staging + - amd-mainline + - pipeline: rocprofiler-register_pipeline + source: \rocprofiler-register + trigger: + branches: + include: + - amd-staging + - amd-mainline + +variables: +- group: common +- template: /.azuredevops/variables-global.yml@pipelines_repo + +trigger: + batch: true + branches: + include: + - amd-staging + - amd-mainline + paths: + exclude: + - docs + - '.github' + - '.jenkins' + - '.*.yaml' + - CODEOWNERS + - Jenkinsfile + - LICENSE.txt + - '**/*.md' + - VERSION + +pr: + autoCancel: true + branches: + include: + - amd-staging + - amd-mainline + paths: + exclude: + - docs + - '.github' + - '.jenkins' + - '.*.yaml' + - CODEOWNERS + - Jenkinsfile + - LICENSE.txt + - '**/.md' + - VERSION + drafts: false + +jobs: + - template: ${{ variables.CI_COMPONENT_PATH }}/HIP.yml@pipelines_repo diff --git a/projects/hip/.clang-format b/projects/hip/.clang-format new file mode 100644 index 0000000000..5572a72cdd --- /dev/null +++ b/projects/hip/.clang-format @@ -0,0 +1,10 @@ +Language: Cpp +BasedOnStyle: Google +AlignEscapedNewlinesLeft: false +AlignOperands: false +ColumnLimit: 100 +AlwaysBreakTemplateDeclarations: false +DerivePointerAlignment: false +IndentFunctionDeclarationAfterType: false +MaxEmptyLinesToKeep: 2 +SortIncludes: false diff --git a/projects/hip/.gitattributes b/projects/hip/.gitattributes new file mode 100644 index 0000000000..55bb72f281 --- /dev/null +++ b/projects/hip/.gitattributes @@ -0,0 +1,21 @@ +# Set the default behavior, in case people don't have core.autolf set. +* text=auto + +# Explicitly declare text files you want to always be normalized and converted +# to have LF line endings on checkout. +*.c text eol=lf +*.cpp text eol=lf +*.cc text eol=lf +*.h text eol=lf +*.hpp text eol=lf +*.txt text eol=lf + +# Define files to support auto-remove trailing white space +# Need to run the command below, before add modified file(s) to the staging area +# git config filter.trimspace.clean 'sed -e "s/[[:space:]]*$//g"' +*.cpp filter=trimspace +*.c filter=trimspace +*.h filter=trimspacecpp +*.hpp filter=trimspace +*.md filter=trimspace + diff --git a/projects/hip/.github/dependabot.yml b/projects/hip/.github/dependabot.yml new file mode 100644 index 0000000000..470629cc14 --- /dev/null +++ b/projects/hip/.github/dependabot.yml @@ -0,0 +1,19 @@ +# To get started with Dependabot version updates, you'll need to specify which +# package ecosystems to update and where the package manifests are located. +# Please see the documentation for all configuration options: +# https://docs.github.com/github/administering-a-repository/configuration-options-for-dependency-updates + +version: 2 +updates: + - package-ecosystem: "pip" # See documentation for possible values + directory: "/docs/sphinx" # Location of package manifests + open-pull-requests-limit: 10 + schedule: + interval: "daily" + labels: + - "documentation" + - "dependencies" + - "ci:docs-only" + target-branch: "docs/develop" + reviewers: + - "samjwu" diff --git a/projects/hip/.github/hooks/clang-format-check.sh b/projects/hip/.github/hooks/clang-format-check.sh new file mode 100644 index 0000000000..e417133217 --- /dev/null +++ b/projects/hip/.github/hooks/clang-format-check.sh @@ -0,0 +1,56 @@ +#!/usr/bin/env bash + +set -euo pipefail + +RANGE="" + +while [[ $# -gt 0 ]]; do + echo $1 + echo $2 + case "$1" in + --range) + RANGE="$2" + shift 2 + ;; + *) + echo "Unknown arg $1" >&2 + exit 64 + ;; + esac +done + +regex='\.(c|cc|cpp|cxx|h|hh|hpp|hxx)$' + +if [[ -n $RANGE ]]; then + files=$(git diff --name-only "$RANGE" | grep -E "$regex" || true) +else + files=$(git diff --cached --name-only --diff-filter=ACMR | grep -E "$regex" || true) +fi +echo "Checking $files" +[[ -z $files ]] && exit 0 + +clang_bin="${CLANG_FORMAT:-clang-format}" +if ! command -v "$clang_bin" >/dev/null 2>&1; then + if [[ -x "/c/Program Files/LLVM/bin/clang-format.exe" ]]; then + clang_bin="/c/Program Files/LLVM/bin/clang-format.exe" + fi +fi + +clang_format_diff="${CLANG_FORMAT_DIFF:-clang-format-diff}" +if ! command -v "$clang_format_diff" >/dev/null 2>&1; then + if [[ -x "/c/Program Files/LLVM/share/clang/clang-format-diff.py" ]]; then + clang_format_diff="/c/Program Files/LLVM/share/clang/clang-format-diff.py" + fi +fi + +for file in $files; do + echo "Checking lines of $file" + + if [[ -n $RANGE ]]; then + diff_output=$(git diff -U0 "$RANGE" -- "$file") + else + diff_output=$(git diff -U0 --cached -- "$file") + fi + + echo "$diff_output" | "$clang_format_diff" -style=file -fallback-style=none -p1 +done diff --git a/projects/hip/.github/hooks/pre-commit b/projects/hip/.github/hooks/pre-commit new file mode 100644 index 0000000000..f42d5a3174 --- /dev/null +++ b/projects/hip/.github/hooks/pre-commit @@ -0,0 +1,2 @@ +#!/usr/bin/env bash +exec "$(git rev-parse --show-toplevel)/.github/hooks/clang-format-check.sh" diff --git a/projects/hip/.github/palamida.yml b/projects/hip/.github/palamida.yml new file mode 100644 index 0000000000..2a54378973 --- /dev/null +++ b/projects/hip/.github/palamida.yml @@ -0,0 +1,5 @@ +disabled: false +scmId: gh-emu-rocm +branchesToScan: + - amd-staging + - amd-mainline diff --git a/projects/hip/.github/pull_request_template.md b/projects/hip/.github/pull_request_template.md new file mode 100644 index 0000000000..3585d2a02f --- /dev/null +++ b/projects/hip/.github/pull_request_template.md @@ -0,0 +1,36 @@ +## Associated JIRA ticket number/Github issue number + + +## What type of PR is this? (check all applicable) + +- [ ] Refactor +- [ ] Feature +- [ ] Bug Fix +- [ ] Optimization +- [ ] Documentation Update +- [ ] Continuous Integration + +## What were the changes? + + + +## Why are these changes needed? + + + +## Updated CHANGELOG? + + + +- [ ] Yes +- [ ] No, Does not apply to this PR. + +## Added/Updated documentation? + +- [ ] Yes +- [ ] No, Does not apply to this PR. + +## Additional Checks + +- [ ] I have added tests relevant to the introduced functionality, and the unit tests are passing locally. +- [ ] Any dependent changes have been merged. diff --git a/projects/hip/.github/scripts/validate_pr_description.py b/projects/hip/.github/scripts/validate_pr_description.py new file mode 100644 index 0000000000..eb282acffd --- /dev/null +++ b/projects/hip/.github/scripts/validate_pr_description.py @@ -0,0 +1,76 @@ +import os, re, sys +from typing import List, Optional + + +def is_checkbox(line: str) -> bool: + return bool(re.match(r"^\s*-\s*\[[ xX]\]\s*.+", line)) + + +def is_checked(line: str) -> bool: + return bool(re.match(r"^\s*-\s*\[[xX]\]\s*.+", line)) + + +def is_comment(line: str) -> bool: + return bool(re.match(r"^\s*\s*$", line)) + + +def text_clean(lines: List[str]) -> str: + text = [line for line in lines if not is_comment(line)] + return "".join("".join(text).strip().split()) + + +def validate_section(section_name: str, lines: List[str]) -> Optional[str]: + has_checkboxes = any(is_checkbox(line) for line in lines) + if has_checkboxes: + if not any(is_checked(line) for line in lines): + return f"Section {section_name} is a checklist without selections" + return None + if not text_clean(lines): + return f"Section {section_name} is empty text section" + return None + + +def check_description(description: str) -> List[str]: + if not description: + # pull_request_template is not merged yet, so treat as valid for now + return [] + # return ["PR description is empty"] + + sections = [] + current_section = None + current_lines = [] + errors = [] + + for line in description.splitlines(): + header_match = re.match(r"^\s*##\s*(.+?)\s*$", line) + if header_match: + if current_section: + sections.append((current_section, current_lines)) + current_section = header_match.group(1) + current_lines = [] + elif current_section: + current_lines.append(line) + + if current_section: + sections.append((current_section, current_lines)) + + if not sections: + return ["No sections available, template is empty"] + + for section_name, section_lines in sections: + error = validate_section(section_name, section_lines) + if error: + errors.append(error) + + return errors + + +if __name__ == "__main__": + pr_description = os.getenv("PR_DESCRIPTION", "") + + errors = check_description(pr_description) + if not errors: + print("All good") + exit(0) + print("\n".join(errors)) + exit(1) diff --git a/projects/hip/.github/workflows/clang-format.yml b/projects/hip/.github/workflows/clang-format.yml new file mode 100644 index 0000000000..0298b5fc76 --- /dev/null +++ b/projects/hip/.github/workflows/clang-format.yml @@ -0,0 +1,22 @@ +name: Clang format check +on: + pull_request: + types: [synchronize, opened] + +jobs: + format: + runs-on: AMD-ROCm-Internal-dev1 + steps: + - uses: actions/checkout@v4 + with: + fetch-depth: 0 + + - name: Install clang-format + run: | + sudo apt update && sudo apt install -y clang-format + + - name: Run clang-format-check + id: clang-format + run: | + chmod +x .github/hooks/clang-format-check.sh + ./.github/hooks/clang-format-check.sh --range "${{ github.event.pull_request.base.sha }}..${{ github.event.pull_request.head.sha }}" diff --git a/projects/hip/.github/workflows/keyword-check.yml b/projects/hip/.github/workflows/keyword-check.yml new file mode 100644 index 0000000000..12108cee14 --- /dev/null +++ b/projects/hip/.github/workflows/keyword-check.yml @@ -0,0 +1,73 @@ +name: Keywords checker + +on: + pull_request: + types: [opened, synchronize, reopened, edited] + branches: + - amd-staging + workflow_dispatch: + +jobs: + check-keywords: + runs-on: AMD-ROCm-Internal-dev1 + env: + KEYWORDS: ${{ vars.KEYWORDS }} + + steps: + - name: Checkout code + uses: actions/checkout@v3 + with: + fetch-depth: 0 + + - name: Check keywords + run: | + set -e + + if [ -z "$KEYWORDS" ]; then + echo "No keywords set. Skipping check" + exit 0 + fi + + IFS=',' read -ra KEYWORDS_ARRAY <<< "$KEYWORDS" + echo "Checking against list of keywords: ${KEYWORDS_ARRAY[*]}" + + MATCHED=0 + BASE_BRANCH=${{github.event.pull_request.base.ref}} + HEAD_BRANCH=${{github.event.pull_request.head.ref}} + PR_TITLE="${{ github.event.pull_request.title }}" + + for file in $(git diff --name-only origin/$BASE_BRANCH..origin/$HEAD_BRANCH); do + if [ -f "$file" ]; then + for keyword in "${KEYWORDS_ARRAY[*]}"; do + grep -in -E "${keyword}" "$file" | while IFS= read -r line; do + echo "Matched in '$file': $line" + MATCHED=1 + done + done + fi + done + + for commit in $(git log --format=%H origin/$BASE_BRANCH..origin/$HEAD_BRANCH); do + msg=$(git log -1 --format=%B "$commit") + for keyword in "${KEYWORDS_ARRAY[*]}"; do + if echo "$msg" | grep -i -q "$keyword"; then + echo "Match in commit $commit: $msg" + MATCHED=1 + fi + done + done + + for keyword in "${KEYWORDS_ARRAY[*]}"; do + if echo "$PR_TITLE" | grep -i -q "$keyword"; then + echo "Match in PR title" + MATCHED=1 + fi + done + + if [ "$MATCHED" -eq 1 ]; then + echo "Keywords found, please see diagnostics higher" + exit 1 + else + echo "No keywords found" + exit 0 + fi diff --git a/projects/hip/.github/workflows/kws-caller.yml b/projects/hip/.github/workflows/kws-caller.yml new file mode 100644 index 0000000000..c0f4f26807 --- /dev/null +++ b/projects/hip/.github/workflows/kws-caller.yml @@ -0,0 +1,15 @@ +name: Rocm Validation Suite KWS +on: + push: + branches: [amd-staging, amd-mainline] + pull_request: + types: [opened, synchronize, reopened] + workflow_dispatch: +jobs: + kws: + if: ${{ github.event_name == 'pull_request' }} + uses: AMD-ROCm-Internal/rocm_ci_infra/.github/workflows/kws.yml@mainline + secrets: inherit + with: + pr_number: ${{github.event.pull_request.number}} + base_branch: ${{github.base_ref}} diff --git a/projects/hip/.github/workflows/linting.yml b/projects/hip/.github/workflows/linting.yml new file mode 100644 index 0000000000..46cf268483 --- /dev/null +++ b/projects/hip/.github/workflows/linting.yml @@ -0,0 +1,20 @@ +name: Linting + +on: + push: + branches: + - develop + - main + - 'docs/*' + - 'roc**' + pull_request: + branches: + - develop + - main + - 'docs/*' + - 'roc**' + +jobs: + call-workflow-passing-data: + name: Documentation + uses: ROCm/rocm-docs-core/.github/workflows/linting.yml@develop diff --git a/projects/hip/.github/workflows/pr-title-validate.yml b/projects/hip/.github/workflows/pr-title-validate.yml new file mode 100644 index 0000000000..f68440d948 --- /dev/null +++ b/projects/hip/.github/workflows/pr-title-validate.yml @@ -0,0 +1,46 @@ +name: Validate PR Title + +on: + pull_request: + types: [opened, edited, synchronize, reopened] + +jobs: + validate-pr-title: + runs-on: ubuntu-latest + steps: + - name: Check PR Title + id: check-pr-title + run: | + PR_TITLE="${{ github.event.pull_request.title }}" + + if [[ ! "$PR_TITLE" =~ ^SWDEV-[0-9]+ ]]; then + echo "::error::PR title must start with a Jira ticket ID, SWDEV-" + exit 1 + else + echo "PR title is valid" + fi + + validate-commit-messages: + runs-on: AMD-ROCm-Internal-dev1 + steps: + - name: Checkout code + uses: actions/checkout@v3 + with: + fetch-depth: 0 + + - name: Check all commit messages + id: validate-commit-messags + run: | + COMMITS=$(git log --format="%H %s" origin/${{ github.event.pull_request.base.ref }}..origin/${{ github.event.pull_request.head.ref }}) + echo "$COMMITS" + echo "$COMMITS" | while read -r hash message; do + echo -e "$hash $message\n " + if [[ "$message" =~ ^SWDEV-[0-9]+ ]]; then + echo "Valid JIRA ticket format" + elif [[ "$message" =~ ^Merge\ branch ]]; then + echo "Merge commits are allowed" + else + echo "::error:: $hash commit should start with Jira ticket ID, SWDEV- or be a merge commit" + exit 1 + fi + done diff --git a/projects/hip/.github/workflows/rocm-ci-caller.yml b/projects/hip/.github/workflows/rocm-ci-caller.yml new file mode 100644 index 0000000000..2a8d0a6c15 --- /dev/null +++ b/projects/hip/.github/workflows/rocm-ci-caller.yml @@ -0,0 +1,25 @@ +name: ROCm CI Caller +on: + pull_request: + branches: [amd-staging, amd-npi-next, release/rocm-rel-*, amd-mainline] + types: [opened, reopened, synchronize] + push: + branches: [amd-mainline] + workflow_dispatch: + issue_comment: + types: [created] + +jobs: + call-workflow: + if: github.event_name != 'issue_comment' ||(github.event_name == 'issue_comment' && github.event.issue.pull_request && (startsWith(github.event.comment.body, '!verify') || startsWith(github.event.comment.body, '!linux-hip-psdb') || startsWith(github.event.comment.body, '!verify release') || startsWith(github.event.comment.body, '!verify retest'))) + uses: AMD-ROCm-Internal/rocm_ci_infra/.github/workflows/rocm_ci.yml@mainline + secrets: inherit + with: + input_sha: ${{github.event_name == 'pull_request' && github.event.pull_request.head.sha || (github.event_name == 'push' && github.sha) || (github.event_name == 'issue_comment' && github.event.issue.pull_request.head.sha) || github.sha}} + input_pr_num: ${{github.event_name == 'pull_request' && github.event.pull_request.number || (github.event_name == 'issue_comment' && github.event.issue.number) || 0}} + input_pr_url: ${{github.event_name == 'pull_request' && github.event.pull_request.html_url || (github.event_name == 'issue_comment' && github.event.issue.pull_request.html_url) || ''}} + input_pr_title: ${{github.event_name == 'pull_request' && github.event.pull_request.title || (github.event_name == 'issue_comment' && github.event.issue.pull_request.title) || ''}} + repository_name: ${{ github.repository }} + base_ref: ${{github.event_name == 'pull_request' && github.event.pull_request.base.ref || (github.event_name == 'issue_comment' && github.event.issue.pull_request.base.ref) || github.ref}} + trigger_event_type: ${{ github.event_name }} + comment_text: ${{ github.event_name == 'issue_comment' && github.event.comment.body || '' }} diff --git a/projects/hip/.github/workflows/validate-pr-description.yml b/projects/hip/.github/workflows/validate-pr-description.yml new file mode 100644 index 0000000000..d9b12b4ba6 --- /dev/null +++ b/projects/hip/.github/workflows/validate-pr-description.yml @@ -0,0 +1,22 @@ +name: Validate PR desription + +on: + pull_request: + types: [opened, edited, synchronize] + +jobs: + validate-pr-description: + runs-on: AMD-ROCm-Internal-dev1 + steps: + - name: Checkout repository + uses: actions/checkout@v4 + + - name: Set up Python + uses: actions/setup-python@v5 + with: + python-version: "3.13" + + - name: Validate PR description + env: + PR_DESCRIPTION: ${{ github.event.pull_request.body }} + run: python .github/scripts/validate_pr_description.py diff --git a/projects/hip/.gitignore b/projects/hip/.gitignore new file mode 100644 index 0000000000..ffb0b5f8c0 --- /dev/null +++ b/projects/hip/.gitignore @@ -0,0 +1,22 @@ +.* +!.gitignore +!.spellcheck.local.yaml +*.o +*.exe +*.swp +lib +packages +build +bin/hipInfo +bin/hipBusBandwidth +bin/hipDispatchLatency +bin/hipify-clang +tags +samples/0_Intro/module_api/runKernel.hip.out +samples/0_Intro/module_api/vcpy_isa.code +samples/0_Intro/module_api/vcpy_isa.hsaco +samples/0_Intro/module_api/vcpy_kernel.co +samples/0_Intro/module_api/vcpy_kernel.code +samples/1_Utils/hipInfo/hipInfo +samples/1_Utils/hipBusBandwidth/hipBusBandwidth +samples/1_Utils/hipDispatchLatency/hipDispatchLatency diff --git a/projects/hip/.jenkins/Jenkinsfile b/projects/hip/.jenkins/Jenkinsfile new file mode 100644 index 0000000000..b345eb0e4e --- /dev/null +++ b/projects/hip/.jenkins/Jenkinsfile @@ -0,0 +1,110 @@ +def hipBuildTest(String backendLabel) { + node(backendLabel) { + stage("SYNC - ${backendLabel}") { + + // Checkout hip repository with the PR patch + dir("${WORKSPACE}/hip") { + checkout scm + env.HIP_DIR = "${WORKSPACE}" + "/hip" + } + + // Clone hip-tests repository + dir("${WORKSPACE}/hip-tests") { + git branch: 'develop', + url: 'https://github.com/ROCm-Developer-Tools/hip-tests' + env.HIP_TESTS_DIR = "${WORKSPACE}" + "/hip-tests" + } + + // Clone clr repository + dir("${WORKSPACE}/clr") { + git branch: 'develop', + credentialsId: 'branch-credentials', + url: 'https://github.com/ROCm-Developer-Tools/clr' + env.CLR_DIR = "${WORKSPACE}" + "/clr" + } + + // Clone hipcc repspoitory + dir("${WORKSPACE}/hipcc") { + git branch: 'develop', + credentialsId: 'branch-credentials', + url: 'https://github.com/ROCm-Developer-Tools/hipcc' + env.HIPCC_DIR = "${WORKSPACE}" + "/hipcc" + } + } + + stage("BUILD HIP - ${backendLabel}") { + // Running the build on clr workspace + dir("${WORKSPACE}/clr") { + sh """#!/usr/bin/env bash + set -x + rm -rf build + mkdir -p build + cd build + # Check if backend label contains string "amd" or backend host is a server with amd gpu + if [[ $backendLabel =~ amd ]]; then + cmake -DCLR_BUILD_HIP=ON -DHIP_PATH=\$PWD/install -DHIPCC_BIN_DIR=\$HIPCC_DIR/bin -DHIP_COMMON_DIR=\$HIP_DIR -DCMAKE_PREFIX_PATH="/opt/rocm/" -DCMAKE_INSTALL_PREFIX=\$PWD/install .. + else + cmake -DCLR_BUILD_HIP=ON -DHIP_PLATFORM=nvidia -DHIPCC_BIN_DIR=\$HIPCC_DIR/bin -DHIP_COMMON_DIR=\$HIP_DIR -DCMAKE_INSTALL_PREFIX=\$PWD/install .. + fi + make -j\$(nproc) + make install -j\$(nproc) + """ + } + } + + stage("BUILD HIP TESTS - ${backendLabel}") { + // Running the build on HIP TESTS workspace + dir("${WORKSPACE}/hip-tests") { + env.HIP_PATH = "${CLR_DIR}" + "/build/install" + sh """#!/usr/bin/env bash + set -x + rm -rf build + mkdir -p build + cd build + echo "testing $HIP_PATH" + # Check if backend label contains string "amd" or backend host is a server with amd gpu + if [[ $backendLabel =~ amd ]]; then + cmake -DHIP_PLATFORM=amd -DHIP_PATH=\$CLR_DIR/build/install ../catch + else + export HIP_PLATFORM=nvidia + cmake -DHIP_PLATFORM=nvidia -DHIP_PATH=\$CLR_DIR/build/install ../catch + fi + make -j\$(nproc) build_tests + """ + } + } + + timeout(time: 1, unit: 'HOURS') { + stage("TEST - ${backendLabel}") { + dir("${WORKSPACE}/hip-tests") { + sh """#!/usr/bin/env bash + set -x + cd build + if [[ $backendLabel =~ amd ]]; then + ctest --overwrite BuildDirectory=. --output-junit hiptest_output_catch_amd.xml + else + ctest --overwrite BuildDirectory=. --output-junit hiptest_output_catch_nvidia.xml -E 'Unit_hipMemcpyHtoD_Positive_Synchronization_Behavior|Unit_hipMemcpy_Positive_Synchronization_Behavior|Unit_hipFreeNegativeHost' + fi + """ + } + } + } + } +} + +timestamps { + node('external-bootstrap') { + skipDefaultCheckout() + + // labels belonging to each backend - AMD, NVIDIA + String[] labels = ['hip-amd-gfx908-ubu2004', 'hip-nvidia-rtx5000-ubu2004'] + buildMap = [:] + + labels.each { backendLabel -> + echo "backendLabel: ${backendLabel}" + buildMap[backendLabel] = { hipBuildTest(backendLabel) } + } + buildMap['failFast'] = false + parallel buildMap + } +} diff --git a/projects/hip/.markdownlint-cli2.yaml b/projects/hip/.markdownlint-cli2.yaml new file mode 100644 index 0000000000..e3d82c4ab5 --- /dev/null +++ b/projects/hip/.markdownlint-cli2.yaml @@ -0,0 +1,3 @@ +ignores: + - RELEASE.md + - docs/doxygen/mainpage.md diff --git a/projects/hip/.readthedocs.yaml b/projects/hip/.readthedocs.yaml new file mode 100644 index 0000000000..76f837ede8 --- /dev/null +++ b/projects/hip/.readthedocs.yaml @@ -0,0 +1,33 @@ +# Read the Docs configuration file +# See https://docs.readthedocs.io/en/stable/config-file/v2.html for details + +version: 2 + +sphinx: + configuration: docs/conf.py + +formats: [] + +python: + install: + - requirements: docs/sphinx/requirements.txt + +conda: + environment: docs/environment.yml # needed until ubuntu ships doxygen >= 1.9.8 + +build: + os: ubuntu-22.04 + tools: + python: "mambaforge-22.9" # needed until ubuntu ships doxygen >= 1.9.8 + apt_packages: + - "gfortran" # For pre-processing fortran sources + - "graphviz" # For dot graphs in doxygen + jobs: + post_checkout: + - if [ -d ../clr ]; then rm -rf ../clr; fi + - if [ -d ../ROCR-Runtime ]; then rm -rf ../ROCR-Runtime; fi + - git clone --depth=1 --single-branch --branch docs/develop https://github.com/ROCm/clr.git ../clr + - git clone --depth=1 --single-branch --branch master https://github.com/ROCm/ROCR-Runtime.git ../ROCR-Runtime + post_build: + - rm -rf ../clr + - rm -rf ../ROCR-Runtime diff --git a/projects/hip/.spellcheck.local.yaml b/projects/hip/.spellcheck.local.yaml new file mode 100644 index 0000000000..a4adbe401f --- /dev/null +++ b/projects/hip/.spellcheck.local.yaml @@ -0,0 +1,10 @@ +matrix: +- name: Markdown + sources: + - ['!docs/doxygen/mainpage.md'] +- name: reST + sources: + - [] +- name: Cpp + sources: + - ['include/hip/*'] diff --git a/projects/hip/.wordlist.txt b/projects/hip/.wordlist.txt new file mode 100644 index 0000000000..1bca54a941 --- /dev/null +++ b/projects/hip/.wordlist.txt @@ -0,0 +1,190 @@ +.hip_fatbin +ALU +ALUs +AmgX +APU +APUs +AQL +AXPY +asm +Asynchronicity +Asynchrony +asynchrony +backtrace +bfloat +Bitcode +bitcode +bitcodes +blockDim +blockIdx +builtins +Builtins +CAS +clr +compilable +constexpr +coroutines +Ctx +cuBLASLt +cuCtx +CUDA +cuda +cuDNN +cuModule +dataflow +deallocate +decompositions +denormal +Dereferencing +DFT +dll +DirectX +EIGEN +enqueue +enqueues +entrypoint +entrypoints +enum +enums +embeded +extern +fatbin +fatbinary +foundationally +framebuffer +frontends +fnuz +FNUZ +fp +gedit +GPGPU +gridDim +GROMACS +GWS +hardcoded +HC +hcBLAS +HIP-Clang +hipcc +hipCtx +hipexamine +hipified +HIPify +hipModule +hipModuleLaunchKernel +hipother +HIPRTC +hyperthreading +icc +IILE +iGPU +inlined +inplace +interop +interoperation +interoperate +interoperation +Interprocess +interprocess +Intrinsics +intrinsics +IPC +IPs +isa +iteratively +Lapack +latencies +libc +libstdc +lifecycle +linearizing +LOC +LUID +ltrace +makefile +Malloc +malloc +MALU +maxregcount +MiB +memset +multicore +multigrid +multithreading +multitenant +MALU +NaN +NCCL +NDRange +nonnegative +NOP +Numa +ns +Nsight +ocp +omnitrace +overindex +overindexing +oversubscription +overutilized +parallelizable +pipelining +parallelized +pixelated +pragmas +preallocated +preconditioners +predefining +prefetched +preprocessor +printf +profilers +PTX +PyHIP +queryable +prefetching +quad +representable +RMW +rocgdb +rocTX +roundtrip +rst +RTC +RTTI +rvalue +SAXPY +scalarizing +sceneries +shaders +SIMT +sinewave +sinf +SOMA +SPMV +structs +SYCL +syntaxes +texel +texels +threadIdx +tradeoffs +templated +toolkits +transfering +typedefs +ULP +ULPs +unintuitive +UMM +uncoalesced +unmap +unmapped +unmapping +unregister +upscaled +variadic +vulkan +warpSize +WinGDB +zc diff --git a/projects/hip/CODEOWNERS b/projects/hip/CODEOWNERS new file mode 100644 index 0000000000..754e825e9f --- /dev/null +++ b/projects/hip/CODEOWNERS @@ -0,0 +1,2 @@ +* @cpaquot_amdeng @gandryey_amdeng @skudchad_amdeng @lmoriche_amdeng + diff --git a/projects/hip/CONTRIBUTING.md b/projects/hip/CONTRIBUTING.md new file mode 100644 index 0000000000..c16a15bff8 --- /dev/null +++ b/projects/hip/CONTRIBUTING.md @@ -0,0 +1,148 @@ +# Contributing to HIP # + +We welcome contributions to the HIP project. Please follow these details to help ensure your contributions will be successfully accepted. +If you want to contribute to our documentation, refer to {doc}`Contribute to ROCm docs `. + +## Issue Discussion ## + +Please use the [GitHub Issue](https://github.com/ROCm/HIP/issues) tab to notify us of issues. + +* Use your best judgement for issue creation. If your issue is already listed, upvote the issue and + comment or post to provide additional details, such as how you reproduced this issue. +* If you're not sure if your issue is the same, err on the side of caution and file your issue. + You can add a comment to include the issue number (and link) for the similar issue. If we evaluate + your issue as being the same as the existing issue, we'll close the duplicate. +* If your issue doesn't exist, use the issue template to file a new issue. + * When filing an issue, be sure to provide as much information as possible, including script output so + we can collect information about your configuration. This helps reduce the time required to + reproduce your issue. + * Check your issue regularly, as we may require additional information to successfully reproduce the + issue. +* You may also open an issue to ask questions to the maintainers about whether a proposed change + meets the acceptance criteria, or to discuss an idea pertaining to the library. + +## Acceptance Criteria ## + +HIP is a C++ Runtime API and Kernel Language that allows developers to create portable applications for AMD and NVIDIA GPUs from single source code. Contributors wishing to submit new HIP Features (ie functions, classes, types) should resemble CUDA APIs. +Differences or limitations of HIP APIs as compared to CUDA APIs should be clearly documented and described. +Some guidelines are outlined below: + +### Add a new HIP API ### + +* Add a translation to the hipify-clang tool ; many examples abound. + * For stat tracking purposes, place the API into an appropriate stat category ("dev", "mem", "stream", etc). +* Add a inlined NVIDIA implementation for the function in /hipnv/include/hip/nvidia_detail/nvidia_hip_runtime_api.h in the repository [hipother](https://github.com/ROCm/hipother). + * These are typically headers +* Add an HIP definition and Doxygen comments for the function in /include/hip/hip_runtime_api.h, in the repository [hip](https://github.com/ROCm/hip). + * Source implementation typically go in clr/hipamd/src/hip_*.cpp in the reposotory [clr](https://github.com/ROCm/clr). The implementation involves calls to HIP runtime (ie for hipStream_t). + +### Run Unit Tests ### + +For new features or bug fixes, it's mandatory to run associate [hip-tests](https://github.com/ROCm/hip-tests). +Please go to the repo and follow the steps. + +For applications and benchmarks outside the hip-tests environment, developments should use a two-step development flow: + +* #1. Compile, link, and install HIP. See {ref}`Building the HIP runtime` notes. +* #2. Relink the target application to include changes in HIP runtime file. + +## Coding Style ## + +* Code Indentation: + * Tabs should be expanded to spaces. + * Use 4 spaces indentation. +* Capitalization and Naming + * Prefer camelCase for HIP interfaces and internal symbols. Note `HIP_CLANG` uses `_` for separator. + This guideline is not yet consistently followed in HIP code * eventual compliance is aspirational. + * Member variables should begin with a leading "_". This allows them to be easily distinguished from other variables or functions. + +* `{}` placement + * namespace should be on same line as `{` and separated by a space. + * Single-line if statement should still use `{/}` pair (even though C++ does not require). + * For functions, the opening `{` should be placed on a new line. + * For if/else blocks, the opening `{` is placed on same line as the if/else. Use a space to separate `{` from if/else. For example, + +```console + if (foo) { + doFoo() + } else { + doFooElse(); + } +``` + +* Miscellaneous + * All references in function parameter lists should be const. + * "ihip" means internal hip structures. These should not be exposed through the HIP API. + * Keyword TODO refers to a note that should be addressed in long-term. Could be style issue, software architecture, or known bugs. + * FIXME refers to a short-term bug that needs to be addressed. + +* `HIP_INIT_API()` should be placed at the start of each top-level HIP API. This function will make sure the HIP runtime is initialized, and also constructs an appropriate API string for tracing and CodeXL marker tracing. The arguments to HIP_INIT_API should match those of the parent function. +* `hipExtGetLastError()` can be called as the AMD platform specific API, to return error code from last HIP API called from the active host thread. `hipGetLastError()` and `hipPeekAtLastError()` can also return the last error that was returned by any of the HIP runtime calls in the same host thread. +* All HIP environment variables should begin with the keyword HIP_ +Environment variables should be long enough to describe their purpose but short enough so they can be remembered * perhaps 10-20 characters, with 3-4 parts separated by underscores. +To see the list of current environment variables, along with their values, set HIP_PRINT_ENV and run any hip applications on ROCm platform. +HIPCC or other tools may support additional environment variables which should follow the above convention. + +## Pull Request Guidelines ## + +By creating a pull request, you agree to the statements made in the code license section. Your pull request should target the default branch. Our current default branch is the develop branch, which serves as our integration branch. + +Follow existing best practice for writing a good Git commit message. + +Some tips: + http://chris.beams.io/posts/git-commit/ + https://robots.thoughtbot.com/5-useful-tips-for-a-better-commit-message + +In particular : + +* Use imperative voice, ie "Fix this bug", "Refactor the XYZ routine", "Update the doc". + Not : "Fixing the bug", "Fixed the bug", "Bug fix", etc. +* Subject should summarize the commit. Do not end subject with a period. Use a blank line + after the subject. + +### Deliverables ### + +HIP is an open source library. Because of this, we include the following license description at the top of every source file. +If you create new source files in the repository, please include this text in them as well (replacing "xx" with the digits for the current year): + +```cpp +// Copyright (c) 20xx Advanced Micro Devices, Inc. All rights reserved. +// +// Permission is hereby granted, free of charge, to any person obtaining a copy +// of this software and associated documentation files (the "Software"), to deal +// in the Software without restriction, including without limitation the rights +// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +// copies of the Software, and to permit persons to whom the Software is +// furnished to do so, subject to the following conditions: +// +// The above copyright notice and this permission notice shall be included in +// all copies or substantial portions of the Software. +// +// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +// THE SOFTWARE. +``` + +### Process ### + +After you create a PR, you can take a look at a diff of the changes you made using the PR's "Files" tab. + +PRs must pass through the checks and the code review described in the [Acceptance Criteria](#acceptance-criteria) section before they can be merged. + +Checks may take some time to complete. You can view their progress in the table near the bottom of the pull request page. You may also be able to use the links in the table +to view logs associated with a check if it fails. + +During code reviews, another developer will take a look through your proposed change. If any modifications are requested (or further discussion about anything is +needed), they may leave a comment. You can follow up and respond to the comment, and/or create comments of your own if you have questions or ideas. +When a modification request has been completed, the conversation thread about it will be marked as resolved. + +To update the code in your PR (eg. in response to a code review discussion), you can simply push another commit to the branch used in your pull request. + +### Doxygen Editing Guidelines ### + +* bugs should be marked with @bugs near the code where the bug might be fixed. The @bug message will appear in the API description and also in the +doxygen bug list. diff --git a/projects/hip/Jenkinsfile b/projects/hip/Jenkinsfile new file mode 100644 index 0000000000..87bf96d99b --- /dev/null +++ b/projects/hip/Jenkinsfile @@ -0,0 +1,442 @@ +#!/usr/bin/env groovy +// Copyright (c) 2017 - 2021 Advanced Micro Devices, Inc. All Rights Reserved. +// Permission is hereby granted, free of charge, to any person obtaining a copy +// of this software and associated documentation files (the "Software"), to deal +// in the Software without restriction, including without limitation the rights +// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +// copies of the Software, and to permit persons to whom the Software is +// furnished to do so, subject to the following conditions: +// +// The above copyright notice and this permission notice shall be included in +// all copies or substantial portions of the Software. +// +// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +// THE SOFTWARE. + +// Generated from snippet generator 'properties; set job properties' +properties([buildDiscarder(logRotator( + artifactDaysToKeepStr: '', + artifactNumToKeepStr: '', + daysToKeepStr: '', + numToKeepStr: '10')), + disableConcurrentBuilds(), + parameters([booleanParam( name: 'push_image_to_docker_hub', defaultValue: false, description: 'Push hip & hcc image to rocm docker-hub' )]), + [$class: 'CopyArtifactPermissionProperty', projectNames: '*'] + ]) + +//////////////////////////////////////////////////////////////////////// +// -- AUXILLARY HELPER FUNCTIONS + +//////////////////////////////////////////////////////////////////////// +// Return build number of upstream job +@NonCPS +int get_upstream_build_num( ) +{ + def upstream_cause = currentBuild.rawBuild.getCause( hudson.model.Cause$UpstreamCause ) + if( upstream_cause == null) + return 0 + + return upstream_cause.getUpstreamBuild() +} + +//////////////////////////////////////////////////////////////////////// +// Return project name of upstream job +@NonCPS +String get_upstream_build_project( ) +{ + def upstream_cause = currentBuild.rawBuild.getCause( hudson.model.Cause$UpstreamCause ) + if( upstream_cause == null) + return null + + return upstream_cause.getUpstreamProject() +} + +//////////////////////////////////////////////////////////////////////// +// Construct the docker build image name +String docker_build_image_name( ) +{ + return "build-ubuntu-16.04" +} + +//////////////////////////////////////////////////////////////////////// +// Construct the relative path of the build directory +String build_directory_rel( String build_config ) +{ + if( build_config.equalsIgnoreCase( 'release' ) ) + { + return "build/release" + } + else + { + return "build/debug" + } +} + +//////////////////////////////////////////////////////////////////////// +// Lots of images are created above; no apparent way to delete images:tags with docker global variable +def docker_clean_images( String org, String image_name ) +{ + // Check if any images exist first grepping for image names + int docker_images = sh( script: "docker images | grep \"${org}/${image_name}\"", returnStatus: true ) + + // The script returns a 0 for success (images were found ) + if( docker_images == 0 ) + { + // Deleting images can fail, if other projects have built on top of that image and are now dependent on it. + // This should not be treated as a hip build failure. This requires cleanup at a later time, possibly through + // another job + try + { + // Best attempt to run bash script to clean images + // deleting images based on hash seems to be more stable than through name:tag values because of tags + sh "docker images | grep \"${org}/${image_name}\" | awk '{print \$1 \":\" \$2}' | xargs docker rmi" + } + catch( err ) + { + println 'Failed to cleanup a few images; probably the images are used as a base for other images' + currentBuild.result = 'SUCCESS' + } + } +} + +//////////////////////////////////////////////////////////////////////// +// -- BUILD RELATED FUNCTIONS + +//////////////////////////////////////////////////////////////////////// +// Checkout source code, source dependencies and update version number numbers +// Returns a relative path to the directory where the source exists in the workspace +String checkout_and_version( String platform ) +{ + String source_dir_rel = "src" + String source_hip_rel = "${source_dir_rel}/hip" + + stage("${platform} clone") + { + dir( "${source_hip_rel}" ) + { + // checkout hip + checkout([ + $class: 'GitSCM', + branches: scm.branches, + doGenerateSubmoduleConfigurations: scm.doGenerateSubmoduleConfigurations, + extensions: scm.extensions + [[$class: 'CleanCheckout']], + userRemoteConfigs: scm.userRemoteConfigs + ]) + } + } + + return source_hip_rel +} + + +//////////////////////////////////////////////////////////////////////// +// This creates the docker image that we use to build the project in +// The docker images contains all dependencies, including OS platform, to build +def docker_build_image( String platform, String org, String optional_build_parm, String source_hip_rel, String from_image ) +{ + String build_image_name = docker_build_image_name( ) + String dockerfile_name = "dockerfile-build-ubuntu-16.04" + def build_image = null + + stage("${platform} build image") + { + dir("${source_hip_rel}") + { + def user_uid = sh( script: 'id -u', returnStdout: true ).trim() + + // Docker 17.05 introduced the ability to use ARG values in FROM statements + // Docker inspect failing on FROM statements with ARG https://issues.jenkins-ci.org/browse/JENKINS-44836 + // build_image = docker.build( "${org}/${build_image_name}:latest", "--pull -f docker/${dockerfile_name} --build-arg user_uid=${user_uid} --build-arg base_image=${from_image} ." ) + + // JENKINS-44836 workaround by using a bash script instead of docker.build() + sh "docker build -t ${org}/${build_image_name}:latest -f docker/${dockerfile_name} ${optional_build_parm} --build-arg user_uid=${user_uid} --build-arg base_image=${from_image} ." + build_image = docker.image( "${org}/${build_image_name}:latest" ) + } + } + + return build_image +} + +//////////////////////////////////////////////////////////////////////// +// This encapsulates the cmake configure, build and package commands +// Leverages docker containers to encapsulate the build in a fixed environment +def docker_build_inside_image( def build_image, String inside_args, String platform, String optional_configure, String build_config, String source_hip_rel, String build_dir_rel ) +{ + String source_hip_abs = pwd() + "/" + source_hip_rel + + build_image.inside( inside_args ) + { + stage("${platform} make ${build_config}") + { + // The rm command needs to run as sudo because the test steps below create files owned by root + sh """#!/usr/bin/env bash + set -x + rm -rf ${build_dir_rel} + mkdir -p ${build_dir_rel} + cd ${build_dir_rel} + cmake -DCMAKE_BUILD_TYPE=${build_config} -DCMAKE_INSTALL_PREFIX=staging ${optional_configure} ${source_hip_abs} + make -j\$(nproc) + """ + } + + // Cap the maximum amount of testing, in case of hangs + // Excluding hipMultiThreadDevice-pyramid & hipMemoryAllocateCoherentDriver tests from automation; due to its flakiness which requires some investigation + timeout(time: 1, unit: 'HOURS') + { + stage("${platform} unit testing") + { + sh """#!/usr/bin/env bash + set -x + cd ${build_dir_rel} + make install -j\$(nproc) + make build_tests -i -j\$(nproc) + ctest --output-on-failure -E "(hipMultiThreadDevice-pyramid|hipMemoryAllocateCoherentDriver)" + """ + // If unit tests output a junit or xunit file in the future, jenkins can parse that file + // to display test results on the dashboard + // junit "${build_dir_rel}/*.xml" + } + } + + // Only create packages from hcc based builds + if( platform.toLowerCase( ).startsWith( 'rocm-' ) ) + { + stage("${platform} packaging") + { + sh """#!/usr/bin/env bash + set -x + cd ${build_dir_rel} + make package + """ + + // No matter the base platform, all packages have the same name + // Only upload 1 set of packages, so we don't have a race condition uploading packages + if( platform.toLowerCase( ).startsWith( 'rocm-head' ) ) + { + archiveArtifacts artifacts: "${build_dir_rel}/*.deb", fingerprint: true + archiveArtifacts artifacts: "${build_dir_rel}/*.rpm", fingerprint: true + } + } + } + } + + return void +} + +//////////////////////////////////////////////////////////////////////// +// This builds a fresh docker image FROM a clean base image, with no build dependencies included +// Uploads the new docker image to internal artifactory +String docker_upload_artifactory( String hcc_ver, String artifactory_org, String from_image, String source_hip_rel, String build_dir_rel ) +{ + def hip_install_image = null + String image_name = "hip-${hcc_ver}-ubuntu-16.04" + + stage( 'artifactory' ) + { + println "artifactory_org: ${artifactory_org}" + + // We copy the docker files into the bin directory where the .deb lives so that it's a clean build everytime + sh "cp -r ${source_hip_rel}/docker/* ${build_dir_rel}" + + // Docker 17.05 introduced the ability to use ARG values in FROM statements + // Docker inspect failing on FROM statements with ARG https://issues.jenkins-ci.org/browse/JENKINS-44836 + // hip_install_image = docker.build( "${artifactory_org}/${image_name}:${env.BUILD_NUMBER}", "--pull -f ${build_dir_rel}/dockerfile-hip-ubuntu-16.04 --build-arg base_image=${from_image} ${build_dir_rel}" ) + + // JENKINS-44836 workaround by using a bash script instead of docker.build() + sh "docker build -t ${artifactory_org}/${image_name} --pull -f ${build_dir_rel}/dockerfile-hip-ubuntu-16.04 --build-arg base_image=${from_image} ${build_dir_rel}" + hip_install_image = docker.image( "${artifactory_org}/${image_name}" ) + + // The connection to artifactory can fail sometimes, but this should not be treated as a build fail + try + { + // Don't push pull requests to artifactory, these tend to accumulate over time + if( env.BRANCH_NAME.toLowerCase( ).startsWith( 'pr-' ) ) + { + println 'Pull Request (PR-xxx) detected; NOT pushing to artifactory' + } + else + { + docker.withRegistry('http://compute-artifactory:5001', 'artifactory-cred' ) + { + hip_install_image.push( "${env.BUILD_NUMBER}" ) + hip_install_image.push( 'latest' ) + } + } + } + catch( err ) + { + currentBuild.result = 'SUCCESS' + } + } + + return image_name +} + +//////////////////////////////////////////////////////////////////////// +// Uploads the new docker image to the public docker-hub +def docker_upload_dockerhub( String local_org, String image_name, String remote_org ) +{ + stage( 'docker-hub' ) + { + // Do not treat failures to push to docker-hub as a build fail + try + { + sh """#!/usr/bin/env bash + set -x + echo inside sh + docker tag ${local_org}/${image_name} ${remote_org}/${image_name} + """ + + docker_hub_image = docker.image( "${remote_org}/${image_name}" ) + + docker.withRegistry('https://registry.hub.docker.com', 'docker-hub-cred' ) + { + docker_hub_image.push( "${env.BUILD_NUMBER}" ) + docker_hub_image.push( 'latest' ) + } + } + catch( err ) + { + currentBuild.result = 'SUCCESS' + } + } +} + +//////////////////////////////////////////////////////////////////////// +// -- MAIN +// Following this line is the start of MAIN of this Jenkinsfile +String build_config = 'Release' +String job_name = env.JOB_NAME.toLowerCase( ) + +// The following launches 3 builds in parallel: rocm-head, rocm-3.3.x and cuda-10.x +parallel rocm_3_3: +{ + node('hip-rocm') + { + String hcc_ver = 'rocm-3.3.x' + String from_image = 'ci_test_nodes/rocm-3.3.x/ubuntu-16.04:latest' + String inside_args = '--device=/dev/kfd --device=/dev/dri --group-add=video' + + // Checkout source code, dependencies and version files + String source_hip_rel = checkout_and_version( hcc_ver ) + + // Create/reuse a docker image that represents the hip build environment + def hip_build_image = docker_build_image( hcc_ver, 'hip', '', source_hip_rel, from_image ) + + // Print system information for the log + hip_build_image.inside( inside_args ) + { + sh """#!/usr/bin/env bash + set -x + /opt/rocm/bin/rocm_agent_enumerator -t ALL + /opt/rocm/bin/hcc --version + """ + } + + // Conctruct a binary directory path based on build config + String build_hip_rel = build_directory_rel( build_config ); + + // Build hip inside of the build environment + docker_build_inside_image( hip_build_image, inside_args, hcc_ver, '', build_config, source_hip_rel, build_hip_rel ) + + // Clean docker build image + docker_clean_images( 'hip', docker_build_image_name( ) ) + + // After a successful build, upload a docker image of the results + /* + String hip_image_name = docker_upload_artifactory( hcc_ver, job_name, from_image, source_hip_rel, build_hip_rel ) + if( params.push_image_to_docker_hub ) + { + docker_upload_dockerhub( job_name, hip_image_name, 'rocm' ) + docker_clean_images( 'rocm', hip_image_name ) + } + docker_clean_images( job_name, hip_image_name ) + */ + } +}, +rocm_head: +{ + node('hip-rocm') + { + String hcc_ver = 'rocm-head' + String from_image = 'ci_test_nodes/rocm-head/ubuntu-16.04:latest' + String inside_args = '--device=/dev/kfd --device=/dev/dri --group-add=video' + + // Checkout source code, dependencies and version files + String source_hip_rel = checkout_and_version( hcc_ver ) + + // Create/reuse a docker image that represents the hip build environment + def hip_build_image = docker_build_image( hcc_ver, 'hip', '', source_hip_rel, from_image ) + + // Print system information for the log + hip_build_image.inside( inside_args ) + { + sh """#!/usr/bin/env bash + set -x + /opt/rocm/bin/rocm_agent_enumerator -t ALL + /opt/rocm/bin/hcc --version + """ + } + + // Conctruct a binary directory path based on build config + String build_hip_rel = build_directory_rel( build_config ); + + // Build hip inside of the build environment + docker_build_inside_image( hip_build_image, inside_args, hcc_ver, '', build_config, source_hip_rel, build_hip_rel ) + + // Clean docker image + docker_clean_images( 'hip', docker_build_image_name( ) ) + + // After a successful build, upload a docker image of the results + /* + String hip_image_name = docker_upload_artifactory( hcc_ver, job_name, from_image, source_hip_rel, build_hip_rel ) + if( params.push_image_to_docker_hub ) + { + docker_upload_dockerhub( job_name, hip_image_name, 'rocm' ) + docker_clean_images( 'rocm', hip_image_name ) + } + docker_clean_images( job_name, hip_image_name ) + */ + } +}, +cuda_10_x: +{ + node('hip-cuda') + { + //////////////////////////////////////////////////////////////////////// + // Block of string constants customizing behavior for cuda + String nvcc_ver = 'cuda-10.x' + String from_image = 'ci_test_nodes/cuda-10.x/ubuntu-16.04:latest' + String inside_args = '--gpus all'; + + // Checkout source code, dependencies and version files + String source_hip_rel = checkout_and_version( nvcc_ver ) + + // Create/reuse a docker image that represents the hip build environment + def hip_build_image = docker_build_image( nvcc_ver, 'hip', '', source_hip_rel, from_image ) + + // Print system information for the log + hip_build_image.inside( inside_args ) + { + sh """#!/usr/bin/env bash + set -x + nvidia-smi + nvcc --version + """ + } + + // Conctruct a binary directory path based on build config + String build_hip_rel = build_directory_rel( build_config ); + + // Build hip inside of the build environment + docker_build_inside_image( hip_build_image, inside_args, nvcc_ver, "-DHIP_NVCC_FLAGS=--Wno-deprecated-gpu-targets", build_config, source_hip_rel, build_hip_rel ) + + // Clean docker image + docker_clean_images( 'hip', docker_build_image_name( ) ) + } +} diff --git a/projects/hip/LICENSE.txt b/projects/hip/LICENSE.txt new file mode 100644 index 0000000000..a8d7060d44 --- /dev/null +++ b/projects/hip/LICENSE.txt @@ -0,0 +1,19 @@ +Copyright (c) 2008 - 2025 Advanced Micro Devices, Inc. + +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: + +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. + +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. diff --git a/projects/hip/README.md b/projects/hip/README.md new file mode 100644 index 0000000000..57ff69619b --- /dev/null +++ b/projects/hip/README.md @@ -0,0 +1,123 @@ +## What is this repository for? + +**HIP is a C++ Runtime API and Kernel Language that allows developers to create portable applications for AMD and NVIDIA GPUs from single source code.** + +Key features include: + +* HIP is very thin and has little or no performance impact over coding directly in CUDA mode. +* HIP allows coding in a single-source C++ programming language including features such as templates, C++11 lambdas, classes, namespaces, and more. +* HIP allows developers to use the "best" development environment and tools on each target platform. +* The [HIPIFY](https://github.com/ROCm/HIPIFY/blob/amd-staging/README.md) tools automatically convert source from CUDA to HIP. +* Developers can specialize for the platform (CUDA or AMD) to tune for performance or handle tricky cases. + +New projects can be developed directly in the portable HIP C++ language and can run on either NVIDIA or AMD platforms. Additionally, HIP provides porting tools which make it easy to port existing CUDA codes to the HIP layer, with no loss of performance as compared to the original CUDA application. HIP is not intended to be a drop-in replacement for CUDA, and developers should expect to do some manual coding and performance tuning work to complete the port. + +> [!NOTE] +> The published documentation is available at [HIP documentation](https://rocm.docs.amd.com/projects/HIP/en/latest/index.html) in an organized, easy-to-read format, with search and a table of contents. The documentation source files reside in the `HIP/docs` folder of this GitHub repository. As with all ROCm projects, the documentation is open source. For more information on contributing to the documentation, see [Contribute to ROCm documentation](https://rocm.docs.amd.com/en/latest/contribute/contributing.html). + +## DISCLAIMER + +The information presented in this document is for informational purposes only and may contain technical inaccuracies, omissions, and typographical errors. The information contained herein is subject to change and may be rendered inaccurate for many reasons, including but not limited to product and roadmap changes, component and motherboard versionchanges, new model and/or product releases, product differences between differing manufacturers, software changes, BIOS flashes, firmware upgrades, or the like. Any computer system has risks of security vulnerabilities that cannot be completely prevented or mitigated.AMD assumes no obligation to update or otherwise correct or revise this information. However, AMD reserves the right to revise this information and to make changes from time to time to the content hereof without obligation of AMD to notify any person of such revisions or changes.THIS INFORMATION IS PROVIDED ‘AS IS.” AMD MAKES NO REPRESENTATIONS OR WARRANTIES WITH RESPECT TO THE CONTENTS HEREOF AND ASSUMES NO RESPONSIBILITY FOR ANY INACCURACIES, ERRORS, OR OMISSIONS THAT MAY APPEAR IN THIS INFORMATION. AMD SPECIFICALLY DISCLAIMS ANY IMPLIED WARRANTIES OF NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR ANY PARTICULAR PURPOSE. IN NO EVENT WILL AMD BE LIABLE TO ANY PERSON FOR ANY RELIANCE, DIRECT, INDIRECT, SPECIAL, OR OTHER CONSEQUENTIAL DAMAGES ARISING FROM THE USE OF ANY INFORMATION CONTAINED HEREIN, EVEN IF AMD IS EXPRESSLY ADVISED OF THE POSSIBILITY OF SUCH DAMAGES. AMD, the AMD Arrow logo, and combinations thereof are trademarks of Advanced Micro Devices, Inc. Other product names used in this publication are for identification purposes only and may be trademarks of their respective companies. + +© 2023 Advanced Micro Devices, Inc. All Rights Reserved. + +## Repository branches + +The HIP repository maintains several branches. The branches that are of importance are: + +* develop branch: This is the default branch, on which the new features are still under development and visible. While this maybe of interest to many, it should be noted that this branch and the features under development might not be stable. +* Main branch: This is the stable branch. It is up to date with the latest release branch, for example, if the latest HIP release is rocm-4.3, main branch will be the repository based on this release. +* Release branches. These are branches corresponding to each ROCM release, listed with release tags, such as rocm-4.2, rocm-4.3, etc. + +## Release tagging + +HIP releases are typically naming convention for each ROCM release to help differentiate them. + +* rocm x.yy: These are the stable releases based on the ROCM release. + This type of release is typically made once a month.* + +## How do I get set up? + +See the [Installation](docs/install/install.rst) notes. + +## Simple Example + +The HIP API includes functions such as hipMalloc, hipMemcpy, and hipFree. +Programmers familiar with CUDA will also be able to quickly learn and start coding with the HIP API. +Compute kernels are launched with the "hipLaunchKernelGGL" macro call. +Here is simple example showing a snippet of HIP API code: + +```cpp +hipMalloc(&A_d, Nbytes); +hipMalloc(&C_d, Nbytes); + +hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice); + +const unsigned blocks = 512; +const unsigned threadsPerBlock = 256; +hipLaunchKernelGGL(vector_square, /* compute kernel*/ + dim3(blocks), dim3(threadsPerBlock), 0/*dynamic shared*/, 0/*stream*/, /* launch config*/ + C_d, A_d, N); /* arguments to the compute kernel */ + +hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost); +``` + +The HIP kernel language defines builtins for determining grid and block coordinates, math functions, short vectors, +atomics, and timer functions. +It also specifies additional defines and keywords for function types, address spaces, and optimization controls (See the [HIP C++ Language Extensions](docs/how-to/hip_cpp_language_extensions.rst) for a full description). +Here's an example of defining a simple 'vector_square' kernel. + +```cpp +template +__global__ void +vector_square(T *C_d, const T *A_d, size_t N) +{ + size_t offset = (blockIdx.x * blockDim.x + threadIdx.x); + size_t stride = blockDim.x * gridDim.x; + + for (size_t i=offset; i $real_sym" >> $1.log + sed -i "s/$mangled_sym/$real_sym/g" $1 +done + +# HC kernels +kernels=$(grep cxxamp_trampoline $1 | cut -d" " -f1 | sort | uniq) +for mangled_sym in $kernels; do + real_sym=$(echo $mangled_sym | sed "s/^/_/g; s/_EC_/$/g" | c++filt -p | cut -d\( -f1 | cut -d" " -f1 --complement | sed 's/ /\\\ /g') + #echo "$mangled_sym => $real_sym" >> $1.log + sed -i "s/$mangled_sym/$real_sym/g" $1 +done diff --git a/projects/hip/cmake/FindHIP.cmake b/projects/hip/cmake/FindHIP.cmake new file mode 100644 index 0000000000..422b708917 --- /dev/null +++ b/projects/hip/cmake/FindHIP.cmake @@ -0,0 +1,777 @@ +# Copyright (c) 2016 - 2021 Advanced Micro Devices, Inc. All Rights Reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in +# all copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +# THE SOFTWARE. + +############################################################################### +# FindHIP.cmake +############################################################################### +include(CheckCXXCompilerFlag) +############################################################################### +# SET: Variable defaults +############################################################################### +# User defined flags +set(HIP_HIPCC_FLAGS "" CACHE STRING "Semicolon delimited flags for HIPCC") +set(HIP_CLANG_FLAGS "" CACHE STRING "Semicolon delimited flags for CLANG") +set(HIP_NVCC_FLAGS "" CACHE STRING "Semicolon delimted flags for NVCC") +mark_as_advanced(HIP_HIPCC_FLAGS HIP_CLANG_FLAGS HIP_NVCC_FLAGS) + +set(_hip_configuration_types ${CMAKE_CONFIGURATION_TYPES} ${CMAKE_BUILD_TYPE} Debug MinSizeRel Release RelWithDebInfo) +list(REMOVE_DUPLICATES _hip_configuration_types) +foreach(config ${_hip_configuration_types}) + string(TOUPPER ${config} config_upper) + set(HIP_HIPCC_FLAGS_${config_upper} "" CACHE STRING "Semicolon delimited flags for HIPCC") + set(HIP_CLANG_FLAGS_${config_upper} "" CACHE STRING "Semicolon delimited flags for CLANG") + set(HIP_NVCC_FLAGS_${config_upper} "" CACHE STRING "Semicolon delimited flags for NVCC") + mark_as_advanced(HIP_HIPCC_FLAGS_${config_upper} HIP_CLANG_FLAGS_${config_upper} HIP_NVCC_FLAGS_${config_upper}) +endforeach() +option(HIP_HOST_COMPILATION_CPP "Host code compilation mode" ON) +option(HIP_VERBOSE_BUILD "Print out the commands run while compiling the HIP source file. With the Makefile generator this defaults to VERBOSE variable specified on the command line, but can be forced on with this option." OFF) +mark_as_advanced(HIP_HOST_COMPILATION_CPP) + +############################################################################### +# FIND: HIP and associated helper binaries +############################################################################### + +if(NOT HIP_CXX_COMPILER) + set(HIP_CXX_COMPILER ${CMAKE_CXX_COMPILER}) +endif() + +# Approach: To find HIP_CLANG_PATH for HIP_CXX_COMPILER types +# For HIP_CXX_COMPILER as *hipcc use hipconfig -l option to get the clang path. +# For HIP_CXX_COMPILER as *clang use real path of HIP_CXX_COMPILER +# IF not successful in getting the CLANG_PATH using HIP_CXX_COMPILER +# fallback to Old Method to find HIP_CLANG_PATH from ENV Vars, ROCMPATH, HIPPATH etc. +if(HIP_CXX_COMPILER MATCHES ".*hipcc") + get_filename_component(HIPCC_PATH "${HIP_CXX_COMPILER}" DIRECTORY) + set(_HIPCONFIG_EXECUTABLE "${HIPCC_PATH}/hipconfig") + execute_process(COMMAND ${_HIPCONFIG_EXECUTABLE} -l + OUTPUT_VARIABLE _HIP_CLANG_INSTALL_PATH + OUTPUT_STRIP_TRAILING_WHITESPACE + ERROR_VARIABLE _HIPCONFIG_EXE_ERROR + ERROR_STRIP_TRAILING_WHITESPACE + RESULT_VARIABLE _HIPCONFIG_EXE_EXIT_CODE) + if( _HIPCONFIG_EXE_ERROR ) + message( STATUS "hipconfig -l option failed with error: ${_HIPCONFIG_EXE_ERROR}" ) + else() #IF hipconfig -l executed with no error + if( "${_HIPCONFIG_EXE_EXIT_CODE}" STREQUAL "0" ) + set(HIP_CLANG_PATH "${_HIP_CLANG_INSTALL_PATH}") + else() + message( STATUS "${_HIPCONFIG_EXECUTABLE} Failed with Exit code: ${_HIPCONFIG_EXE_EXIT_CODE}" ) + endif() + endif() +elseif (HIP_CXX_COMPILER MATCHES ".*clang\\+\\+") + get_filename_component(_HIP_CLANG_REAL_PATH "${HIP_CXX_COMPILER}" REALPATH) + get_filename_component(_HIP_CLANG_BIN_PATH "${_HIP_CLANG_REAL_PATH}" DIRECTORY) + set(HIP_CLANG_PATH "${_HIP_CLANG_BIN_PATH}") +endif() + +if(WIN32) + # In windows FindHIP.cmake is installed in /cmake + get_filename_component(_IMPORT_PREFIX "${CMAKE_CURRENT_LIST_DIR}/../" REALPATH) +else() + # In Linux FindHIP.cmake is installed in /lib/cmake/hip + # RealPath: /lib/cmake/hip/FindHIP.cmake + # Go 4 level up to get IMPORT PREFIX as + get_filename_component(_FILE_PATH "${CMAKE_CURRENT_LIST_FILE}" REALPATH) + get_filename_component(_IMPORT_PREFIX "${_FILE_PATH}/../../../../" ABSOLUTE) +endif() + +# HIP is currently not supported for apple +if(NOT APPLE) + # Search for HIP installation + if(NOT HIP_ROOT_DIR) + # Search in user specified path first + find_path( + HIP_ROOT_DIR + NAMES bin/hipconfig + PATHS + "$ENV{ROCM_PATH}" + "$ENV{ROCM_PATH}/hip" + ENV HIP_PATH + ${_IMPORT_PREFIX} + DOC "HIP installed location" + NO_DEFAULT_PATH + ) + if(NOT EXISTS ${HIP_ROOT_DIR}) + if(HIP_FIND_REQUIRED) + message(FATAL_ERROR "Specify HIP_ROOT_DIR") + elseif(NOT HIP_FIND_QUIETLY) + message("HIP_ROOT_DIR not found or specified") + endif() + endif() + # And push it back to the cache + set(HIP_ROOT_DIR ${HIP_ROOT_DIR} CACHE PATH "HIP installed location" FORCE) + endif() + + # Find HIPCC executable + find_program( + HIP_HIPCC_EXECUTABLE + NAMES hipcc + PATHS + "${HIP_ROOT_DIR}" + ENV ROCM_PATH + ENV HIP_PATH + PATH_SUFFIXES bin + NO_DEFAULT_PATH + ) + if(NOT HIP_HIPCC_EXECUTABLE) + # Now search in default paths + find_program(HIP_HIPCC_EXECUTABLE hipcc) + endif() + + # Find HIPCONFIG executable + find_program( + HIP_HIPCONFIG_EXECUTABLE + NAMES hipconfig + PATHS + "${HIP_ROOT_DIR}" + ENV ROCM_PATH + ENV HIP_PATH + PATH_SUFFIXES bin + NO_DEFAULT_PATH + ) + if(NOT HIP_HIPCONFIG_EXECUTABLE) + # Now search in default paths + find_program(HIP_HIPCONFIG_EXECUTABLE hipconfig) + endif() + if(NOT UNIX) + get_filename_component(HIPCONFIG_EXECUTABLE_EXT ${HIP_HIPCONFIG_EXECUTABLE} EXT) + if(NOT HIPCONFIG_EXECUTABLE_EXT STREQUAL ".bat") + set(HIP_HIPCONFIG_EXECUTABLE "${HIP_HIPCONFIG_EXECUTABLE}.bat") + set(HIP_HIPCC_EXECUTABLE "${HIP_HIPCC_EXECUTABLE}.bat") + endif() + endif() + mark_as_advanced(HIP_HIPCONFIG_EXECUTABLE) + mark_as_advanced(HIP_HIPCC_EXECUTABLE) + + # Find HIPCC_CMAKE_LINKER_HELPER executable + find_program( + HIP_HIPCC_CMAKE_LINKER_HELPER + NAMES hipcc_cmake_linker_helper + PATHS + "${HIP_ROOT_DIR}" + ENV ROCM_PATH + ENV HIP_PATH + PATH_SUFFIXES bin + NO_DEFAULT_PATH + ) + if(NOT HIP_HIPCC_CMAKE_LINKER_HELPER) + # Now search in default paths + find_program(HIP_HIPCC_CMAKE_LINKER_HELPER hipcc_cmake_linker_helper) + endif() + mark_as_advanced(HIP_HIPCC_CMAKE_LINKER_HELPER) + + if(HIP_HIPCONFIG_EXECUTABLE AND NOT HIP_VERSION) + # Compute the version + execute_process( + COMMAND ${HIP_HIPCONFIG_EXECUTABLE} --version + OUTPUT_VARIABLE _hip_version + ERROR_VARIABLE _hip_error + OUTPUT_STRIP_TRAILING_WHITESPACE + ERROR_STRIP_TRAILING_WHITESPACE + ) + if(NOT _hip_error) + set(HIP_VERSION ${_hip_version} CACHE STRING "Version of HIP as computed from hipcc") + else() + set(HIP_VERSION "0.0.0" CACHE STRING "Version of HIP as computed by FindHIP()") + endif() + mark_as_advanced(HIP_VERSION) + endif() + if(HIP_VERSION) + string(REPLACE "." ";" _hip_version_list "${HIP_VERSION}") + list(GET _hip_version_list 0 HIP_VERSION_MAJOR) + list(GET _hip_version_list 1 HIP_VERSION_MINOR) + list(GET _hip_version_list 2 HIP_VERSION_PATCH) + set(HIP_VERSION_STRING "${HIP_VERSION}") + endif() + + if(HIP_HIPCONFIG_EXECUTABLE AND NOT HIP_PLATFORM) + # Compute the platform + execute_process( + COMMAND ${HIP_HIPCONFIG_EXECUTABLE} --platform + OUTPUT_VARIABLE _hip_platform + OUTPUT_STRIP_TRAILING_WHITESPACE + ) + set(HIP_PLATFORM ${_hip_platform} CACHE STRING "HIP platform as computed by hipconfig") + mark_as_advanced(HIP_PLATFORM) + endif() + + if(HIP_HIPCONFIG_EXECUTABLE AND NOT HIP_COMPILER) + # Compute the compiler + execute_process( + COMMAND ${HIP_HIPCONFIG_EXECUTABLE} --compiler + OUTPUT_VARIABLE _hip_compiler + OUTPUT_STRIP_TRAILING_WHITESPACE + ) + set(HIP_COMPILER ${_hip_compiler} CACHE STRING "HIP compiler as computed by hipconfig") + mark_as_advanced(HIP_COMPILER) + endif() + + if(HIP_HIPCONFIG_EXECUTABLE AND NOT HIP_RUNTIME) + # Compute the runtime + execute_process( + COMMAND ${HIP_HIPCONFIG_EXECUTABLE} --runtime + OUTPUT_VARIABLE _hip_runtime + OUTPUT_STRIP_TRAILING_WHITESPACE + ) + set(HIP_RUNTIME ${_hip_runtime} CACHE STRING "HIP runtime as computed by hipconfig") + mark_as_advanced(HIP_RUNTIME) + endif() +endif() + +include(FindPackageHandleStandardArgs) +find_package_handle_standard_args( + HIP + REQUIRED_VARS + HIP_ROOT_DIR + HIP_HIPCC_EXECUTABLE + HIP_HIPCONFIG_EXECUTABLE + HIP_PLATFORM + HIP_COMPILER + HIP_RUNTIME + VERSION_VAR HIP_VERSION + ) + +############################################################################### +# Set HIP CMAKE Flags +############################################################################### +# Copy the invocation styles from CXX to HIP +set(CMAKE_HIP_ARCHIVE_CREATE ${CMAKE_CXX_ARCHIVE_CREATE}) +set(CMAKE_HIP_ARCHIVE_APPEND ${CMAKE_CXX_ARCHIVE_APPEND}) +set(CMAKE_HIP_ARCHIVE_FINISH ${CMAKE_CXX_ARCHIVE_FINISH}) +set(CMAKE_SHARED_LIBRARY_SONAME_HIP_FLAG ${CMAKE_SHARED_LIBRARY_SONAME_CXX_FLAG}) +set(CMAKE_SHARED_LIBRARY_CREATE_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_CREATE_CXX_FLAGS}) +set(CMAKE_SHARED_LIBRARY_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_CXX_FLAGS}) +#set(CMAKE_SHARED_LIBRARY_LINK_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_LINK_CXX_FLAGS}) +set(CMAKE_SHARED_LIBRARY_RUNTIME_HIP_FLAG ${CMAKE_SHARED_LIBRARY_RUNTIME_CXX_FLAG}) +set(CMAKE_SHARED_LIBRARY_RUNTIME_HIP_FLAG_SEP ${CMAKE_SHARED_LIBRARY_RUNTIME_CXX_FLAG_SEP}) +set(CMAKE_SHARED_LIBRARY_LINK_STATIC_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_LINK_STATIC_CXX_FLAGS}) +set(CMAKE_SHARED_LIBRARY_LINK_DYNAMIC_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_LINK_DYNAMIC_CXX_FLAGS}) + +set(HIP_CLANG_PARALLEL_BUILD_COMPILE_OPTIONS "") +set(HIP_CLANG_PARALLEL_BUILD_LINK_OPTIONS "") + +if("${HIP_COMPILER}" STREQUAL "nvcc") + # Set the CMake Flags to use the nvcc Compiler. + set(CMAKE_HIP_CREATE_SHARED_LIBRARY "${HIP_HIPCC_CMAKE_LINKER_HELPER} -o ") + set(CMAKE_HIP_CREATE_SHARED_MODULE "${HIP_HIPCC_CMAKE_LINKER_HELPER} -o -shared" ) + set(CMAKE_HIP_LINK_EXECUTABLE "${HIP_HIPCC_CMAKE_LINKER_HELPER} -o ") +elseif("${HIP_COMPILER}" STREQUAL "clang") + #Set HIP_CLANG_PATH + if("x${HIP_CLANG_PATH}" STREQUAL "x") + # IF HIP_CLANG_INSTALL_DIR is Found + if( HIP_CLANG_INSTALL_DIR ) + set(HIP_CLANG_PATH ${HIP_CLANG_INSTALL_DIR}) + else() # IF HIP_CLANG_INSTALL_DIR is not found + if(DEFINED ENV{HIP_CLANG_PATH}) + set(HIP_CLANG_PATH $ENV{HIP_CLANG_PATH}) + elseif(DEFINED ENV{ROCM_PATH}) + set(HIP_CLANG_PATH "$ENV{ROCM_PATH}/llvm/bin") + elseif(DEFINED ENV{HIP_PATH}) + if(EXISTS "$ENV{HIP_PATH}/llvm/bin") #File Reorg backward compatibility + set(HIP_CLANG_PATH "$ENV{HIP_PATH}/llvm/bin") + else() + set(HIP_CLANG_PATH "$ENV{HIP_PATH}/../llvm/bin") + endif() + elseif(DEFINED HIP_PATH) + if(EXISTS "${HIP_PATH}/llvm/bin") #File Reorg backward compatibility + set(HIP_CLANG_PATH "${HIP_PATH}/llvm/bin") + else() + set(HIP_CLANG_PATH "${HIP_PATH}/../llvm/bin") + endif() + # Handle the case where ROCM_PATH is defined and not set in ENV + elseif(DEFINED ROCM_PATH) + set(HIP_CLANG_PATH "${ROCM_PATH}/llvm/bin") + else() + message(FATAL_ERROR "Unable to find the clang compiler path. Set ROCM_PATH or HIP_PATH in env ") + endif() + endif() # HIP_CLANG_INSTALL_DIR Check + endif() # Set HIP_CLANG_PATH + + #Number of parallel jobs by default is 1 + if(NOT DEFINED HIP_CLANG_NUM_PARALLEL_JOBS) + set(HIP_CLANG_NUM_PARALLEL_JOBS 1) + endif() + #Add support for parallel build and link + if(${CMAKE_CXX_COMPILER_ID} STREQUAL "Clang") + check_cxx_compiler_flag("-parallel-jobs=1" HIP_CLANG_SUPPORTS_PARALLEL_JOBS) + endif() + if(HIP_CLANG_NUM_PARALLEL_JOBS GREATER 1) + if(${HIP_CLANG_SUPPORTS_PARALLEL_JOBS}) + set(HIP_CLANG_PARALLEL_BUILD_COMPILE_OPTIONS "-Wno-format-nonliteral -parallel-jobs=${HIP_CLANG_NUM_PARALLEL_JOBS}") + set(HIP_CLANG_PARALLEL_BUILD_LINK_OPTIONS "-parallel-jobs=${HIP_CLANG_NUM_PARALLEL_JOBS}") + else() + message("clang compiler doesn't support parallel jobs") + endif() + endif() + + # Set the CMake Flags to use the HIP-Clang Compiler. + set(CMAKE_HIP_CREATE_SHARED_LIBRARY "${HIP_HIPCC_CMAKE_LINKER_HELPER} ${HIP_CLANG_PATH} ${HIP_CLANG_PARALLEL_BUILD_LINK_OPTIONS} -o ") + set(CMAKE_HIP_CREATE_SHARED_MODULE "${HIP_HIPCC_CMAKE_LINKER_HELPER} ${HIP_CLANG_PATH} ${HIP_CLANG_PARALLEL_BUILD_LINK_OPTIONS} -o -shared" ) + set(CMAKE_HIP_LINK_EXECUTABLE "${HIP_HIPCC_CMAKE_LINKER_HELPER} ${HIP_CLANG_PATH} ${HIP_CLANG_PARALLEL_BUILD_LINK_OPTIONS} -o ") + + if("${HIP_RUNTIME}" STREQUAL "rocclr") + if(TARGET host) + message(STATUS "host interface - found") + set(HIP_HOST_INTERFACE host) + endif() + endif() +endif() + +############################################################################### +# MACRO: Locate helper files +############################################################################### +macro(HIP_FIND_HELPER_FILE _name _extension) + set(_hip_full_name "${_name}.${_extension}") + get_filename_component(CMAKE_CURRENT_LIST_DIR "${CMAKE_CURRENT_LIST_FILE}" PATH) + set(HIP_${_name} "${CMAKE_CURRENT_LIST_DIR}/FindHIP/${_hip_full_name}") + if(NOT EXISTS "${HIP_${_name}}") + set(error_message "${_hip_full_name} not found in ${CMAKE_CURRENT_LIST_DIR}/FindHIP") + if(HIP_FIND_REQUIRED) + message(FATAL_ERROR "${error_message}") + else() + if(NOT HIP_FIND_QUIETLY) + message(STATUS "${error_message}") + endif() + endif() + endif() + # Set this variable as internal, so the user isn't bugged with it. + set(HIP_${_name} ${HIP_${_name}} CACHE INTERNAL "Location of ${_full_name}" FORCE) +endmacro() + +############################################################################### +hip_find_helper_file(run_make2cmake cmake) +hip_find_helper_file(run_hipcc cmake) +############################################################################### + +############################################################################### +# MACRO: Reset compiler flags +############################################################################### +macro(HIP_RESET_FLAGS) + unset(HIP_HIPCC_FLAGS) + unset(HIP_CLANG_FLAGS) + unset(HIP_NVCC_FLAGS) + foreach(config ${_hip_configuration_types}) + string(TOUPPER ${config} config_upper) + unset(HIP_HIPCC_FLAGS_${config_upper}) + unset(HIP_CLANG_FLAGS_${config_upper}) + unset(HIP_NVCC_FLAGS_${config_upper}) + endforeach() +endmacro() + +############################################################################### +# MACRO: Separate the options from the sources +############################################################################### +macro(HIP_GET_SOURCES_AND_OPTIONS _sources _cmake_options _hipcc_options _clang_options _nvcc_options) + set(${_sources}) + set(${_cmake_options}) + set(${_hipcc_options}) + set(${_clang_options}) + set(${_nvcc_options}) + set(_hipcc_found_options FALSE) + set(_clang_found_options FALSE) + set(_nvcc_found_options FALSE) + foreach(arg ${ARGN}) + if("x${arg}" STREQUAL "xHIPCC_OPTIONS") + set(_hipcc_found_options TRUE) + set(_clang_found_options FALSE) + set(_nvcc_found_options FALSE) + elseif("x${arg}" STREQUAL "xCLANG_OPTIONS") + set(_hipcc_found_options FALSE) + set(_clang_found_options TRUE) + set(_nvcc_found_options FALSE) + elseif("x${arg}" STREQUAL "xNVCC_OPTIONS") + set(_hipcc_found_options FALSE) + set(_clang_found_options FALSE) + set(_nvcc_found_options TRUE) + elseif( + "x${arg}" STREQUAL "xEXCLUDE_FROM_ALL" OR + "x${arg}" STREQUAL "xSTATIC" OR + "x${arg}" STREQUAL "xSHARED" OR + "x${arg}" STREQUAL "xMODULE" + ) + list(APPEND ${_cmake_options} ${arg}) + else() + if(_hipcc_found_options) + list(APPEND ${_hipcc_options} ${arg}) + elseif(_clang_found_options) + list(APPEND ${_clang_options} ${arg}) + elseif(_nvcc_found_options) + list(APPEND ${_nvcc_options} ${arg}) + else() + # Assume this is a file + list(APPEND ${_sources} ${arg}) + endif() + endif() + endforeach() +endmacro() + +############################################################################### +# MACRO: Add include directories to pass to the hipcc command +############################################################################### +set(HIP_HIPCC_INCLUDE_ARGS_USER "") +macro(HIP_INCLUDE_DIRECTORIES) + foreach(dir ${ARGN}) + list(APPEND HIP_HIPCC_INCLUDE_ARGS_USER $<$:-I${dir}>) + endforeach() +endmacro() + +############################################################################### +# FUNCTION: Helper to avoid clashes of files with the same basename but different paths +############################################################################### +function(HIP_COMPUTE_BUILD_PATH path build_path) + # Convert to cmake style paths + file(TO_CMAKE_PATH "${path}" bpath) + if(IS_ABSOLUTE "${bpath}") + string(FIND "${bpath}" "${CMAKE_CURRENT_BINARY_DIR}" _binary_dir_pos) + if(_binary_dir_pos EQUAL 0) + file(RELATIVE_PATH bpath "${CMAKE_CURRENT_BINARY_DIR}" "${bpath}") + else() + file(RELATIVE_PATH bpath "${CMAKE_CURRENT_SOURCE_DIR}" "${bpath}") + endif() + endif() + + # Remove leading / + string(REGEX REPLACE "^[/]+" "" bpath "${bpath}") + # Avoid absolute paths by removing ':' + string(REPLACE ":" "_" bpath "${bpath}") + # Avoid relative paths that go up the tree + string(REPLACE "../" "__/" bpath "${bpath}") + # Avoid spaces + string(REPLACE " " "_" bpath "${bpath}") + # Strip off the filename + get_filename_component(bpath "${bpath}" PATH) + + set(${build_path} "${bpath}" PARENT_SCOPE) +endfunction() + +############################################################################### +# MACRO: Parse OPTIONS from ARGN & set variables prefixed by _option_prefix +############################################################################### +macro(HIP_PARSE_HIPCC_OPTIONS _option_prefix) + set(_hip_found_config) + foreach(arg ${ARGN}) + # Determine if we are dealing with a per-configuration flag + foreach(config ${_hip_configuration_types}) + string(TOUPPER ${config} config_upper) + if(arg STREQUAL "${config_upper}") + set(_hip_found_config _${arg}) + # Clear arg to prevent it from being processed anymore + set(arg) + endif() + endforeach() + if(arg) + list(APPEND ${_option_prefix}${_hip_found_config} "${arg}") + endif() + endforeach() +endmacro() + +############################################################################### +# MACRO: Try and include dependency file if it exists +############################################################################### +macro(HIP_INCLUDE_HIPCC_DEPENDENCIES dependency_file) + set(HIP_HIPCC_DEPEND) + set(HIP_HIPCC_DEPEND_REGENERATE FALSE) + + # Create the dependency file if it doesn't exist + if(NOT EXISTS ${dependency_file}) + file(WRITE ${dependency_file} "# Generated by: FindHIP.cmake. Do not edit.\n") + endif() + # Include the dependency file + include(${dependency_file}) + + # Verify the existence of all the included files + if(HIP_HIPCC_DEPEND) + foreach(f ${HIP_HIPCC_DEPEND}) + if(NOT EXISTS ${f}) + # If they aren't there, regenerate the file again + set(HIP_HIPCC_DEPEND_REGENERATE TRUE) + endif() + endforeach() + else() + # No dependencies, so regenerate the file + set(HIP_HIPCC_DEPEND_REGENERATE TRUE) + endif() + + # Regenerate the dependency file if needed + if(HIP_HIPCC_DEPEND_REGENERATE) + set(HIP_HIPCC_DEPEND ${dependency_file}) + file(WRITE ${dependency_file} "# Generated by: FindHIP.cmake. Do not edit.\n") + endif() +endmacro() + +############################################################################### +# MACRO: Prepare cmake commands for the target +############################################################################### +macro(HIP_PREPARE_TARGET_COMMANDS _target _format _generated_files _source_files) + set(_hip_flags "") + string(TOUPPER "${CMAKE_BUILD_TYPE}" _hip_build_configuration) + if(HIP_HOST_COMPILATION_CPP) + set(HIP_C_OR_CXX CXX) + else() + set(HIP_C_OR_CXX C) + endif() + set(generated_extension ${CMAKE_${HIP_C_OR_CXX}_OUTPUT_EXTENSION}) + + # Initialize list of includes with those specified by the user. Append with + # ones specified to cmake directly. + set(HIP_HIPCC_INCLUDE_ARGS ${HIP_HIPCC_INCLUDE_ARGS_USER}) + + # Add the include directories + set(include_directories_generator "$") + list(APPEND HIP_HIPCC_INCLUDE_ARGS "$<$:-I$>") + + get_directory_property(_hip_include_directories INCLUDE_DIRECTORIES) + list(REMOVE_DUPLICATES _hip_include_directories) + if(_hip_include_directories) + foreach(dir ${_hip_include_directories}) + list(APPEND HIP_HIPCC_INCLUDE_ARGS $<$:-I${dir}>) + endforeach() + endif() + + HIP_GET_SOURCES_AND_OPTIONS(_hip_sources _hip_cmake_options _hipcc_options _clang_options _nvcc_options ${ARGN}) + HIP_PARSE_HIPCC_OPTIONS(HIP_HIPCC_FLAGS ${_hipcc_options}) + HIP_PARSE_HIPCC_OPTIONS(HIP_CLANG_FLAGS ${_clang_options}) + HIP_PARSE_HIPCC_OPTIONS(HIP_NVCC_FLAGS ${_nvcc_options}) + + # Add the compile definitions + set(compile_definition_generator "$") + list(APPEND HIP_HIPCC_FLAGS "$<$:-D$>") + + # Check if we are building shared library. + set(_hip_build_shared_libs FALSE) + list(FIND _hip_cmake_options SHARED _hip_found_SHARED) + list(FIND _hip_cmake_options MODULE _hip_found_MODULE) + if(_hip_found_SHARED GREATER -1 OR _hip_found_MODULE GREATER -1) + set(_hip_build_shared_libs TRUE) + endif() + list(FIND _hip_cmake_options STATIC _hip_found_STATIC) + if(_hip_found_STATIC GREATER -1) + set(_hip_build_shared_libs FALSE) + endif() + + # If we are building a shared library, add extra flags to HIP_HIPCC_FLAGS + if(_hip_build_shared_libs) + list(APPEND HIP_CLANG_FLAGS "-fPIC") + list(APPEND HIP_NVCC_FLAGS "--shared -Xcompiler '-fPIC'") + endif() + + # Set host compiler + set(HIP_HOST_COMPILER "${CMAKE_${HIP_C_OR_CXX}_COMPILER}") + + # Set compiler flags + set(_HIP_HOST_FLAGS "set(CMAKE_HOST_FLAGS ${CMAKE_${HIP_C_OR_CXX}_FLAGS})") + set(_HIP_HIPCC_FLAGS "set(HIP_HIPCC_FLAGS ${HIP_HIPCC_FLAGS})") + set(_HIP_CLANG_FLAGS "set(HIP_CLANG_FLAGS ${HIP_CLANG_FLAGS})") + set(_HIP_NVCC_FLAGS "set(HIP_NVCC_FLAGS ${HIP_NVCC_FLAGS})") + foreach(config ${_hip_configuration_types}) + string(TOUPPER ${config} config_upper) + set(_HIP_HOST_FLAGS "${_HIP_HOST_FLAGS}\nset(CMAKE_HOST_FLAGS_${config_upper} ${CMAKE_${HIP_C_OR_CXX}_FLAGS_${config_upper}})") + set(_HIP_HIPCC_FLAGS "${_HIP_HIPCC_FLAGS}\nset(HIP_HIPCC_FLAGS_${config_upper} ${HIP_HIPCC_FLAGS_${config_upper}})") + set(_HIP_CLANG_FLAGS "${_HIP_CLANG_FLAGS}\nset(HIP_CLANG_FLAGS_${config_upper} ${HIP_CLANG_FLAGS_${config_upper}})") + set(_HIP_NVCC_FLAGS "${_HIP_NVCC_FLAGS}\nset(HIP_NVCC_FLAGS_${config_upper} ${HIP_NVCC_FLAGS_${config_upper}})") + endforeach() + + # Reset the output variable + set(_hip_generated_files "") + set(_hip_source_files "") + + # Iterate over all arguments and create custom commands for all source files + foreach(file ${ARGN}) + # Ignore any file marked as a HEADER_FILE_ONLY + get_source_file_property(_is_header ${file} HEADER_FILE_ONLY) + # Allow per source file overrides of the format. Also allows compiling non .cu files. + get_source_file_property(_hip_source_format ${file} HIP_SOURCE_PROPERTY_FORMAT) + if((${file} MATCHES "\\.cu$" OR _hip_source_format) AND NOT _is_header) + set(host_flag FALSE) + else() + set(host_flag TRUE) + endif() + + if(NOT host_flag) + # Determine output directory + HIP_COMPUTE_BUILD_PATH("${file}" hip_build_path) + set(hip_compile_output_dir "${CMAKE_CURRENT_BINARY_DIR}/CMakeFiles/${_target}.dir/${hip_build_path}") + + get_filename_component(basename ${file} NAME) + set(generated_file_path "${hip_compile_output_dir}/${CMAKE_CFG_INTDIR}") + set(generated_file_basename "${_target}_generated_${basename}${generated_extension}") + + # Set file names + set(generated_file "${generated_file_path}/${generated_file_basename}") + set(cmake_dependency_file "${hip_compile_output_dir}/${generated_file_basename}.depend") + set(custom_target_script_pregen "${hip_compile_output_dir}/${generated_file_basename}.cmake.pre-gen") + set(custom_target_script "${hip_compile_output_dir}/${generated_file_basename}.cmake") + + # Set properties for object files + set_source_files_properties("${generated_file}" + PROPERTIES + EXTERNAL_OBJECT true # This is an object file not to be compiled, but only be linked + ) + + # Don't add CMAKE_CURRENT_SOURCE_DIR if the path is already an absolute path + get_filename_component(file_path "${file}" PATH) + if(IS_ABSOLUTE "${file_path}") + set(source_file "${file}") + else() + set(source_file "${CMAKE_CURRENT_SOURCE_DIR}/${file}") + endif() + + # Bring in the dependencies + HIP_INCLUDE_HIPCC_DEPENDENCIES(${cmake_dependency_file}) + + # Configure the build script + configure_file("${HIP_run_hipcc}" "${custom_target_script_pregen}" @ONLY) + file(GENERATE + OUTPUT "${custom_target_script}" + INPUT "${custom_target_script_pregen}" + ) + set(main_dep DEPENDS ${source_file}) + if(CMAKE_GENERATOR MATCHES "Makefiles") + set(verbose_output "$(VERBOSE)") + elseif(HIP_VERBOSE_BUILD) + set(verbose_output ON) + else() + set(verbose_output OFF) + endif() + + # Create up the comment string + file(RELATIVE_PATH generated_file_relative_path "${CMAKE_BINARY_DIR}" "${generated_file}") + set(hip_build_comment_string "Building HIPCC object ${generated_file_relative_path}") + + # Build the generated file and dependency file + add_custom_command( + OUTPUT ${generated_file} + # These output files depend on the source_file and the contents of cmake_dependency_file + ${main_dep} + DEPENDS ${HIP_HIPCC_DEPEND} + DEPENDS ${custom_target_script} + # Make sure the output directory exists before trying to write to it. + COMMAND ${CMAKE_COMMAND} -E make_directory "${generated_file_path}" + COMMAND ${CMAKE_COMMAND} ARGS + -D verbose:BOOL=${verbose_output} + -D build_configuration:STRING=${_hip_build_configuration} + -D "generated_file:STRING=${generated_file}" + -P "${custom_target_script}" + WORKING_DIRECTORY "${hip_compile_output_dir}" + COMMENT "${hip_build_comment_string}" + ) + + # Make sure the build system knows the file is generated + set_source_files_properties(${generated_file} PROPERTIES GENERATED TRUE) + list(APPEND _hip_generated_files ${generated_file}) + list(APPEND _hip_source_files ${file}) + endif() + endforeach() + + # Set the return parameter + set(${_generated_files} ${_hip_generated_files}) + set(${_source_files} ${_hip_source_files}) +endmacro() + +############################################################################### +# HIP_ADD_EXECUTABLE +############################################################################### +macro(HIP_ADD_EXECUTABLE hip_target) + # Separate the sources from the options + HIP_GET_SOURCES_AND_OPTIONS(_sources _cmake_options _hipcc_options _clang_options _nvcc_options ${ARGN}) + HIP_PREPARE_TARGET_COMMANDS(${hip_target} OBJ _generated_files _source_files ${_sources} HIPCC_OPTIONS ${_hipcc_options} CLANG_OPTIONS ${_clang_options} NVCC_OPTIONS ${_nvcc_options}) + if(_source_files) + list(REMOVE_ITEM _sources ${_source_files}) + endif() + if("${HIP_COMPILER}" STREQUAL "clang") + if("x${HIP_CLANG_PATH}" STREQUAL "x") + # IF HIP_CLANG_INSTALL_DIR is Found + if( HIP_CLANG_INSTALL_DIR ) + set(HIP_CLANG_PATH ${HIP_CLANG_INSTALL_DIR}) + else() # IF HIP_CLANG_INSTALL_DIR is not found + if(DEFINED ENV{HIP_CLANG_PATH}) + set(HIP_CLANG_PATH $ENV{HIP_CLANG_PATH}) + elseif(DEFINED ENV{ROCM_PATH}) + set(HIP_CLANG_PATH "$ENV{ROCM_PATH}/llvm/bin") + elseif(DEFINED ENV{HIP_PATH}) + if(EXISTS "$ENV{HIP_PATH}/llvm/bin") #file reorg backward compatibility + set(HIP_CLANG_PATH "$ENV{HIP_PATH}/llvm/bin") + else() + set(HIP_CLANG_PATH "$ENV{HIP_PATH}/../llvm/bin") + endif() + elseif(DEFINED HIP_PATH) + if(EXISTS "${HIP_PATH}/llvm/bin") #file reorg backward compatibility + set(HIP_CLANG_PATH "${HIP_PATH}/llvm/bin") + else() + set(HIP_CLANG_PATH "${HIP_PATH}/../llvm/bin") + endif() + # Handle the case where ROCM_PATH is defined and not set in ENV + elseif(DEFINED ROCM_PATH) + set(HIP_CLANG_PATH "${ROCM_PATH}/llvm/bin") + else() + message(FATAL_ERROR "Unable to find the clang compiler path. Set ROCM_PATH or HIP_PATH in env") + endif() + endif() # HIP_CLANG_INSTALL_DIR Check + endif() + + set(CMAKE_HIP_LINK_EXECUTABLE "${HIP_HIPCC_CMAKE_LINKER_HELPER} ${HIP_CLANG_PATH} ${HIP_CLANG_PARALLEL_BUILD_LINK_OPTIONS} -o ") + else() + set(CMAKE_HIP_LINK_EXECUTABLE "${HIP_HIPCC_CMAKE_LINKER_HELPER} -o ") + endif() + if ("${_sources}" STREQUAL "") + add_executable(${hip_target} ${_cmake_options} ${_generated_files} "") + else() + add_executable(${hip_target} ${_cmake_options} ${_generated_files} ${_sources}) + endif() + #LINK_OPTIONS + if("${HIP_COMPILER}" STREQUAL "nvcc") + # Some arch flags need be sent to linker. _nvcc_options mixes compiling and linker flags. + string(REPLACE ";" " " _nvcc_flags "${_nvcc_options}") # Replace ',' with space + if(NOT "x${_nvcc_flags}" STREQUAL "x") + set_target_properties(${hip_target} PROPERTIES LINK_FLAGS "${_nvcc_flags}") + endif() + endif() + set_target_properties(${hip_target} PROPERTIES LINKER_LANGUAGE HIP) + # Link with host + if (HIP_HOST_INTERFACE) + # hip rt should be rocclr, compiler should be clang + target_link_libraries(${hip_target} ${HIP_HOST_INTERFACE}) + endif() +endmacro() + +############################################################################### +# HIP_ADD_LIBRARY +############################################################################### +macro(HIP_ADD_LIBRARY hip_target) + # Separate the sources from the options + HIP_GET_SOURCES_AND_OPTIONS(_sources _cmake_options _hipcc_options _clang_options _nvcc_options ${ARGN}) + HIP_PREPARE_TARGET_COMMANDS(${hip_target} OBJ _generated_files _source_files ${_sources} ${_cmake_options} HIPCC_OPTIONS ${_hipcc_options} CLANG_OPTIONS ${_clang_options} NVCC_OPTIONS ${_nvcc_options}) + if(_source_files) + list(REMOVE_ITEM _sources ${_source_files}) + endif() + if ("${_sources}" STREQUAL "") + add_library(${hip_target} ${_cmake_options} ${_generated_files} "") + else() + add_library(${hip_target} ${_cmake_options} ${_generated_files} ${_sources}) + endif() + set_target_properties(${hip_target} PROPERTIES LINKER_LANGUAGE ${HIP_C_OR_CXX}) + # Link with host + if (HIP_HOST_INTERFACE) + # hip rt should be rocclr, compiler should be clang + target_link_libraries(${hip_target} ${HIP_HOST_INTERFACE}) + endif() +endmacro() + +# vim: ts=4:sw=4:expandtab:smartindent diff --git a/projects/hip/cmake/FindHIP/run_hipcc.cmake b/projects/hip/cmake/FindHIP/run_hipcc.cmake new file mode 100644 index 0000000000..aeb943c15f --- /dev/null +++ b/projects/hip/cmake/FindHIP/run_hipcc.cmake @@ -0,0 +1,194 @@ +# Copyright (c) 2016 - 2021 Advanced Micro Devices, Inc. All Rights Reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in +# all copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +# THE SOFTWARE. + +############################################################################### +# Runs commands using HIPCC +############################################################################### + +############################################################################### +# This file runs the hipcc commands to produce the desired output file +# along with the dependency file needed by CMake to compute dependencies. +# +# Input variables: +# +# verbose:BOOL=<> OFF: Be as quiet as possible (default) +# ON : Describe each step +# build_configuration:STRING=<> Build configuration. Defaults to Debug. +# generated_file:STRING=<> File to generate. Mandatory argument. + +if(NOT build_configuration) + set(build_configuration Debug) +endif() +if(NOT generated_file) + message(FATAL_ERROR "You must specify generated_file on the command line") +endif() + +# Set these up as variables to make reading the generated file easier +set(HIP_HIPCC_EXECUTABLE "@HIP_HIPCC_EXECUTABLE@") # path +set(HIP_HIPCONFIG_EXECUTABLE "@HIP_HIPCONFIG_EXECUTABLE@") #path +set(HIP_HOST_COMPILER "@HIP_HOST_COMPILER@") # path +set(CMAKE_COMMAND "@CMAKE_COMMAND@") # path +set(HIP_run_make2cmake "@HIP_run_make2cmake@") # path +set(HIP_CLANG_PATH "@HIP_CLANG_PATH@") #path +set(HIP_CLANG_PARALLEL_BUILD_COMPILE_OPTIONS "@HIP_CLANG_PARALLEL_BUILD_COMPILE_OPTIONS@") + +@HIP_HOST_FLAGS@ +@_HIP_HIPCC_FLAGS@ +@_HIP_CLANG_FLAGS@ +@_HIP_NVCC_FLAGS@ +#Needed to bring the HIP_HIPCC_INCLUDE_ARGS variable in scope +set(HIP_HIPCC_INCLUDE_ARGS @HIP_HIPCC_INCLUDE_ARGS@) # list + +set(cmake_dependency_file "@cmake_dependency_file@") # path +set(source_file "@source_file@") # path +set(host_flag "@host_flag@") # bool + +# Determine compiler and compiler flags +execute_process(COMMAND ${HIP_HIPCONFIG_EXECUTABLE} --platform OUTPUT_VARIABLE HIP_PLATFORM OUTPUT_STRIP_TRAILING_WHITESPACE) +execute_process(COMMAND ${HIP_HIPCONFIG_EXECUTABLE} --compiler OUTPUT_VARIABLE HIP_COMPILER OUTPUT_STRIP_TRAILING_WHITESPACE) +execute_process(COMMAND ${HIP_HIPCONFIG_EXECUTABLE} --runtime OUTPUT_VARIABLE HIP_RUNTIME OUTPUT_STRIP_TRAILING_WHITESPACE) +if(NOT host_flag) + set(__CC ${HIP_HIPCC_EXECUTABLE}) + if("${HIP_PLATFORM}" STREQUAL "amd") + if("${HIP_COMPILER}" STREQUAL "clang") + if(NOT "x${HIP_CLANG_PATH}" STREQUAL "x") + set(ENV{HIP_CLANG_PATH} ${HIP_CLANG_PATH}) + endif() + set(__CC_FLAGS ${HIP_CLANG_PARALLEL_BUILD_COMPILE_OPTIONS} ${HIP_HIPCC_FLAGS} ${HIP_CLANG_FLAGS} ${HIP_HIPCC_FLAGS_${build_configuration}} ${HIP_CLANG_FLAGS_${build_configuration}}) + endif() + else() + set(__CC_FLAGS ${HIP_HIPCC_FLAGS} ${HIP_NVCC_FLAGS} ${HIP_HIPCC_FLAGS_${build_configuration}} ${HIP_NVCC_FLAGS_${build_configuration}}) + endif() +else() + set(__CC ${HIP_HOST_COMPILER}) + set(__CC_FLAGS ${CMAKE_HOST_FLAGS} ${CMAKE_HOST_FLAGS_${build_configuration}}) +endif() +set(__CC_INCLUDES ${HIP_HIPCC_INCLUDE_ARGS}) + +# hip_execute_process - Executes a command with optional command echo and status message. +# status - Status message to print if verbose is true +# command - COMMAND argument from the usual execute_process argument structure +# ARGN - Remaining arguments are the command with arguments +# HIP_result - Return value from running the command +macro(hip_execute_process status command) + set(_command ${command}) + if(NOT "x${_command}" STREQUAL "xCOMMAND") + message(FATAL_ERROR "Malformed call to hip_execute_process. Missing COMMAND as second argument. (command = ${command})") + endif() + if(verbose) + execute_process(COMMAND "${CMAKE_COMMAND}" -E echo -- ${status}) + # Build command string to print + set(hip_execute_process_string) + foreach(arg ${ARGN}) + # Escape quotes if any + string(REPLACE "\"" "\\\"" arg ${arg}) + # Surround args with spaces with quotes + if(arg MATCHES " ") + list(APPEND hip_execute_process_string "\"${arg}\"") + else() + list(APPEND hip_execute_process_string ${arg}) + endif() + endforeach() + # Echo the command + execute_process(COMMAND ${CMAKE_COMMAND} -E echo ${hip_execute_process_string}) + endif() + # Run the command + execute_process(COMMAND ${ARGN} RESULT_VARIABLE HIP_result) +endmacro() + +# Delete the target file +hip_execute_process( + "Removing ${generated_file}" + COMMAND "${CMAKE_COMMAND}" -E remove "${generated_file}" + ) + +# Generate the dependency file +hip_execute_process( + "Generating dependency file: ${cmake_dependency_file}.pre" + COMMAND "${__CC}" + -M + "${source_file}" + -o "${cmake_dependency_file}.pre" + ${__CC_FLAGS} + ${__CC_INCLUDES} + ) + +if(HIP_result) + message(FATAL_ERROR "Error generating ${generated_file}") +endif() + +# Generate the cmake readable dependency file to a temp file +hip_execute_process( + "Generating temporary cmake readable file: ${cmake_dependency_file}.tmp" + COMMAND "${CMAKE_COMMAND}" + -D "input_file:FILEPATH=${cmake_dependency_file}.pre" + -D "output_file:FILEPATH=${cmake_dependency_file}.tmp" + -D "verbose=${verbose}" + -P "${HIP_run_make2cmake}" + ) + +if(HIP_result) + message(FATAL_ERROR "Error generating ${generated_file}") +endif() + +# Copy the file if it is different +hip_execute_process( + "Copy if different ${cmake_dependency_file}.tmp to ${cmake_dependency_file}" + COMMAND "${CMAKE_COMMAND}" -E copy_if_different "${cmake_dependency_file}.tmp" "${cmake_dependency_file}" + ) + +if(HIP_result) + message(FATAL_ERROR "Error generating ${generated_file}") +endif() + +# Delete the temporary file +hip_execute_process( + "Removing ${cmake_dependency_file}.tmp and ${cmake_dependency_file}.pre" + COMMAND "${CMAKE_COMMAND}" -E remove "${cmake_dependency_file}.tmp" "${cmake_dependency_file}.pre" + ) + +if(HIP_result) + message(FATAL_ERROR "Error generating ${generated_file}") +endif() + +# Generate the output file +hip_execute_process( + "Generating ${generated_file}" + COMMAND "${__CC}" + -c + "${source_file}" + -o "${generated_file}" + ${__CC_FLAGS} + ${__CC_INCLUDES} + ) + +if(HIP_result) + # Make sure that we delete the output file + hip_execute_process( + "Removing ${generated_file}" + COMMAND "${CMAKE_COMMAND}" -E remove "${generated_file}" + ) + message(FATAL_ERROR "Error generating file ${generated_file}") +else() + if(verbose) + message("Generated ${generated_file} successfully.") + endif() +endif() +# vim: ts=4:sw=4:expandtab:smartindent diff --git a/projects/hip/cmake/FindHIP/run_make2cmake.cmake b/projects/hip/cmake/FindHIP/run_make2cmake.cmake new file mode 100644 index 0000000000..0563725aac --- /dev/null +++ b/projects/hip/cmake/FindHIP/run_make2cmake.cmake @@ -0,0 +1,70 @@ +# Copyright (c) 2016 - 2021 Advanced Micro Devices, Inc. All Rights Reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in +# all copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +# THE SOFTWARE. + +############################################################################### +# Computes dependencies using HIPCC +############################################################################### + +############################################################################### +# This file converts dependency files generated using hipcc to a format that +# cmake can understand. + +# Input variables: +# +# input_file:STRING=<> Dependency file to parse. Required argument +# output_file:STRING=<> Output file to generate. Required argument + +if(NOT input_file OR NOT output_file) + message(FATAL_ERROR "You must specify input_file and output_file on the command line") +endif() + +file(READ ${input_file} depend_text) + +if (NOT "${depend_text}" STREQUAL "") + string(REPLACE " /" "\n/" depend_text ${depend_text}) + string(REGEX REPLACE "^.*:" "" depend_text ${depend_text}) + string(REGEX REPLACE "[ \\\\]*\n" ";" depend_text ${depend_text}) + + set(dependency_list "") + + foreach(file ${depend_text}) + string(REGEX REPLACE "^ +" "" file ${file}) + if(NOT EXISTS "${file}") + message(WARNING " Removing non-existent dependency file: ${file}") + set(file "") + endif() + + if(NOT IS_DIRECTORY "${file}") + get_filename_component(file_absolute "${file}" ABSOLUTE) + list(APPEND dependency_list "${file_absolute}") + endif() + endforeach() +endif() + +# Remove the duplicate entries and sort them. +list(REMOVE_DUPLICATES dependency_list) +list(SORT dependency_list) + +foreach(file ${dependency_list}) + set(hip_hipcc_depend "${hip_hipcc_depend} \"${file}\"\n") +endforeach() + +file(WRITE ${output_file} "# Generated by: FindHIP.cmake. Do not edit.\nSET(HIP_HIPCC_DEPEND\n ${hip_hipcc_depend})\n\n") +# vim: ts=4:sw=4:expandtab:smartindent diff --git a/projects/hip/configure b/projects/hip/configure new file mode 100644 index 0000000000..e69de29bb2 diff --git a/projects/hip/docker/dockerfile-build-ubuntu-16.04 b/projects/hip/docker/dockerfile-build-ubuntu-16.04 new file mode 100644 index 0000000000..747cefc54a --- /dev/null +++ b/projects/hip/docker/dockerfile-build-ubuntu-16.04 @@ -0,0 +1,34 @@ +# Copyright (c) 2017 - 2021 Advanced Micro Devices, Inc. All Rights Reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in +# all copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +# THE SOFTWARE. + +# Parameters related to building hip +ARG base_image + +FROM ${base_image} +MAINTAINER Maneesh Gupta + +ARG user_uid + +# docker pipeline runs containers with particular uid +# create a jenkins user with this specific uid so it can use sudo priviledges +# Grant any member of sudo group password-less sudo privileges +RUN useradd --create-home -u ${user_uid} -G sudo,video --shell /bin/bash jenkins && \ + mkdir -p /etc/sudoers.d/ && \ + echo '%sudo ALL=(ALL) NOPASSWD:ALL' | tee /etc/sudoers.d/sudo-nopasswd diff --git a/projects/hip/docker/dockerfile-hip-ubuntu-16.04 b/projects/hip/docker/dockerfile-hip-ubuntu-16.04 new file mode 100644 index 0000000000..8c833f7918 --- /dev/null +++ b/projects/hip/docker/dockerfile-hip-ubuntu-16.04 @@ -0,0 +1,39 @@ +# Copyright (c) 2017 - 2021 Advanced Micro Devices, Inc. All Rights Reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in +# all copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +# THE SOFTWARE. + +# Parameters related to building hip +ARG base_image + +FROM ${base_image} +MAINTAINER Kent Knox + +# Copy the debian package of hip into the container from host +COPY *.deb /tmp/ + +# Install the debian package +RUN apt-get update && DEBIAN_FRONTEND=noninteractive apt-get install --no-install-recommends -y curl \ + && apt-get update && DEBIAN_FRONTEND=noninteractive apt-get install --no-install-recommends --allow-unauthenticated -y \ + /tmp/hip-devel-*.deb \ + /tmp/hip-runtime-amd-*.deb \ + /tmp/hip-doc-*.deb \ + /tmp/hip-samples-* \ + && rm -f /tmp/*.deb \ + && apt-get clean \ + && rm -rf /var/lib/apt/lists/* diff --git a/projects/hip/docs/.gitignore b/projects/hip/docs/.gitignore new file mode 100644 index 0000000000..76d890c082 --- /dev/null +++ b/projects/hip/docs/.gitignore @@ -0,0 +1,9 @@ +/_build +/_doxygen +/_images +/_static +/_templates +/doxygen/html +/doxygen/xml +/sphinx/_toc.yml +__pycache__ diff --git a/projects/hip/docs/conf.py b/projects/hip/docs/conf.py new file mode 100644 index 0000000000..a9d7e3f534 --- /dev/null +++ b/projects/hip/docs/conf.py @@ -0,0 +1,65 @@ +# Configuration file for the Sphinx documentation builder. +# +# This file only contains a selection of the most common options. For a full +# list see the documentation: +# https://www.sphinx-doc.org/en/master/usage/configuration.html + +import re +import sys +import subprocess +from pathlib import Path +from typing import Any, Dict, List + +from rocm_docs import ROCmDocs + +version_numbers = [] +version_file = open("../VERSION", "r") +lines = version_file.readlines() +for line in lines: + if line[0] == '#': + continue + version_numbers.append(line.strip()) +version_number = ".".join(version_numbers) +left_nav_title = f"HIP {version_number} Documentation" + +# for PDF output on Read the Docs +project = "HIP Documentation" +author = "Advanced Micro Devices, Inc." +copyright = "Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved." +version = version_number +release = version_number + +external_toc_path = "./sphinx/_toc.yml" + +docs_core = ROCmDocs(left_nav_title) +docs_core.run_doxygen(doxygen_root="doxygen", doxygen_path="doxygen/xml") +docs_core.enable_api_reference() +docs_core.setup() + +external_projects_current_project = "hip" + +for sphinx_var in ROCmDocs.SPHINX_VARS: + globals()[sphinx_var] = getattr(docs_core, sphinx_var) + +# Add the _extensions directory to Python's search path +sys.path.append(str(Path(__file__).parent / 'extension')) + +extensions += ["sphinxcontrib.doxylink", "custom_directive"] + +cpp_id_attributes = ["__global__", "__device__", "__host__", "__forceinline__", "static"] +cpp_paren_attributes = ["__declspec"] + +suppress_warnings = ["etoc.toctree"] + +numfig = False + +exclude_patterns = [ + "doxygen/mainpage.md", + "understand/glossary.md", + 'how-to/debugging_env.rst', + "data/env_variables_hip.rst" +] + +git_url = subprocess.check_output(['git', 'config', '--get', 'remote.origin.url']).strip().decode('ascii') +if git_url.find("git:") != -1: + html_theme_options = {"repository_url": "https://github.com/ROCm/hip"} \ No newline at end of file diff --git a/projects/hip/docs/data/env_variables_hip.rst b/projects/hip/docs/data/env_variables_hip.rst new file mode 100644 index 0000000000..4192db7387 --- /dev/null +++ b/projects/hip/docs/data/env_variables_hip.rst @@ -0,0 +1,278 @@ +.. meta:: + :description: HIP environment variables + :keywords: AMD, HIP, environment variables, environment + +HIP GPU isolation variables +-------------------------------------------------------------------------------- + +The GPU isolation environment variables in HIP are collected in the following table. + +.. _hip-env-isolation: +.. list-table:: + :header-rows: 1 + :widths: 70,30 + + * - **Environment variable** + - **Value** + + * - | ``ROCR_VISIBLE_DEVICES`` + | A list of device indices or UUIDs that will be exposed to applications. + - Example: ``0,GPU-DEADBEEFDEADBEEF`` + + * - | ``GPU_DEVICE_ORDINAL`` + | Devices indices exposed to OpenCL and HIP applications. + - Example: ``0,2`` + + * - | ``HIP_VISIBLE_DEVICES`` or ``CUDA_VISIBLE_DEVICES`` + | Device indices exposed to HIP applications. + - Example: ``0,2`` + +HIP profiling variables +-------------------------------------------------------------------------------- + +The profiling environment variables in HIP are collected in the following table. + +.. _hip-env-prof: +.. list-table:: + :header-rows: 1 + :widths: 70,30 + + * - **Environment variable** + - **Value** + + * - | ``HSA_CU_MASK`` + | Sets the mask on a lower level of queue creation in the driver, + | this mask will also be set for queues being profiled. + - Example: ``1:0-8`` + + * - | ``ROC_GLOBAL_CU_MASK`` + | Sets the mask on queues created by the HIP or the OpenCL runtimes, + | this mask will also be set for queues being profiled. + - Example: ``0xf``, enables only 4 CUs + + * - | ``HIP_FORCE_QUEUE_PROFILING`` + | Used to run the app as if it were run in rocprof. Forces command queue + | profiling on by default. + - | 0: Disable + | 1: Enable + +HIP debug variables +-------------------------------------------------------------------------------- + +The debugging environment variables in HIP are collected in the following table. + +.. _hip-env-debug: +.. list-table:: + :header-rows: 1 + :widths: 35,14,51 + + * - **Environment variable** + - **Default value** + - **Value** + + * - | ``AMD_LOG_LEVEL`` + | Enables HIP log on various level. + - ``0`` + - | 0: Disable log. + | 1: Enables error logs. + | 2: Enables warning logs next to lower-level logs. + | 3: Enables information logs next to lower-level logs. + | 4: Enables debug logs next to lower-level logs. + | 5: Enables debug extra logs next to lower-level logs. + + * - | ``AMD_LOG_LEVEL_FILE`` + | Sets output file for ``AMD_LOG_LEVEL``. + - stderr output + - + + * - | ``AMD_LOG_MASK`` + | Specifies HIP log filters. Here is the ` complete list of log masks `_. + - ``0x7FFFFFFF`` + - | 0x1: Log API calls. + | 0x2: Kernel and copy commands and barriers. + | 0x4: Synchronization and waiting for commands to finish. + | 0x8: Decode and display AQL packets. + | 0x10: Queue commands and queue contents. + | 0x20: Signal creation, allocation, pool. + | 0x40: Locks and thread-safety code. + | 0x80: Kernel creations and arguments, etc. + | 0x100: Copy debug. + | 0x200: Detailed copy debug. + | 0x400: Resource allocation, performance-impacting events. + | 0x800: Initialization and shutdown. + | 0x1000: Misc debug, not yet classified. + | 0x2000: Show raw bytes of AQL packet. + | 0x4000: Show code creation debug. + | 0x8000: More detailed command info, including barrier commands. + | 0x10000: Log message location. + | 0x20000: Memory allocation. + | 0x40000: Memory pool allocation, including memory in graphs. + | 0x80000: Timestamp details. + | 0xFFFFFFFF: Log always even mask flag is zero. + + * - | ``HIP_LAUNCH_BLOCKING`` + | Used for serialization on kernel execution. + - ``0`` + - | 0: Disable. Kernel executes normally. + | 1: Enable. Serializes kernel enqueue, behaves the same as ``AMD_SERIALIZE_KERNEL``. + + * - | ``HIP_VISIBLE_DEVICES`` (or ``CUDA_VISIBLE_DEVICES``) + | Only devices whose index is present in the sequence are visible to HIP + - Unset by default. + - 0,1,2: Depending on the number of devices on the system. + + * - | ``GPU_DUMP_CODE_OBJECT`` + | Dump code object. + - ``0`` + - | 0: Disable + | 1: Enable + + * - | ``AMD_SERIALIZE_KERNEL`` + | Serialize kernel enqueue. + - ``0`` + - | 0: Disable + | 1: Wait for completion before enqueue. + | 2: Wait for completion after enqueue. + | 3: Both + + * - | ``AMD_SERIALIZE_COPY`` + | Serialize copies + - ``0`` + - | 0: Disable + | 1: Wait for completion before enqueue. + | 2: Wait for completion after enqueue. + | 3: Both + + * - | ``AMD_DIRECT_DISPATCH`` + | Enable direct kernel dispatch (Currently for Linux; under development for Windows). + - ``1`` + - | 0: Disable + | 1: Enable + + * - | ``GPU_MAX_HW_QUEUES`` + | The maximum number of hardware queues allocated per device. + - ``4`` + - The variable controls how many independent hardware queues HIP runtime can create per process, + per device. If an application allocates more HIP streams than this number, then HIP runtime reuses + the same hardware queues for the new streams in a round-robin manner. Note that this maximum + number does not apply to hardware queues that are created for CU-masked HIP streams, or + cooperative queues for HIP Cooperative Groups (single queue per device). + +HIP memory management related variables +-------------------------------------------------------------------------------- + +The memory management related environment variables in HIP are collected in the +following table. + +.. _hip-env-memory: +.. list-table:: + :header-rows: 1 + :widths: 35,14,51 + + * - **Environment variable** + - **Default value** + - **Value** + + * - | ``HIP_HIDDEN_FREE_MEM`` + | Amount of memory to hide from the free memory reported by hipMemGetInfo. + - ``0`` + - | 0: Disable + | Unit: megabyte (MB) + + * - | ``HIP_HOST_COHERENT`` + | Specifies if the memory is coherent between the host and GPU in ``hipHostMalloc``. + - ``0`` + - | 0: Memory is not coherent. + | 1: Memory is coherent. + | Environment variable has effect, if the following conditions are statisfied: + | - One of the ``hipHostMallocDefault``, ``hipHostMallocPortable``, ``hipHostMallocWriteCombined`` or ``hipHostMallocNumaUser`` flag set to 1. + | - ``hipHostMallocCoherent``, ``hipHostMallocNonCoherent`` and ``hipHostMallocMapped`` flags set to 0. + + * - | ``HIP_INITIAL_DM_SIZE`` + | Set initial heap size for device malloc. + - ``8388608`` + - | Unit: Byte + | The default value corresponds to 8 MB. + + * - | ``HIP_MEM_POOL_SUPPORT`` + | Enables memory pool support in HIP. + - ``0`` + - | 0: Disable + | 1: Enable + + * - | ``HIP_MEM_POOL_USE_VM`` + | Enables memory pool support in HIP. + - | ``0``: other OS + | ``1``: Windows + - | 0: Disable + | 1: Enable + + * - | ``HIP_VMEM_MANAGE_SUPPORT`` + | Virtual Memory Management Support. + - ``1`` + - | 0: Disable + | 1: Enable + + * - | ``GPU_MAX_HEAP_SIZE`` + | Set maximum size of the GPU heap to % of board memory. + - ``100`` + - | Unit: Percentage + + * - | ``GPU_MAX_REMOTE_MEM_SIZE`` + | Maximum size that allows device memory substitution with system. + - ``2`` + - | Unit: kilobyte (KB) + + * - | ``GPU_NUM_MEM_DEPENDENCY`` + | Number of memory objects for dependency tracking. + - ``256`` + - + + * - | ``GPU_STREAMOPS_CP_WAIT`` + | Force the stream memory operation to wait on CP. + - ``0`` + - | 0: Disable + | 1: Enable + + * - | ``HSA_LOCAL_MEMORY_ENABLE`` + | Enable HSA device local memory usage. + - ``1`` + - | 0: Disable + | 1: Enable + + * - | ``PAL_ALWAYS_RESIDENT`` + | Force memory resources to become resident at allocation time. + - ``0`` + - | 0: Disable + | 1: Enable + + * - | ``PAL_PREPINNED_MEMORY_SIZE`` + | Size of prepinned memory. + - ``64`` + - | Unit: kilobyte (KB) + + * - | ``REMOTE_ALLOC`` + | Use remote memory for the global heap allocation. + - ``0`` + - | 0: Disable + | 1: Enable + +HIP miscellaneous variables +-------------------------------------------------------------------------------- + +The following table lists environment variables that are useful but relate to +different features in HIP. + +.. _hip-env-other: +.. list-table:: + :header-rows: 1 + :widths: 35,14,51 + + * - **Environment variable** + - **Default value** + - **Value** + + * - | ``HIPRTC_COMPILE_OPTIONS_APPEND`` + | Sets compile options needed for ``hiprtc`` compilation. + - None + - ``--gpu-architecture=gfx906:sramecc+:xnack``, ``-fgpu-rdc`` diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/asynchronous/sequential_async_event.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/asynchronous/sequential_async_event.drawio new file mode 100644 index 0000000000..2ea9376cf3 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/asynchronous/sequential_async_event.drawio @@ -0,0 +1,274 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/asynchronous/sequential_async_event.svg b/projects/hip/docs/data/how-to/hip_runtime_api/asynchronous/sequential_async_event.svg new file mode 100644 index 0000000000..fe52799858 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/asynchronous/sequential_async_event.svg @@ -0,0 +1,2 @@ +
time
time
default stream
default stream
H2D
data1
H2D...
H2D
data2
H2D...
kernel
data1
kernel...
kernel
data2
kernel...
D2H
data1
D2H...
D2H
data2
D2H...
H2D
data2
H2D...
kernel
data2
kernel...
stream2
stream2
D2H
data2
D2H...
H2D
data1
H2D...
kernel
data1
kernel...
stream1
stream1
D2H
data1
D2H...
Seqeuntial calls:
Seqeuntial calls:
Asynchronous calls:
Asynchronous calls:
Asynchronous calls with hipEvent:
Asynchronous calls with hipEvent: +
H2D
data2
H2D...
kernel
data2
kernel...
stream2
stream2
H2D
data1
H2D...
kernel
data1
kernel...
stream1
stream1
D2H
data2
D2H...
event
event
D2H
data1
D2H...
eventA
eventB
eventA...Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_bottom.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_bottom.drawio new file mode 100644 index 0000000000..4f1ff494f2 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_bottom.drawio @@ -0,0 +1,904 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_bottom.svg b/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_bottom.svg new file mode 100644 index 0000000000..298cd48218 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_bottom.svg @@ -0,0 +1 @@ +Block
Thread-block tile
Thread-block tile
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Thread-block tile
Thread-block tile
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
WarpBlock
Coalesced group
Coalesced group
Warp
Warp
Warp
Warp
Warp
Warp
Coalesced group
Coalesced group
Warp
Warp
Warp
Warp
Warp
WarpGridText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_top.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_top.drawio new file mode 100644 index 0000000000..e4c0c90d2d --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_top.drawio @@ -0,0 +1,5152 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_top.svg b/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_top.svg new file mode 100644 index 0000000000..ebe4794576 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/cooperative_groups/thread_hierarchy_coop_top.svg @@ -0,0 +1 @@ +Grid
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
WarpGridMulti Grid
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
WarpText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph.drawio new file mode 100644 index 0000000000..03569ac734 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph.drawio @@ -0,0 +1,76 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph.svg b/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph.svg new file mode 100644 index 0000000000..6eed6b92e5 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph.svg @@ -0,0 +1,4 @@ + + + +Stream 1
Kernel B
Kernel BStream 2
Kernel A
Kernel A
hipDeviceSynchronize
hipDeviceSynchronize
Kernel C
Kernel C
hipDeviceSynchronize
hipDeviceSynchronize
Kernel D
Kernel D
Kernel A
Kernel A
Kernel B
Kernel B
Kernel C
Kernel C
Kernel D
Kernel DText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph_speedup.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph_speedup.drawio new file mode 100644 index 0000000000..7802785f6b --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph_speedup.drawio @@ -0,0 +1,162 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph_speedup.svg b/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph_speedup.svg new file mode 100644 index 0000000000..f16123b9e2 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/hipgraph/hip_graph_speedup.svg @@ -0,0 +1,4 @@ + + + +Streams
kernel A
kernel launch A
kernel B
kernel C
kernel launch B
kernel launch C
host activity
device activity
time
kernel launch D
kernel D
device idling due to kernel launch congestion
kernel A
kernel B
kernel C
graph launch
host activity
device activity
kernel D
Graph
speedup
 \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/pageable_pinned.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/pageable_pinned.drawio new file mode 100644 index 0000000000..602c7e501d --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/pageable_pinned.drawio @@ -0,0 +1,106 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/pageable_pinned.svg b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/pageable_pinned.svg new file mode 100644 index 0000000000..8ffb8aa965 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/pageable_pinned.svg @@ -0,0 +1 @@ +Pageable data transfer
Pinned memory
Pinned memoryPinned data transfer
Pageable memory
Pageable memory
Device memory
Device memory
Pinned memory
Pinned memory
Device memory
Device memory
Host
Host
Device
Device
Device
Device
Host
HostText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/border.png b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/border.png new file mode 100644 index 0000000000..e616610c15 Binary files /dev/null and b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/border.png differ diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/clamp.png b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/clamp.png new file mode 100644 index 0000000000..63ed5f116f Binary files /dev/null and b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/clamp.png differ diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/linear.png b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/linear.png new file mode 100644 index 0000000000..2c85c0b11a Binary files /dev/null and b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/linear.png differ diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/mirror.png b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/mirror.png new file mode 100644 index 0000000000..d26a241183 Binary files /dev/null and b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/mirror.png differ diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/nearest.png b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/nearest.png new file mode 100644 index 0000000000..edfbf8cfbe Binary files /dev/null and b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/nearest.png differ diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/original.png b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/original.png new file mode 100644 index 0000000000..eaf6e7f7be Binary files /dev/null and b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/original.png differ diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/wrap.png b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/wrap.png new file mode 100644 index 0000000000..68e80befe8 Binary files /dev/null and b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/textures/wrap.png differ diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/unified_memory/um.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/unified_memory/um.drawio new file mode 100644 index 0000000000..1deeca61f5 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/unified_memory/um.drawio @@ -0,0 +1,1880 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/unified_memory/um.svg b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/unified_memory/um.svg new file mode 100644 index 0000000000..83accc3b27 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/memory_management/unified_memory/um.svg @@ -0,0 +1,9 @@ +Explicit Memory Management
CPU cores
CPU cores +
CPU
CPU +
GPU Memory
(HBM)
GPU Memory...Unified Memory Management
Unified Memory
Unified Memory
CPU Memory (RAM)
CPU Memory (RAM)
GPU
GPU +
GPU cores
GPU cores +
GPU
GPU +
CPU cores
CPU cores +
CPU
CPU +
GPU cores
GPU cores +Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/runtimes.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/runtimes.drawio new file mode 100644 index 0000000000..a01c453452 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/runtimes.drawio @@ -0,0 +1,127 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/runtimes.svg b/projects/hip/docs/data/how-to/hip_runtime_api/runtimes.svg new file mode 100644 index 0000000000..a64a7a54dc --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/runtimes.svg @@ -0,0 +1,2 @@ +
HIP Runtime API
HIP Runtime API
CUDA Driver API
CUDA Driver API
CUDA runtime
CUDA runtime
ROCr runtime
ROCr runtime
PAL
PAL
CLR
CLR
AMD Platform
AMD Platform +
NVIDIA Platform
NVIDIA Platform
hipother
hipotherText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/stream_management.drawio b/projects/hip/docs/data/how-to/hip_runtime_api/stream_management.drawio new file mode 100644 index 0000000000..2b443fe3f0 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/stream_management.drawio @@ -0,0 +1,46 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/how-to/hip_runtime_api/stream_management.svg b/projects/hip/docs/data/how-to/hip_runtime_api/stream_management.svg new file mode 100644 index 0000000000..c7a05657f1 --- /dev/null +++ b/projects/hip/docs/data/how-to/hip_runtime_api/stream_management.svg @@ -0,0 +1 @@ +Stream 1
Kernel A
Kernel AStream 2
Memory Copy
Memory Copy
hipDeviceSynchronize
hipDeviceSynchronize
Kernel B
Kernel B
Kernel C
Kernel C
Memory Copy
Memory Copy
Memory Copy
Memory Copy
Memory Copy
Memory CopyText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/conflict_free_reduction.drawio b/projects/hip/docs/data/tutorial/reduction/conflict_free_reduction.drawio new file mode 100644 index 0000000000..b1f0b51074 --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/conflict_free_reduction.drawio @@ -0,0 +1,448 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/conflict_free_reduction.svg b/projects/hip/docs/data/tutorial/reduction/conflict_free_reduction.svg new file mode 100644 index 0000000000..71eb0660ed --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/conflict_free_reduction.svg @@ -0,0 +1 @@ +Bank conflict free reduction
Shared
Shared
5
5
13
13
0
0
8
8
2
2
1
1
7
7
42
42
2
2
23
23
10
10
3
3
Shared
Shared
7
7
42
42
23
23
10
10
7
7
42
42
23
23
10
10
0
0
1
1
Shared
Shared
23
23
42
42
23
23
10
10
7
7
42
42
23
23
10
10
0
0
Shared
Shared
42
42
42
42
23
23
10
10
7
7
42
42
23
23
10
10
Thread IDs
Thread...
Data snapshot
Data sn...
Thread IDs
Thread...
Data snapshot
Data sn...
Thread IDs
Thread...
Data snapshot
Data sn...
Data snapshot
Data sn...Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/foldl.drawio b/projects/hip/docs/data/tutorial/reduction/foldl.drawio new file mode 100644 index 0000000000..1d5228da9e --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/foldl.drawio @@ -0,0 +1,142 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/foldl.svg b/projects/hip/docs/data/tutorial/reduction/foldl.svg new file mode 100644 index 0000000000..7603080193 --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/foldl.svg @@ -0,0 +1 @@ +Fold-left
Input
Input
8
8
13
13
5
5
z
z
5
5
f(z,5)
f(z,5)
5
5
13
13
f(f(z,5),13)
f(f(z,5...
13
13
8
8
f(f(f(z,5),13),8)
f(f(f(z...
Result
Result
13
13Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/naive_reduction.drawio b/projects/hip/docs/data/tutorial/reduction/naive_reduction.drawio new file mode 100644 index 0000000000..b186c58aad --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/naive_reduction.drawio @@ -0,0 +1,442 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/naive_reduction.svg b/projects/hip/docs/data/tutorial/reduction/naive_reduction.svg new file mode 100644 index 0000000000..922bfff1e9 --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/naive_reduction.svg @@ -0,0 +1 @@ +Naive Shared Reduction
Shared
Shared
5
5
13
13
0
0
8
8
2
2
2
2
7
7
42
42
4
4
23
23
10
10
6
6
Shared
Shared
13
13
13
13
0
0
8
8
2
2
42
42
42
42
4
4
23
23
10
10
Shared
Shared
13
13
13
13
0
0
8
8
2
2
42
42
42
42
23
23
10
10
Shared
Shared
42
42
13
13
8
8
2
2
42
42
42
42
23
23
10
10
Thread IDs
Thread...
Data snapshot
Data sn...
Thread IDs
Thread...
Data snapshot
Data sn...
Thread IDs
Thread...
Data snapshot
Data sn...
Data snapshot
Data sn...Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/parallel_foldl.drawio b/projects/hip/docs/data/tutorial/reduction/parallel_foldl.drawio new file mode 100644 index 0000000000..6b04c73cc2 --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/parallel_foldl.drawio @@ -0,0 +1,142 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/parallel_foldl.svg b/projects/hip/docs/data/tutorial/reduction/parallel_foldl.svg new file mode 100644 index 0000000000..d5edb0accb --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/parallel_foldl.svg @@ -0,0 +1 @@ +Parallel Reduce
Input
Input
8
8
13
13
5
5
z
z
5
5
f(z,5)
f(z,5)
13
13
8
8
f(13,8)
f(13,8)
5
5
13
13
f(f(z,5),f(13,8))
f(f(z,5...
Result
Result
13
13Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/reduced_divergence_reduction.drawio b/projects/hip/docs/data/tutorial/reduction/reduced_divergence_reduction.drawio new file mode 100644 index 0000000000..0f1bd277ad --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/reduced_divergence_reduction.drawio @@ -0,0 +1,442 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/reduced_divergence_reduction.svg b/projects/hip/docs/data/tutorial/reduction/reduced_divergence_reduction.svg new file mode 100644 index 0000000000..9661e05115 --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/reduced_divergence_reduction.svg @@ -0,0 +1 @@ +Reduced Divergence Reduction
Shared
Shared
5
5
13
13
0
0
8
8
2
2
1
1
7
7
42
42
2
2
23
23
10
10
3
3
Shared
Shared
13
13
13
13
0
0
8
8
2
2
42
42
42
42
1
1
23
23
10
10
Shared
Shared
13
13
13
13
0
0
8
8
2
2
42
42
42
42
23
23
10
10
Shared
Shared
42
42
13
13
8
8
2
2
42
42
42
42
23
23
10
10
Thread IDs
Thread...
Data snapshot
Data sn...
Thread IDs
Thread...
Data snapshot
Data sn...
Thread IDs
Thread...
Data snapshot
Data sn...
Data snapshot
Data sn...Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/warp_reduction.drawio b/projects/hip/docs/data/tutorial/reduction/warp_reduction.drawio new file mode 100644 index 0000000000..583f90cdd2 --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/warp_reduction.drawio @@ -0,0 +1,421 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/warp_reduction.svg b/projects/hip/docs/data/tutorial/reduction/warp_reduction.svg new file mode 100644 index 0000000000..ec8d0a829b --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/warp_reduction.svg @@ -0,0 +1,2 @@ +Warp reduction
Local
Local
5
5
13
13
0
0
8
8
2
2
1
1
7
7
42
42
2
2
23
23
10
10
3
3
Local
Local
7
7
42
42
23
23
10
10
7
7
42
42
23
23
10
10
0
0
1
1
Local
Local
23
23
42
42
23
23
10
10
7
7
42
42
23
23
10
10
0
0
Local
Local +
42
42
42
42
23
23
10
10
7
7
42
42
23
23
10
10
Thread IDs
Thread...
Data snapshot
Data sn...
Thread IDs
Thread...
Data snapshot
Data sn...
Thread IDs
Thread...
Data snapshot
Data sn...
Data snapshot
Data sn...Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/warp_reduction_with_shared.drawio b/projects/hip/docs/data/tutorial/reduction/warp_reduction_with_shared.drawio new file mode 100644 index 0000000000..338407f45e --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/warp_reduction_with_shared.drawio @@ -0,0 +1,707 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/tutorial/reduction/warp_reduction_with_shared.svg b/projects/hip/docs/data/tutorial/reduction/warp_reduction_with_shared.svg new file mode 100644 index 0000000000..65b6d642b8 --- /dev/null +++ b/projects/hip/docs/data/tutorial/reduction/warp_reduction_with_shared.svg @@ -0,0 +1,3 @@ +Warp reduction
Local
Local
5
5
13
13
8
8
2
2
7
7
42
42
23
23
10
10
7
7
42
42
23
23
10
10
7
7
42
42
23
23
10
10
Local
Local
23
23
42
42
23
23
10
10
7
7
42
42
23
23
10
10
Local
Local +
42
42
42
42
23
23
10
10
7
7
42
42
23
23
10
10Warp reduction
Local
Local
3
3
2
2
4
4
1
1
7
7
11
11
8
8
14
14
10
10
13
13
12
12
15
15
7
7
11
11
8
8
14
14
Local
Local
22
22
28
28
12
12
15
15
7
7
11
11
8
8
14
14
Local
Local +
50
50
28
28
12
12
15
15
7
7
11
11
8
8
14
14
Local
Local
92
92
50
50
Local
Local
42
42
50
50
Shared
Shared
42
42
50
50Warp reduction with shared memory
Local
Local
Local
LocalText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/hardware_implementation/cdna2_gcd.png b/projects/hip/docs/data/understand/hardware_implementation/cdna2_gcd.png new file mode 100644 index 0000000000..4344d7c81a Binary files /dev/null and b/projects/hip/docs/data/understand/hardware_implementation/cdna2_gcd.png differ diff --git a/projects/hip/docs/data/understand/hardware_implementation/cdna3_cu.png b/projects/hip/docs/data/understand/hardware_implementation/cdna3_cu.png new file mode 100644 index 0000000000..a1917adc5f Binary files /dev/null and b/projects/hip/docs/data/understand/hardware_implementation/cdna3_cu.png differ diff --git a/projects/hip/docs/data/understand/hardware_implementation/compute_unit.drawio b/projects/hip/docs/data/understand/hardware_implementation/compute_unit.drawio new file mode 100644 index 0000000000..3e1d18a8a8 --- /dev/null +++ b/projects/hip/docs/data/understand/hardware_implementation/compute_unit.drawio @@ -0,0 +1,187 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + diff --git a/projects/hip/docs/data/understand/hardware_implementation/compute_unit.svg b/projects/hip/docs/data/understand/hardware_implementation/compute_unit.svg new file mode 100644 index 0000000000..49adbc8e45 --- /dev/null +++ b/projects/hip/docs/data/understand/hardware_implementation/compute_unit.svg @@ -0,0 +1,4 @@ + + + +Compute Unit
Vector Caches
Vector Caches
SIMD 0
SIMD 0
Vector ALU
Vector ALU
Vector
Register
File
Vector...
SIMD 2
SIMD 2
Vector ALU
Vector ALU
Vector
Register
File
Vector...
SIMD 3
SIMD 3
Vector ALU
Vector ALU
Vector
Register
File
Vector...
SIMD 1
SIMD 1
Vector ALU
Vector ALU
Vector
Register
File
Vector...
Scalar ALU
Scalar ALU
Scalar Unit
Scalar Unit
Scalar
Register
File
Scalar...
Local Data Share
Local Data ShareText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/hardware_implementation/rdna3_cu.png b/projects/hip/docs/data/understand/hardware_implementation/rdna3_cu.png new file mode 100644 index 0000000000..2d38d2a508 Binary files /dev/null and b/projects/hip/docs/data/understand/hardware_implementation/rdna3_cu.png differ diff --git a/projects/hip/docs/data/understand/programming_model/cdna2_gcd.png b/projects/hip/docs/data/understand/programming_model/cdna2_gcd.png new file mode 100644 index 0000000000..4344d7c81a Binary files /dev/null and b/projects/hip/docs/data/understand/programming_model/cdna2_gcd.png differ diff --git a/projects/hip/docs/data/understand/programming_model/cdna3_cu.png b/projects/hip/docs/data/understand/programming_model/cdna3_cu.png new file mode 100644 index 0000000000..a1917adc5f Binary files /dev/null and b/projects/hip/docs/data/understand/programming_model/cdna3_cu.png differ diff --git a/projects/hip/docs/data/understand/programming_model/cpu-gpu-comparison.drawio b/projects/hip/docs/data/understand/programming_model/cpu-gpu-comparison.drawio new file mode 100644 index 0000000000..a7e851b3d5 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/cpu-gpu-comparison.drawio @@ -0,0 +1,181 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/cpu-gpu-comparison.svg b/projects/hip/docs/data/understand/programming_model/cpu-gpu-comparison.svg new file mode 100644 index 0000000000..552290299f --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/cpu-gpu-comparison.svg @@ -0,0 +1 @@ +
CPU versus GPU Architecture
CPU versus GPU Archite...
CPU
CPU
CPU Core
CPU Core
CPU Core
CPU Core
CPU Core
CPU Core
CPU Core
CPU Core
GPU
GPU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
CU
Large Complex Cores
Large Complex Cores
High Clock Speed (3-5 GHz)
High Clock Speed (3-5 GHz)
Many Simple Cores
Many Simple Cores
Lower Clock Speed (1-2 GHz)
Lower Clock Speed (1-2 GHz)
Large Cache per Core
Large Cache per Core
Shared Memory across Cores
Shared Memory across CoresText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/host-device-flow.drawio b/projects/hip/docs/data/understand/programming_model/host-device-flow.drawio new file mode 100644 index 0000000000..2ee8c43ae9 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/host-device-flow.drawio @@ -0,0 +1,61 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/host-device-flow.svg b/projects/hip/docs/data/understand/programming_model/host-device-flow.svg new file mode 100644 index 0000000000..02bce96c5d --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/host-device-flow.svg @@ -0,0 +1 @@ +
Host-Device Data Flow
Host-Device Data Flow
Host (CPU)
Host (CPU)
Device (GPU)
Device (GPU)
1. Initialize
1. Initialize
2. Transfer Data
2. Transfer Data
3. Execute Kernel
3. Execute Kernel
4. Return Results
4. Return ResultsText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/memory-access.drawio b/projects/hip/docs/data/understand/programming_model/memory-access.drawio new file mode 100644 index 0000000000..3577772532 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/memory-access.drawio @@ -0,0 +1,237 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/memory-access.svg b/projects/hip/docs/data/understand/programming_model/memory-access.svg new file mode 100644 index 0000000000..5f0dbd8aae --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/memory-access.svg @@ -0,0 +1 @@ +
Memory Access Patterns
Memory Access Patterns
Uncoalesced Access
Uncoalesced Access
Threads
Threads
Memory
Memory
Coalesced Access
Coalesced Access
Threads
Threads
Memory
Memory
0
0
...
...
...
...
63
63
0
0
...
...
...
...
63
63
0
0
...
...
...
...
63
63
0
0
...
...
...
...
63
63Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/memory_hierarchy.drawio b/projects/hip/docs/data/understand/programming_model/memory_hierarchy.drawio new file mode 100644 index 0000000000..21c801a62d --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/memory_hierarchy.drawio @@ -0,0 +1,400 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/memory_hierarchy.svg b/projects/hip/docs/data/understand/programming_model/memory_hierarchy.svg new file mode 100644 index 0000000000..7599e7b5d3 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/memory_hierarchy.svg @@ -0,0 +1 @@ +Grid
Block
Block
%3CmxGraphModel%3E%3Croot%3E%3CmxCell%20id%3D%220%22%2F%3E%3CmxCell%20id%3D%221%22%20parent%3D%220%22%2F%3E%3CmxCell%20id%3D%222%22%20value%3D%22%26lt%3Bfont%20face%3D%26quot%3BKlavika%26quot%3B%20style%3D%26quot%3Bfont-size%3A%2017px%3B%26quot%3B%26gt%3BCluster%20shared%26lt%3B%2Ffont%26gt%3B%22%20style%3D%22text%3Bhtml%3D1%3BstrokeColor%3Dnone%3BfillColor%3Dnone%3Balign%3Dcenter%3BverticalAlign%3Dmiddle%3BwhiteSpace%3Dwrap%3Brounded%3D0%3BfontSize%3D17%3BfontColor%3D%23FFFFFF%3B%22%20vertex%3D%221%22%20parent%3D%221%22%3E%3CmxGeometry%20x%3D%221007.5%22%20y%3D%22150%22%20width%3D%22115%22%20height%3D%2230%22%20as%3D%22geometry%22%2F%3E%3C%2FmxCell%3E%3C%2Froot%3E%3C%2FmxGraphModel%3E
%3CmxGraphModel%3E...
Warp
Warp
Warp
Warp
Local
Local
Shared
Shared
Block
Block
Warp
Warp
Warp
Warp
Local
Local
Shared
Shared
Global
Global
%3CmxGraphModel%3E%3Croot%3E%3CmxCell%20id%3D%220%22%2F%3E%3CmxCell%20id%3D%221%22%20parent%3D%220%22%2F%3E%3CmxCell%20id%3D%222%22%20value%3D%22%26lt%3Bfont%20face%3D%26quot%3BKlavika%26quot%3B%20style%3D%26quot%3Bfont-size%3A%2017px%3B%26quot%3B%26gt%3BCluster%20shared%26lt%3B%2Ffont%26gt%3B%22%20style%3D%22text%3Bhtml%3D1%3BstrokeColor%3Dnone%3BfillColor%3Dnone%3Balign%3Dcenter%3BverticalAlign%3Dmiddle%3BwhiteSpace%3Dwrap%3Brounded%3D0%3BfontSize%3D17%3BfontColor%3D%23FFFFFF%3B%22%20vertex%3D%221%22%20parent%3D%221%22%3E%3CmxGeometry%20x%3D%221007.5%22%20y%3D%22150%22%20width%3D%22115%22%20height%3D%2230%22%20as%3D%22geometry%22%2F%3E%3C%2FmxCell%3E%3C%2Froot%3E%3C%2FmxGraphModel%3E
%3CmxGraphModel...
Constant
Constant
Texture
Texture
Surface
SurfaceText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/multi-gpu.drawio b/projects/hip/docs/data/understand/programming_model/multi-gpu.drawio new file mode 100644 index 0000000000..17eca3c318 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/multi-gpu.drawio @@ -0,0 +1,64 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/multi-gpu.svg b/projects/hip/docs/data/understand/programming_model/multi-gpu.svg new file mode 100644 index 0000000000..190f2593d2 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/multi-gpu.svg @@ -0,0 +1 @@ +
Multi-GPU Workload Distribution
Multi-GPU Workload Distribution
Host CPU
Host CPU
GPU 0
GPU 0
GPU 1
GPU 1
GPU 2
GPU 2
GPU 3
GPU 3
25%
25%
25%
25%
25%
25%
25%
25%Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/rdna3_cu.png b/projects/hip/docs/data/understand/programming_model/rdna3_cu.png new file mode 100644 index 0000000000..2d38d2a508 Binary files /dev/null and b/projects/hip/docs/data/understand/programming_model/rdna3_cu.png differ diff --git a/projects/hip/docs/data/understand/programming_model/simt-execution.drawio b/projects/hip/docs/data/understand/programming_model/simt-execution.drawio new file mode 100644 index 0000000000..1e2652f51f --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/simt-execution.drawio @@ -0,0 +1,124 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/simt-execution.svg b/projects/hip/docs/data/understand/programming_model/simt-execution.svg new file mode 100644 index 0000000000..412b9265e7 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/simt-execution.svg @@ -0,0 +1 @@ +
SIMT Execution Model
SIMT Execution Model
a[i] = b[i] + c[i]
a[i] = b[i] + c[i]
Thread 0
Thread 0
b[0] = 5
b[0] = 5
c[0] = 3
c[0] = 3
a[0] = 8
a[0] = 8
Thread 1
Thread 1
b[1] = 2
b[1] = 2
c[1] = 4
c[1] = 4
a[1] = 6
a[1] = 6
Thread 2
Thread 2
b[2] = 7
b[2] = 7
c[2] = 1
c[2] = 1
a[2] = 8
a[2] = 8
Thread 3
Thread 3
b[3] = 3
b[3] = 3
c[3] = 5
c[3] = 5
a[3] = 8
a[3] = 8Text is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/stream-workflow.drawio b/projects/hip/docs/data/understand/programming_model/stream-workflow.drawio new file mode 100644 index 0000000000..616dd28d78 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/stream-workflow.drawio @@ -0,0 +1,97 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/stream-workflow.svg b/projects/hip/docs/data/understand/programming_model/stream-workflow.svg new file mode 100644 index 0000000000..9648351cad --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/stream-workflow.svg @@ -0,0 +1 @@ +
Stream and Event Workflow
Stream and Event Workf...
Stream 1
Stream 1
Stream 2
Stream 2
Stream 3
Stream 3
Operation
Operation
Event
EventText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/thread_hierarchy.drawio b/projects/hip/docs/data/understand/programming_model/thread_hierarchy.drawio new file mode 100644 index 0000000000..61ac9aa59c --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/thread_hierarchy.drawio @@ -0,0 +1,8332 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/understand/programming_model/thread_hierarchy.svg b/projects/hip/docs/data/understand/programming_model/thread_hierarchy.svg new file mode 100644 index 0000000000..62a1b14d50 --- /dev/null +++ b/projects/hip/docs/data/understand/programming_model/thread_hierarchy.svg @@ -0,0 +1,3 @@ + + +Grid
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Block
Block
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
Warp
WarpText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/data/what_is_hip/hip.drawio b/projects/hip/docs/data/what_is_hip/hip.drawio new file mode 100644 index 0000000000..1a47e4b097 --- /dev/null +++ b/projects/hip/docs/data/what_is_hip/hip.drawio @@ -0,0 +1,157 @@ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + \ No newline at end of file diff --git a/projects/hip/docs/data/what_is_hip/hip.svg b/projects/hip/docs/data/what_is_hip/hip.svg new file mode 100644 index 0000000000..c151dc8717 --- /dev/null +++ b/projects/hip/docs/data/what_is_hip/hip.svg @@ -0,0 +1,2 @@ +
NVIDIA runtime
NVIDIA runtime
NVIDIA Platform
NVIDIA Platform
HIP
HIP
AMD runtime
AMD runtime
AMD Platform
AMD Platform +
hipLibrary
hipLibrary
rocLibrary
rocLibrary
cuLibrary
cuLibrary
Application Implementation
Application Implementation
Application
Application
runtime API
runtime API
kernel language
kernel languageText is not SVG - cannot display \ No newline at end of file diff --git a/projects/hip/docs/device_md_gen.py b/projects/hip/docs/device_md_gen.py new file mode 100644 index 0000000000..3a261bcdda --- /dev/null +++ b/projects/hip/docs/device_md_gen.py @@ -0,0 +1,509 @@ +""" +Copyright (c) 2015-2021 Advanced Micro Devices, Inc. All rights reserved. + +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +""" + +""" +1. This files uses Python3 to run + +List of device functions: +acosf +acoshf +asinf +asinhf +atan2f +atanf +atanhf +cbrtf +ceilf +copysignf +cosf +coshf +cospif +cyl_bessel_i0f +cyl_bessel_i1f +erfcf +erfcinvf +erfcxf +erff +erfinvf +exp10f +exp2f +expf +expm1f +fabsf +fdimf +fdividef +floorf +fmaf +fmaxf +fminf +fmodf +frexpf +hypotf +ilogbf +isfinite +isinf +isnan +j0f +j1f +jnf +ldexpf +lgammaf +llrintf +llroundf +log10f +log1pf +logbf +lrintf +lroundf +modff +nanf +nearbyintf +nextafterf +norm3df +norm4df +normcdff +normcdfinvf +normf +powf +rcbrtf +remainderf +remquof +rhypotf +rintf +rnorm3df +rnorm4df +rnormf +roundf +rsqrtf +scalblnf +scalbnf +signbit +sincosf +sincospif +sinf +sinhf +sinpif +sqrtf +tanf +tanhf +tgammaf +truncf +y0f +y1f +ynf +acos +acosh +asin +asinh +atan +atan2 +atanh +cbrt +ceil +copysign +cos +cosh +cospi +cyl_bessel_i0 +cyl_bessel_i1 +erf +erfc +erfcinv +erfcx +erfinv +exp +exp10 +exp2 +expm1 +fabs +fdim +floor +fma +fmax +fmin +fmod +frexp +hypot +ilogb +isfinite +isinf +isnan +j0 +j1 +jn +ldexp +lgamma +llrint +llround +log +log10 +log1p +log2 +logb +lrint +lround +modf +nan +nearbyint +nextafter +norm +norm3d +norm4d +normcdf +normcdfinv +pow +rcbrt +remainder +remquo +rhypot +rint +rnorm +rnorm3d +rnorm4d +round +rsqrt +scalbln +scalbn +signbit +sin +sincos +sincospi +sinh +sinpi +sqrt +tan +tanh +tgamma +trunc +y0 +y1 +yn +__cosf +__exp10f +__expf +__fadd_rd +__fadd_rn +__fadd_ru +__fadd_rz +__fdiv_rd +__fdiv_rn +__fdiv_ru +__fdiv_rz +__fdividef +__fmaf_rd +__fmaf_rn +__fmaf_ru +__fmaf_rz +__fmul_rd +__fmul_rn +__fmul_ru +__fmul_rz +__frcp_rd +__frcp_rn +__frcp_ru +__frcp_rz +__frsqrt_rn +__fsqrt_rd +__fsqrt_rn +__fsqrt_ru +__fsqrt_rz +__fsub_rd +__fsub_rn +__fsub_ru +__log10f +__log2f +__logf +__powf +__saturatef +__sincosf +__sinf +__tanf +__dadd_rd +__dadd_rn +__dadd_ru +__dadd_rz +__ddiv_rd +__ddiv_rn +__ddiv_ru +__ddiv_rz +__dmul_rd +__dmul_rn +__dmul_ru +__dmul_rz +__drcp_rd +__drcp_rn +__drcp_ru +__drcp_rz +__dsqrt_rd +__dsqrt_rn +__dsqrt_ru +__dsqrt_rz +__dsub_rd +__dsub_rn +__dsub_ru +__dsub_rz +__fma_rd +__fma_rn +__fma_ru +__fma_rz +__brev +__brevll +__byte_perm +__clz +__clzll +__ffs +__ffsll +__hadd +__mul24 +__mul64hi +__mulhi +__popc +__popcll +__rhadd +__sad +__uhadd +__umul24 +__umul64hi +__umulhi +__urhadd +__usad +__double2float_rd +__double2float_rn +__double2float_ru +__double2float_rz +__double2hiint +__double2int_rd +__double2int_rn +__double2int_ru +__double2int_rz +__double2ll_rd +__double2ll_rn +__double2ll_ru +__double2ll_rz +__double2loint +__double2uint_rd +__double2uint_rn +__double2uint_ru +__double2uint_rz +__double2ull_rd +__double2ull_rn +__double2ull_ru +__double2ull_rz +__double_as_longlong +__float2half_rn +__half2float +__float2half_rn +__half2float +__float2int_rd +__float2int_rn +__float2int_ru +__float2int_rz +__float2ll_rd +__float2ll_rn +__float2ll_ru +__float2ll_rz +__float2uint_rd +__float2uint_rn +__float2uint_ru +__float2uint_rz +__float2ull_rd +__float2ull_rn +__float2ull_ru +__float2ull_rz +__float_as_int +__float_as_uint +__hiloint2double +__int2double_rn +__int2float_rd +__int2float_rn +__int2float_ru +__int2float_rz +__int_as_float +__ll2double_rd +__ll2double_rn +__ll2double_ru +__ll2double_rz +__ll2float_rd +__ll2float_rn +__ll2float_ru +__ll2float_rz +__longlong_as_double +__uint2double_rn +__uint2float_rd +__uint2float_rn +__uint2float_ru +__uint2float_rz +__uint_as_float +__ull2double_rd +__ull2double_rn +__ull2double_ru +__ull2double_rz +__ull2float_rd +__ull2float_rn +__ull2float_ru +__ull2float_rz +__heq +__hge +__hgt +__hisinf +__hisnan +__hle +__hlt +__hne +__hbeq2 +__hbge2 +__hbgt2 +__hble2 +__hblt2 +__hbne2 +__heq2 +__hge2 +__hgt2 +__hisnan2 +__hle2 +__hlt2 +__hne2 +__float22half2_rn +__float2half +__float2half2_rn +__float2half_rd +__float2half_rn +__float2half_ru +__float2half_rz +__floats2half2_rn +__half22float2 +__half2float +half2half2 +__half2int_rd +__half2int_rn +__half2int_ru +__half2int_rz +__half2ll_rd +__half2ll_rn +__half2ll_ru +__half2ll_rz +__half2short_rd +__half2short_rn +__half2short_ru +__half2short_rz +__half2uint_rd +__half2uint_rn +__half2uint_ru +__half2uint_rz +__half2ull_rd +__half2ull_rn +__half2ull_ru +__half2ull_rz +__half2ushort_rd +__half2ushort_rn +__half2ushort_ru +__half2ushort_rz +__half_as_short +__half_as_ushort +__halves2half2 +__high2float +__high2half +__high2half2 +__highs2half2 +__int2half_rd +__int2half_rn +__int2half_ru +__int2half_rz +__ll2half_rd +__ll2half_rn +__ll2half_ru +__ll2half_rz +__low2float +__low2half +__low2half2 +__low2half2 +__lowhigh2highlow +__lows2half2 +__short2half_rd +__short2half_rn +__short2half_ru +__short2half_rz +__uint2half_rd +__uint2half_rn +__uint2half_ru +__uint2half_rz +__ull2half_rd +__ull2half_rn +__ull2half_ru +__ull2half_rz +__ushort2half_rd +__ushort2half_rn +__ushort2half_ru +__ushort2half_rz +__ushort_as_half +""" +# The dictionary is to place description of each device function. Expand it to all the device functions +deviceFuncDesc = {'acosf': "This function returns floating point of arc cosine from a floating point input"} + +fnames = ["../../include/hip/amd_detail/math_functions.h","../../include/hip/amd_detail/device_functions.h","../../include/hip/amd_detail/hip_fp16.h"] +markdownFileName = "./hip-math-api.md" + +preamble = "# HIP MATH APIs Documentation \n"+\ +"HIP supports most of the device functions supported by CUDA. Way to find the unsupported one is to search for the function and check its description\n" + \ +"Note: This document is not human generated. Any changes to this file will be discarded. Please make changes to Python3 script docs/markdown/device_md_gen.py\n\n" + \ +"## For Developers \n" + \ +"If you add or fixed a device function, make sure to add a signature of the function and definition later.\n" + \ +"For example, if you want to add `__device__ float __dotf(float4, float4)`, which does a dot product on 4 float vector components \n" + \ +"The way to add to the header is, \n" + \ +"```cpp \n" + \ +"__device__ static float __dotf(float4, float4); \n" + \ +"/*Way down in the file....*/\n" + \ +"__device__ static inline float __dotf(float4 x, float4 y) { \n" + \ +" /*implementation*/\n}\n" + \ +"```\n\n" + \ +"This helps python script to add the device function newly declared into markdown documentation (as it looks at functions with `;` at the end and `__device__` at the beginning)\n\n" + \ +"The next step would be to add Description to `deviceFuncDesc` dictionary in python script.\n" + \ +"From the above example, it can be writtern as,\n`deviceFuncDesc['__dotf'] = 'This functions takes 2 4 component float vector and outputs dot product across them'`\n\n" + +def generateSnippet(name, description, signature): + return "### " + name + "\n" + \ + "```cpp \n" + signature + "\n```\n" + \ + "**Description:** " + description + "\n\n\n" + +def getName(line): + l1 = line.split('(') + l2 = l1[0].split(' ') + return l2[-1] + +with open(markdownFileName, 'w') as mdfd: + mdfd.truncate() + mdfd.write(preamble) + for fname in fnames: + with open(fname) as fd: + lines = fd.readlines() + for line in lines: + if line.find('HIP_FAST_MATH') != -1: + break; + if line.find('__device__') != -1 and line.find(';') != -1 and line.find('hip') == -1: + name = getName(line) + if line.find('//') == -1: + if name in deviceFuncDesc: + mdfd.write(generateSnippet(name, deviceFuncDesc[name], line)) + else: + mdfd.write(generateSnippet(name, "Supported", line)) + else: + mdfd.write(generateSnippet(name, "**NOT Supported**", line)) + fd.close() + mdfd.close() diff --git a/projects/hip/docs/doxygen-input/doxy.cfg b/projects/hip/docs/doxygen-input/doxy.cfg new file mode 100644 index 0000000000..f4e08a678f --- /dev/null +++ b/projects/hip/docs/doxygen-input/doxy.cfg @@ -0,0 +1,2583 @@ +# Doxyfile 1.8.20 + +# This file describes the settings to be used by the documentation system +# doxygen (www.doxygen.org) for a project. +# +# All text after a double hash (##) is considered a comment and is placed in +# front of the TAG it is preceding. +# +# All text after a single hash (#) is considered a comment and will be ignored. +# The format is: +# TAG = value [value, ...] +# For lists, items can also be appended using: +# TAG += value [value, ...] +# Values that contain spaces should be placed between quotes (\" \"). + +#--------------------------------------------------------------------------- +# Project related configuration options +#--------------------------------------------------------------------------- + +# This tag specifies the encoding used for all characters in the configuration +# file that follow. The default is UTF-8 which is also the encoding used for all +# text before the first occurrence of this tag. Doxygen uses libiconv (or the +# iconv built into libc) for the transcoding. See +# https://www.gnu.org/software/libiconv/ for the list of possible encodings. +# The default value is: UTF-8. + +DOXYFILE_ENCODING = UTF-8 + +# The PROJECT_NAME tag is a single word (or a sequence of words surrounded by +# double-quotes, unless you are using Doxywizard) that should identify the +# project for which the documentation is generated. This name is used in the +# title of most generated pages and in a few other places. +# The default value is: My Project. + +PROJECT_NAME = "HIP: Heterogenous-computing Interface for Portability" + +# The PROJECT_NUMBER tag can be used to enter a project or revision number. This +# could be handy for archiving the generated documentation or if some version +# control system is used. + +PROJECT_NUMBER = + +# Using the PROJECT_BRIEF tag one can provide an optional one line description +# for a project that appears at the top of each page and should give viewer a +# quick idea about the purpose of the project. Keep the description short. + +PROJECT_BRIEF = + +# With the PROJECT_LOGO tag one can specify a logo or an icon that is included +# in the documentation. The maximum height of the logo should not exceed 55 +# pixels and the maximum width should not exceed 200 pixels. Doxygen will copy +# the logo to the output directory. + +PROJECT_LOGO = + +# The OUTPUT_DIRECTORY tag is used to specify the (relative or absolute) path +# into which the generated documentation will be written. If a relative path is +# entered, it will be relative to the location where doxygen was started. If +# left blank the current directory will be used. + +OUTPUT_DIRECTORY = RuntimeAPI + +# If the CREATE_SUBDIRS tag is set to YES then doxygen will create 4096 sub- +# directories (in 2 levels) under the output directory of each output format and +# will distribute the generated files over these directories. Enabling this +# option can be useful when feeding doxygen a huge amount of source files, where +# putting all generated files in the same directory would otherwise causes +# performance problems for the file system. +# The default value is: NO. + +CREATE_SUBDIRS = NO + +# If the ALLOW_UNICODE_NAMES tag is set to YES, doxygen will allow non-ASCII +# characters to appear in the names of generated files. If set to NO, non-ASCII +# characters will be escaped, for example _xE3_x81_x84 will be used for Unicode +# U+3044. +# The default value is: NO. + +ALLOW_UNICODE_NAMES = NO + +# The OUTPUT_LANGUAGE tag is used to specify the language in which all +# documentation generated by doxygen is written. Doxygen will use this +# information to generate all constant output in the proper language. +# Possible values are: Afrikaans, Arabic, Armenian, Brazilian, Catalan, Chinese, +# Chinese-Traditional, Croatian, Czech, Danish, Dutch, English (United States), +# Esperanto, Farsi (Persian), Finnish, French, German, Greek, Hungarian, +# Indonesian, Italian, Japanese, Japanese-en (Japanese with English messages), +# Korean, Korean-en (Korean with English messages), Latvian, Lithuanian, +# Macedonian, Norwegian, Persian (Farsi), Polish, Portuguese, Romanian, Russian, +# Serbian, Serbian-Cyrillic, Slovak, Slovene, Spanish, Swedish, Turkish, +# Ukrainian and Vietnamese. +# The default value is: English. + +OUTPUT_LANGUAGE = English + +# The OUTPUT_TEXT_DIRECTION tag is used to specify the direction in which all +# documentation generated by doxygen is written. Doxygen will use this +# information to generate all generated output in the proper direction. +# Possible values are: None, LTR, RTL and Context. +# The default value is: None. + +OUTPUT_TEXT_DIRECTION = None + +# If the BRIEF_MEMBER_DESC tag is set to YES, doxygen will include brief member +# descriptions after the members that are listed in the file and class +# documentation (similar to Javadoc). Set to NO to disable this. +# The default value is: YES. + +BRIEF_MEMBER_DESC = YES + +# If the REPEAT_BRIEF tag is set to YES, doxygen will prepend the brief +# description of a member or function before the detailed description +# +# Note: If both HIDE_UNDOC_MEMBERS and BRIEF_MEMBER_DESC are set to NO, the +# brief descriptions will be completely suppressed. +# The default value is: YES. + +REPEAT_BRIEF = YES + +# This tag implements a quasi-intelligent brief description abbreviator that is +# used to form the text in various listings. Each string in this list, if found +# as the leading text of the brief description, will be stripped from the text +# and the result, after processing the whole list, is used as the annotated +# text. Otherwise, the brief description is used as-is. If left blank, the +# following values are used ($name is automatically replaced with the name of +# the entity):The $name class, The $name widget, The $name file, is, provides, +# specifies, contains, represents, a, an and the. + +ABBREVIATE_BRIEF = "The $name class" \ + "The $name widget" \ + "The $name file" \ + is \ + provides \ + specifies \ + contains \ + represents \ + a \ + an \ + the + +# If the ALWAYS_DETAILED_SEC and REPEAT_BRIEF tags are both set to YES then +# doxygen will generate a detailed section even if there is only a brief +# description. +# The default value is: NO. + +ALWAYS_DETAILED_SEC = NO + +# If the INLINE_INHERITED_MEMB tag is set to YES, doxygen will show all +# inherited members of a class in the documentation of that class as if those +# members were ordinary class members. Constructors, destructors and assignment +# operators of the base classes will not be shown. +# The default value is: NO. + +INLINE_INHERITED_MEMB = NO + +# If the FULL_PATH_NAMES tag is set to YES, doxygen will prepend the full path +# before files name in the file list and in the header files. If set to NO the +# shortest path that makes the file name unique will be used +# The default value is: YES. + +FULL_PATH_NAMES = YES + +# The STRIP_FROM_PATH tag can be used to strip a user-defined part of the path. +# Stripping is only done if one of the specified strings matches the left-hand +# part of the path. The tag can be used to show relative paths in the file list. +# If left blank the directory from which doxygen is run is used as the path to +# strip. +# +# Note that you can specify absolute paths here, but also relative paths, which +# will be relative from the directory where doxygen is started. +# This tag requires that the tag FULL_PATH_NAMES is set to YES. + +STRIP_FROM_PATH = + +# The STRIP_FROM_INC_PATH tag can be used to strip a user-defined part of the +# path mentioned in the documentation of a class, which tells the reader which +# header file to include in order to use a class. If left blank only the name of +# the header file containing the class definition is used. Otherwise one should +# specify the list of include paths that are normally passed to the compiler +# using the -I flag. + +STRIP_FROM_INC_PATH = + +# If the SHORT_NAMES tag is set to YES, doxygen will generate much shorter (but +# less readable) file names. This can be useful is your file systems doesn't +# support long names like on DOS, Mac, or CD-ROM. +# The default value is: NO. + +SHORT_NAMES = YES + +# If the JAVADOC_AUTOBRIEF tag is set to YES then doxygen will interpret the +# first line (until the first dot) of a Javadoc-style comment as the brief +# description. If set to NO, the Javadoc-style will behave just like regular Qt- +# style comments (thus requiring an explicit @brief command for a brief +# description.) +# The default value is: NO. + +JAVADOC_AUTOBRIEF = NO + +# If the JAVADOC_BANNER tag is set to YES then doxygen will interpret a line +# such as +# /*************** +# as being the beginning of a Javadoc-style comment "banner". If set to NO, the +# Javadoc-style will behave just like regular comments and it will not be +# interpreted by doxygen. +# The default value is: NO. + +JAVADOC_BANNER = NO + +# If the QT_AUTOBRIEF tag is set to YES then doxygen will interpret the first +# line (until the first dot) of a Qt-style comment as the brief description. If +# set to NO, the Qt-style will behave just like regular Qt-style comments (thus +# requiring an explicit \brief command for a brief description.) +# The default value is: NO. + +QT_AUTOBRIEF = NO + +# The MULTILINE_CPP_IS_BRIEF tag can be set to YES to make doxygen treat a +# multi-line C++ special comment block (i.e. a block of //! or /// comments) as +# a brief description. This used to be the default behavior. The new default is +# to treat a multi-line C++ comment block as a detailed description. Set this +# tag to YES if you prefer the old behavior instead. +# +# Note that setting this tag to YES also means that rational rose comments are +# not recognized any more. +# The default value is: NO. + +MULTILINE_CPP_IS_BRIEF = NO + +# By default Python docstrings are displayed as preformatted text and doxygen's +# special commands cannot be used. By setting PYTHON_DOCSTRING to NO the +# doxygen's special commands can be used and the contents of the docstring +# documentation blocks is shown as doxygen documentation. +# The default value is: YES. + +PYTHON_DOCSTRING = YES + +# If the INHERIT_DOCS tag is set to YES then an undocumented member inherits the +# documentation from any documented member that it re-implements. +# The default value is: YES. + +INHERIT_DOCS = YES + +# If the SEPARATE_MEMBER_PAGES tag is set to YES then doxygen will produce a new +# page for each member. If set to NO, the documentation of a member will be part +# of the file/class/namespace that contains it. +# The default value is: NO. + +SEPARATE_MEMBER_PAGES = NO + +# The TAB_SIZE tag can be used to set the number of spaces in a tab. Doxygen +# uses this value to replace tabs by spaces in code fragments. +# Minimum value: 1, maximum value: 16, default value: 4. + +TAB_SIZE = 4 + +# This tag can be used to specify a number of aliases that act as commands in +# the documentation. An alias has the form: +# name=value +# For example adding +# "sideeffect=@par Side Effects:\n" +# will allow you to put the command \sideeffect (or @sideeffect) in the +# documentation, which will result in a user-defined paragraph with heading +# "Side Effects:". You can put \n's in the value part of an alias to insert +# newlines (in the resulting output). You can put ^^ in the value part of an +# alias to insert a newline as if a physical newline was in the original file. +# When you need a literal { or } or , in the value part of an alias you have to +# escape them by means of a backslash (\), this can lead to conflicts with the +# commands \{ and \} for these it is advised to use the version @{ and @} or use +# a double escape (\\{ and \\}) + +ALIASES = + +# Set the OPTIMIZE_OUTPUT_FOR_C tag to YES if your project consists of C sources +# only. Doxygen will then generate output that is more tailored for C. For +# instance, some of the names that are used will be different. The list of all +# members will be omitted, etc. +# The default value is: NO. + +OPTIMIZE_OUTPUT_FOR_C = NO + +# Set the OPTIMIZE_OUTPUT_JAVA tag to YES if your project consists of Java or +# Python sources only. Doxygen will then generate output that is more tailored +# for that language. For instance, namespaces will be presented as packages, +# qualified scopes will look different, etc. +# The default value is: NO. + +OPTIMIZE_OUTPUT_JAVA = NO + +# Set the OPTIMIZE_FOR_FORTRAN tag to YES if your project consists of Fortran +# sources. Doxygen will then generate output that is tailored for Fortran. +# The default value is: NO. + +OPTIMIZE_FOR_FORTRAN = NO + +# Set the OPTIMIZE_OUTPUT_VHDL tag to YES if your project consists of VHDL +# sources. Doxygen will then generate output that is tailored for VHDL. +# The default value is: NO. + +OPTIMIZE_OUTPUT_VHDL = NO + +# Set the OPTIMIZE_OUTPUT_SLICE tag to YES if your project consists of Slice +# sources only. Doxygen will then generate output that is more tailored for that +# language. For instance, namespaces will be presented as modules, types will be +# separated into more groups, etc. +# The default value is: NO. + +OPTIMIZE_OUTPUT_SLICE = NO + +# Doxygen selects the parser to use depending on the extension of the files it +# parses. With this tag you can assign which parser to use for a given +# extension. Doxygen has a built-in mapping, but you can override or extend it +# using this tag. The format is ext=language, where ext is a file extension, and +# language is one of the parsers supported by doxygen: IDL, Java, JavaScript, +# Csharp (C#), C, C++, D, PHP, md (Markdown), Objective-C, Python, Slice, VHDL, +# Fortran (fixed format Fortran: FortranFixed, free formatted Fortran: +# FortranFree, unknown formatted Fortran: Fortran. In the later case the parser +# tries to guess whether the code is fixed or free formatted code, this is the +# default for Fortran type files). For instance to make doxygen treat .inc files +# as Fortran files (default is PHP), and .f files as C (default is Fortran), +# use: inc=Fortran f=C. +# +# Note: For files without extension you can use no_extension as a placeholder. +# +# Note that for custom extensions you also need to set FILE_PATTERNS otherwise +# the files are not read by doxygen. + +EXTENSION_MAPPING = + +# If the MARKDOWN_SUPPORT tag is enabled then doxygen pre-processes all comments +# according to the Markdown format, which allows for more readable +# documentation. See https://daringfireball.net/projects/markdown/ for details. +# The output of markdown processing is further processed by doxygen, so you can +# mix doxygen, HTML, and XML commands with Markdown formatting. Disable only in +# case of backward compatibilities issues. +# The default value is: YES. + +MARKDOWN_SUPPORT = YES + +# When the TOC_INCLUDE_HEADINGS tag is set to a non-zero value, all headings up +# to that level are automatically included in the table of contents, even if +# they do not have an id attribute. +# Note: This feature currently applies only to Markdown headings. +# Minimum value: 0, maximum value: 99, default value: 5. +# This tag requires that the tag MARKDOWN_SUPPORT is set to YES. + +TOC_INCLUDE_HEADINGS = 5 + +# When enabled doxygen tries to link words that correspond to documented +# classes, or namespaces to their corresponding documentation. Such a link can +# be prevented in individual cases by putting a % sign in front of the word or +# globally by setting AUTOLINK_SUPPORT to NO. +# The default value is: YES. + +AUTOLINK_SUPPORT = YES + +# If you use STL classes (i.e. std::string, std::vector, etc.) but do not want +# to include (a tag file for) the STL sources as input, then you should set this +# tag to YES in order to let doxygen match functions declarations and +# definitions whose arguments contain STL classes (e.g. func(std::string); +# versus func(std::string) {}). This also make the inheritance and collaboration +# diagrams that involve STL classes more complete and accurate. +# The default value is: NO. + +BUILTIN_STL_SUPPORT = NO + +# If you use Microsoft's C++/CLI language, you should set this option to YES to +# enable parsing support. +# The default value is: NO. + +CPP_CLI_SUPPORT = NO + +# Set the SIP_SUPPORT tag to YES if your project consists of sip (see: +# https://www.riverbankcomputing.com/software/sip/intro) sources only. Doxygen +# will parse them like normal C++ but will assume all classes use public instead +# of private inheritance when no explicit protection keyword is present. +# The default value is: NO. + +SIP_SUPPORT = NO + +# For Microsoft's IDL there are propget and propput attributes to indicate +# getter and setter methods for a property. Setting this option to YES will make +# doxygen to replace the get and set methods by a property in the documentation. +# This will only work if the methods are indeed getting or setting a simple +# type. If this is not the case, or you want to show the methods anyway, you +# should set this option to NO. +# The default value is: YES. + +IDL_PROPERTY_SUPPORT = YES + +# If member grouping is used in the documentation and the DISTRIBUTE_GROUP_DOC +# tag is set to YES then doxygen will reuse the documentation of the first +# member in the group (if any) for the other members of the group. By default +# all members of a group must be documented explicitly. +# The default value is: NO. + +DISTRIBUTE_GROUP_DOC = NO + +# If one adds a struct or class to a group and this option is enabled, then also +# any nested class or struct is added to the same group. By default this option +# is disabled and one has to add nested compounds explicitly via \ingroup. +# The default value is: NO. + +GROUP_NESTED_COMPOUNDS = NO + +# Set the SUBGROUPING tag to YES to allow class member groups of the same type +# (for instance a group of public functions) to be put as a subgroup of that +# type (e.g. under the Public Functions section). Set it to NO to prevent +# subgrouping. Alternatively, this can be done per class using the +# \nosubgrouping command. +# The default value is: YES. + +SUBGROUPING = YES + +# When the INLINE_GROUPED_CLASSES tag is set to YES, classes, structs and unions +# are shown inside the group in which they are included (e.g. using \ingroup) +# instead of on a separate page (for HTML and Man pages) or section (for LaTeX +# and RTF). +# +# Note that this feature does not work in combination with +# SEPARATE_MEMBER_PAGES. +# The default value is: NO. + +INLINE_GROUPED_CLASSES = NO + +# When the INLINE_SIMPLE_STRUCTS tag is set to YES, structs, classes, and unions +# with only public data fields or simple typedef fields will be shown inline in +# the documentation of the scope in which they are defined (i.e. file, +# namespace, or group documentation), provided this scope is documented. If set +# to NO, structs, classes, and unions are shown on a separate page (for HTML and +# Man pages) or section (for LaTeX and RTF). +# The default value is: NO. + +INLINE_SIMPLE_STRUCTS = NO + +# When TYPEDEF_HIDES_STRUCT tag is enabled, a typedef of a struct, union, or +# enum is documented as struct, union, or enum with the name of the typedef. So +# typedef struct TypeS {} TypeT, will appear in the documentation as a struct +# with name TypeT. When disabled the typedef will appear as a member of a file, +# namespace, or class. And the struct will be named TypeS. This can typically be +# useful for C code in case the coding convention dictates that all compound +# types are typedef'ed and only the typedef is referenced, never the tag name. +# The default value is: NO. + +TYPEDEF_HIDES_STRUCT = NO + +# The size of the symbol lookup cache can be set using LOOKUP_CACHE_SIZE. This +# cache is used to resolve symbols given their name and scope. Since this can be +# an expensive process and often the same symbol appears multiple times in the +# code, doxygen keeps a cache of pre-resolved symbols. If the cache is too small +# doxygen will become slower. If the cache is too large, memory is wasted. The +# cache size is given by this formula: 2^(16+LOOKUP_CACHE_SIZE). The valid range +# is 0..9, the default is 0, corresponding to a cache size of 2^16=65536 +# symbols. At the end of a run doxygen will report the cache usage and suggest +# the optimal cache size from a speed point of view. +# Minimum value: 0, maximum value: 9, default value: 0. + +LOOKUP_CACHE_SIZE = 0 + +# The NUM_PROC_THREADS specifies the number threads doxygen is allowed to use +# during processing. When set to 0 doxygen will based this on the number of +# cores available in the system. You can set it explicitly to a value larger +# than 0 to get more control over the balance between CPU load and processing +# speed. At this moment only the input processing can be done using multiple +# threads. Since this is still an experimental feature the default is set to 1, +# which efficively disables parallel processing. Please report any issues you +# encounter. Generating dot graphs in parallel is controlled by the +# DOT_NUM_THREADS setting. +# Minimum value: 0, maximum value: 32, default value: 1. + +NUM_PROC_THREADS = 1 + +#--------------------------------------------------------------------------- +# Build related configuration options +#--------------------------------------------------------------------------- + +# If the EXTRACT_ALL tag is set to YES, doxygen will assume all entities in +# documentation are documented, even if no documentation was available. Private +# class members and static file members will be hidden unless the +# EXTRACT_PRIVATE respectively EXTRACT_STATIC tags are set to YES. +# Note: This will also disable the warnings about undocumented members that are +# normally produced when WARNINGS is set to YES. +# The default value is: NO. + +EXTRACT_ALL = NO + +# If the EXTRACT_PRIVATE tag is set to YES, all private members of a class will +# be included in the documentation. +# The default value is: NO. + +EXTRACT_PRIVATE = NO + +# If the EXTRACT_PRIV_VIRTUAL tag is set to YES, documented private virtual +# methods of a class will be included in the documentation. +# The default value is: NO. + +EXTRACT_PRIV_VIRTUAL = NO + +# If the EXTRACT_PACKAGE tag is set to YES, all members with package or internal +# scope will be included in the documentation. +# The default value is: NO. + +EXTRACT_PACKAGE = NO + +# If the EXTRACT_STATIC tag is set to YES, all static members of a file will be +# included in the documentation. +# The default value is: NO. + +EXTRACT_STATIC = NO + +# If the EXTRACT_LOCAL_CLASSES tag is set to YES, classes (and structs) defined +# locally in source files will be included in the documentation. If set to NO, +# only classes defined in header files are included. Does not have any effect +# for Java sources. +# The default value is: YES. + +EXTRACT_LOCAL_CLASSES = YES + +# This flag is only useful for Objective-C code. If set to YES, local methods, +# which are defined in the implementation section but not in the interface are +# included in the documentation. If set to NO, only methods in the interface are +# included. +# The default value is: NO. + +EXTRACT_LOCAL_METHODS = NO + +# If this flag is set to YES, the members of anonymous namespaces will be +# extracted and appear in the documentation as a namespace called +# 'anonymous_namespace{file}', where file will be replaced with the base name of +# the file that contains the anonymous namespace. By default anonymous namespace +# are hidden. +# The default value is: NO. + +EXTRACT_ANON_NSPACES = NO + +# If the HIDE_UNDOC_MEMBERS tag is set to YES, doxygen will hide all +# undocumented members inside documented classes or files. If set to NO these +# members will be included in the various overviews, but no documentation +# section is generated. This option has no effect if EXTRACT_ALL is enabled. +# The default value is: NO. + +HIDE_UNDOC_MEMBERS = NO + +# If the HIDE_UNDOC_CLASSES tag is set to YES, doxygen will hide all +# undocumented classes that are normally visible in the class hierarchy. If set +# to NO, these classes will be included in the various overviews. This option +# has no effect if EXTRACT_ALL is enabled. +# The default value is: NO. + +HIDE_UNDOC_CLASSES = NO + +# If the HIDE_FRIEND_COMPOUNDS tag is set to YES, doxygen will hide all friend +# declarations. If set to NO, these declarations will be included in the +# documentation. +# The default value is: NO. + +HIDE_FRIEND_COMPOUNDS = NO + +# If the HIDE_IN_BODY_DOCS tag is set to YES, doxygen will hide any +# documentation blocks found inside the body of a function. If set to NO, these +# blocks will be appended to the function's detailed documentation block. +# The default value is: NO. + +HIDE_IN_BODY_DOCS = NO + +# The INTERNAL_DOCS tag determines if documentation that is typed after a +# \internal command is included. If the tag is set to NO then the documentation +# will be excluded. Set it to YES to include the internal documentation. +# The default value is: NO. + +INTERNAL_DOCS = YES + +# If the CASE_SENSE_NAMES tag is set to NO then doxygen will only generate file +# names in lower-case letters. If set to YES, upper-case letters are also +# allowed. This is useful if you have classes or files whose names only differ +# in case and if your file system supports case sensitive file names. Windows +# (including Cygwin) and Mac users are advised to set this option to NO. +# The default value is: system dependent. + +CASE_SENSE_NAMES = YES + +# If the HIDE_SCOPE_NAMES tag is set to NO then doxygen will show members with +# their full class and namespace scopes in the documentation. If set to YES, the +# scope will be hidden. +# The default value is: NO. + +HIDE_SCOPE_NAMES = NO + +# If the HIDE_COMPOUND_REFERENCE tag is set to NO (default) then doxygen will +# append additional text to a page's title, such as Class Reference. If set to +# YES the compound reference will be hidden. +# The default value is: NO. + +HIDE_COMPOUND_REFERENCE= NO + +# If the SHOW_INCLUDE_FILES tag is set to YES then doxygen will put a list of +# the files that are included by a file in the documentation of that file. +# The default value is: YES. + +SHOW_INCLUDE_FILES = YES + +# If the SHOW_GROUPED_MEMB_INC tag is set to YES then Doxygen will add for each +# grouped member an include statement to the documentation, telling the reader +# which file to include in order to use the member. +# The default value is: NO. + +SHOW_GROUPED_MEMB_INC = NO + +# If the FORCE_LOCAL_INCLUDES tag is set to YES then doxygen will list include +# files with double quotes in the documentation rather than with sharp brackets. +# The default value is: NO. + +FORCE_LOCAL_INCLUDES = NO + +# If the INLINE_INFO tag is set to YES then a tag [inline] is inserted in the +# documentation for inline members. +# The default value is: YES. + +INLINE_INFO = YES + +# If the SORT_MEMBER_DOCS tag is set to YES then doxygen will sort the +# (detailed) documentation of file and class members alphabetically by member +# name. If set to NO, the members will appear in declaration order. +# The default value is: YES. + +SORT_MEMBER_DOCS = YES + +# If the SORT_BRIEF_DOCS tag is set to YES then doxygen will sort the brief +# descriptions of file, namespace and class members alphabetically by member +# name. If set to NO, the members will appear in declaration order. Note that +# this will also influence the order of the classes in the class list. +# The default value is: NO. + +SORT_BRIEF_DOCS = NO + +# If the SORT_MEMBERS_CTORS_1ST tag is set to YES then doxygen will sort the +# (brief and detailed) documentation of class members so that constructors and +# destructors are listed first. If set to NO the constructors will appear in the +# respective orders defined by SORT_BRIEF_DOCS and SORT_MEMBER_DOCS. +# Note: If SORT_BRIEF_DOCS is set to NO this option is ignored for sorting brief +# member documentation. +# Note: If SORT_MEMBER_DOCS is set to NO this option is ignored for sorting +# detailed member documentation. +# The default value is: NO. + +SORT_MEMBERS_CTORS_1ST = NO + +# If the SORT_GROUP_NAMES tag is set to YES then doxygen will sort the hierarchy +# of group names into alphabetical order. If set to NO the group names will +# appear in their defined order. +# The default value is: NO. + +SORT_GROUP_NAMES = NO + +# If the SORT_BY_SCOPE_NAME tag is set to YES, the class list will be sorted by +# fully-qualified names, including namespaces. If set to NO, the class list will +# be sorted only by class name, not including the namespace part. +# Note: This option is not very useful if HIDE_SCOPE_NAMES is set to YES. +# Note: This option applies only to the class list, not to the alphabetical +# list. +# The default value is: NO. + +SORT_BY_SCOPE_NAME = NO + +# If the STRICT_PROTO_MATCHING option is enabled and doxygen fails to do proper +# type resolution of all parameters of a function it will reject a match between +# the prototype and the implementation of a member function even if there is +# only one candidate or it is obvious which candidate to choose by doing a +# simple string match. By disabling STRICT_PROTO_MATCHING doxygen will still +# accept a match between prototype and implementation in such cases. +# The default value is: NO. + +STRICT_PROTO_MATCHING = NO + +# The GENERATE_TODOLIST tag can be used to enable (YES) or disable (NO) the todo +# list. This list is created by putting \todo commands in the documentation. +# The default value is: YES. + +GENERATE_TODOLIST = YES + +# The GENERATE_TESTLIST tag can be used to enable (YES) or disable (NO) the test +# list. This list is created by putting \test commands in the documentation. +# The default value is: YES. + +GENERATE_TESTLIST = YES + +# The GENERATE_BUGLIST tag can be used to enable (YES) or disable (NO) the bug +# list. This list is created by putting \bug commands in the documentation. +# The default value is: YES. + +GENERATE_BUGLIST = NO + +# The GENERATE_DEPRECATEDLIST tag can be used to enable (YES) or disable (NO) +# the deprecated list. This list is created by putting \deprecated commands in +# the documentation. +# The default value is: YES. + +GENERATE_DEPRECATEDLIST= NO + +# The ENABLED_SECTIONS tag can be used to enable conditional documentation +# sections, marked by \if ... \endif and \cond +# ... \endcond blocks. + +ENABLED_SECTIONS = + +# The MAX_INITIALIZER_LINES tag determines the maximum number of lines that the +# initial value of a variable or macro / define can have for it to appear in the +# documentation. If the initializer consists of more lines than specified here +# it will be hidden. Use a value of 0 to hide initializers completely. The +# appearance of the value of individual variables and macros / defines can be +# controlled using \showinitializer or \hideinitializer command in the +# documentation regardless of this setting. +# Minimum value: 0, maximum value: 10000, default value: 30. + +MAX_INITIALIZER_LINES = 30 + +# Set the SHOW_USED_FILES tag to NO to disable the list of files generated at +# the bottom of the documentation of classes and structs. If set to YES, the +# list will mention the files that were used to generate the documentation. +# The default value is: YES. + +SHOW_USED_FILES = YES + +# Set the SHOW_FILES tag to NO to disable the generation of the Files page. This +# will remove the Files entry from the Quick Index and from the Folder Tree View +# (if specified). +# The default value is: YES. + +SHOW_FILES = YES + +# Set the SHOW_NAMESPACES tag to NO to disable the generation of the Namespaces +# page. This will remove the Namespaces entry from the Quick Index and from the +# Folder Tree View (if specified). +# The default value is: YES. + +SHOW_NAMESPACES = YES + +# The FILE_VERSION_FILTER tag can be used to specify a program or script that +# doxygen should invoke to get the current version for each file (typically from +# the version control system). Doxygen will invoke the program by executing (via +# popen()) the command command input-file, where command is the value of the +# FILE_VERSION_FILTER tag, and input-file is the name of an input file provided +# by doxygen. Whatever the program writes to standard output is used as the file +# version. For an example see the documentation. + +FILE_VERSION_FILTER = + +# The LAYOUT_FILE tag can be used to specify a layout file which will be parsed +# by doxygen. The layout file controls the global structure of the generated +# output files in an output format independent way. To create the layout file +# that represents doxygen's defaults, run doxygen with the -l option. You can +# optionally specify a file name after the option, if omitted DoxygenLayout.xml +# will be used as the name of the layout file. +# +# Note that if you run doxygen from a directory containing a file called +# DoxygenLayout.xml, doxygen will parse it automatically even if the LAYOUT_FILE +# tag is left empty. + +LAYOUT_FILE = + +# The CITE_BIB_FILES tag can be used to specify one or more bib files containing +# the reference definitions. This must be a list of .bib files. The .bib +# extension is automatically appended if omitted. This requires the bibtex tool +# to be installed. See also https://en.wikipedia.org/wiki/BibTeX for more info. +# For LaTeX the style of the bibliography can be controlled using +# LATEX_BIB_STYLE. To use this feature you need bibtex and perl available in the +# search path. See also \cite for info how to create references. + +CITE_BIB_FILES = + +#--------------------------------------------------------------------------- +# Configuration options related to warning and progress messages +#--------------------------------------------------------------------------- + +# The QUIET tag can be used to turn on/off the messages that are generated to +# standard output by doxygen. If QUIET is set to YES this implies that the +# messages are off. +# The default value is: NO. + +QUIET = NO + +# The WARNINGS tag can be used to turn on/off the warning messages that are +# generated to standard error (stderr) by doxygen. If WARNINGS is set to YES +# this implies that the warnings are on. +# +# Tip: Turn warnings on while writing the documentation. +# The default value is: YES. + +WARNINGS = YES + +# If the WARN_IF_UNDOCUMENTED tag is set to YES then doxygen will generate +# warnings for undocumented members. If EXTRACT_ALL is set to YES then this flag +# will automatically be disabled. +# The default value is: YES. + +WARN_IF_UNDOCUMENTED = YES + +# If the WARN_IF_DOC_ERROR tag is set to YES, doxygen will generate warnings for +# potential errors in the documentation, such as not documenting some parameters +# in a documented function, or documenting parameters that don't exist or using +# markup commands wrongly. +# The default value is: YES. + +WARN_IF_DOC_ERROR = YES + +# This WARN_NO_PARAMDOC option can be enabled to get warnings for functions that +# are documented, but have no documentation for their parameters or return +# value. If set to NO, doxygen will only warn about wrong or incomplete +# parameter documentation, but not about the absence of documentation. If +# EXTRACT_ALL is set to YES then this flag will automatically be disabled. +# The default value is: NO. + +WARN_NO_PARAMDOC = NO + +# If the WARN_AS_ERROR tag is set to YES then doxygen will immediately stop when +# a warning is encountered. +# The default value is: NO. + +WARN_AS_ERROR = NO + +# The WARN_FORMAT tag determines the format of the warning messages that doxygen +# can produce. The string should contain the $file, $line, and $text tags, which +# will be replaced by the file and line number from which the warning originated +# and the warning text. Optionally the format may contain $version, which will +# be replaced by the version of the file (if it could be obtained via +# FILE_VERSION_FILTER) +# The default value is: $file:$line: $text. + +WARN_FORMAT = "$file:$line: $text" + +# The WARN_LOGFILE tag can be used to specify a file to which warning and error +# messages should be written. If left blank the output is written to standard +# error (stderr). + +WARN_LOGFILE = + +#--------------------------------------------------------------------------- +# Configuration options related to the input files +#--------------------------------------------------------------------------- + +# The INPUT tag is used to specify the files and/or directories that contain +# documented source files. You may enter file names like myfile.cpp or +# directories like /usr/src/myproject. Separate the files or directories with +# spaces. See also FILE_PATTERNS and EXTENSION_MAPPING +# Note: If this tag is empty the current directory is searched. + +INPUT = mainpage.txt \ + ../../README.md \ + ../../CONTRIBUTING.md \ + sync.txt \ + ../../INSTALL.md \ + ../../docs/markdown \ + ../../include/hip \ + ../../../hipamd/include/hip/amd_detail/ + +# This tag can be used to specify the character encoding of the source files +# that doxygen parses. Internally doxygen uses the UTF-8 encoding. Doxygen uses +# libiconv (or the iconv built into libc) for the transcoding. See the libiconv +# documentation (see: https://www.gnu.org/software/libiconv/) for the list of +# possible encodings. +# The default value is: UTF-8. + +INPUT_ENCODING = UTF-8 + +# If the value of the INPUT tag contains directories, you can use the +# FILE_PATTERNS tag to specify one or more wildcard patterns (like *.cpp and +# *.h) to filter out the source-files in the directories. +# +# Note that for custom extensions or not directly supported extensions you also +# need to set EXTENSION_MAPPING for the extension otherwise the files are not +# read by doxygen. +# +# If left blank the following patterns are tested:*.c, *.cc, *.cxx, *.cpp, +# *.c++, *.java, *.ii, *.ixx, *.ipp, *.i++, *.inl, *.idl, *.ddl, *.odl, *.h, +# *.hh, *.hxx, *.hpp, *.h++, *.cs, *.d, *.php, *.php4, *.php5, *.phtml, *.inc, +# *.m, *.markdown, *.md, *.mm, *.dox (to be provided as doxygen C comment), +# *.doc (to be provided as doxygen C comment), *.txt (to be provided as doxygen +# C comment), *.py, *.pyw, *.f90, *.f95, *.f03, *.f08, *.f18, *.f, *.for, *.vhd, +# *.vhdl, *.ucf, *.qsf and *.ice. + +FILE_PATTERNS = *.c \ + *.cc \ + *.cxx \ + *.cpp \ + *.h \ + *.hpp \ + *.md \ + *.dox \ + *.doc \ + *.txt + +# The RECURSIVE tag can be used to specify whether or not subdirectories should +# be searched for input files as well. +# The default value is: NO. + +RECURSIVE = YES + +# The EXCLUDE tag can be used to specify files and/or directories that should be +# excluded from the INPUT source files. This way you can easily exclude a +# subdirectory from a directory tree whose root is specified with the INPUT tag. +# +# Note that relative paths are relative to the directory from which doxygen is +# run. + +EXCLUDE = + +# The EXCLUDE_SYMLINKS tag can be used to select whether or not files or +# directories that are symbolic links (a Unix file system feature) are excluded +# from the input. +# The default value is: NO. + +EXCLUDE_SYMLINKS = NO + +# If the value of the INPUT tag contains directories, you can use the +# EXCLUDE_PATTERNS tag to specify one or more wildcard patterns to exclude +# certain files from those directories. +# +# Note that the wildcards are matched against the file with absolute path, so to +# exclude all test directories for example use the pattern */test/* + +EXCLUDE_PATTERNS = + +# The EXCLUDE_SYMBOLS tag can be used to specify one or more symbol names +# (namespaces, classes, functions, etc.) that should be excluded from the +# output. The symbol name can be a fully qualified name, a word, or if the +# wildcard * is used, a substring. Examples: ANamespace, AClass, +# AClass::ANamespace, ANamespace::*Test +# +# Note that the wildcards are matched against the file with absolute path, so to +# exclude all test directories use the pattern */test/* + +EXCLUDE_SYMBOLS = + +# The EXAMPLE_PATH tag can be used to specify one or more files or directories +# that contain example code fragments that are included (see the \include +# command). + +EXAMPLE_PATH = ./examples + +# If the value of the EXAMPLE_PATH tag contains directories, you can use the +# EXAMPLE_PATTERNS tag to specify one or more wildcard pattern (like *.cpp and +# *.h) to filter out the source-files in the directories. If left blank all +# files are included. + +EXAMPLE_PATTERNS = + +# If the EXAMPLE_RECURSIVE tag is set to YES then subdirectories will be +# searched for input files to be used with the \include or \dontinclude commands +# irrespective of the value of the RECURSIVE tag. +# The default value is: NO. + +EXAMPLE_RECURSIVE = NO + +# The IMAGE_PATH tag can be used to specify one or more files or directories +# that contain images that are to be included in the documentation (see the +# \image command). + +IMAGE_PATH = images + +# The INPUT_FILTER tag can be used to specify a program that doxygen should +# invoke to filter for each input file. Doxygen will invoke the filter program +# by executing (via popen()) the command: +# +# +# +# where is the value of the INPUT_FILTER tag, and is the +# name of an input file. Doxygen will then use the output that the filter +# program writes to standard output. If FILTER_PATTERNS is specified, this tag +# will be ignored. +# +# Note that the filter must not add or remove lines; it is applied before the +# code is scanned, but not when the output code is generated. If lines are added +# or removed, the anchors will not be placed correctly. +# +# Note that for custom extensions or not directly supported extensions you also +# need to set EXTENSION_MAPPING for the extension otherwise the files are not +# properly processed by doxygen. + +INPUT_FILTER = + +# The FILTER_PATTERNS tag can be used to specify filters on a per file pattern +# basis. Doxygen will compare the file name with each pattern and apply the +# filter if there is a match. The filters are a list of the form: pattern=filter +# (like *.cpp=my_cpp_filter). See INPUT_FILTER for further information on how +# filters are used. If the FILTER_PATTERNS tag is empty or if none of the +# patterns match the file name, INPUT_FILTER is applied. +# +# Note that for custom extensions or not directly supported extensions you also +# need to set EXTENSION_MAPPING for the extension otherwise the files are not +# properly processed by doxygen. + +FILTER_PATTERNS = + +# If the FILTER_SOURCE_FILES tag is set to YES, the input filter (if set using +# INPUT_FILTER) will also be used to filter the input files that are used for +# producing the source files to browse (i.e. when SOURCE_BROWSER is set to YES). +# The default value is: NO. + +FILTER_SOURCE_FILES = NO + +# The FILTER_SOURCE_PATTERNS tag can be used to specify source filters per file +# pattern. A pattern will override the setting for FILTER_PATTERN (if any) and +# it is also possible to disable source filtering for a specific pattern using +# *.ext= (so without naming a filter). +# This tag requires that the tag FILTER_SOURCE_FILES is set to YES. + +FILTER_SOURCE_PATTERNS = + +# If the USE_MDFILE_AS_MAINPAGE tag refers to the name of a markdown file that +# is part of the input, its contents will be placed on the main page +# (index.html). This can be useful if you have a project on for instance GitHub +# and want to reuse the introduction page also for the doxygen output. + +USE_MDFILE_AS_MAINPAGE = + +#--------------------------------------------------------------------------- +# Configuration options related to source browsing +#--------------------------------------------------------------------------- + +# If the SOURCE_BROWSER tag is set to YES then a list of source files will be +# generated. Documented entities will be cross-referenced with these sources. +# +# Note: To get rid of all source code in the generated output, make sure that +# also VERBATIM_HEADERS is set to NO. +# The default value is: NO. + +SOURCE_BROWSER = NO + +# Setting the INLINE_SOURCES tag to YES will include the body of functions, +# classes and enums directly into the documentation. +# The default value is: NO. + +INLINE_SOURCES = NO + +# Setting the STRIP_CODE_COMMENTS tag to YES will instruct doxygen to hide any +# special comment blocks from generated source code fragments. Normal C, C++ and +# Fortran comments will always remain visible. +# The default value is: YES. + +STRIP_CODE_COMMENTS = YES + +# If the REFERENCED_BY_RELATION tag is set to YES then for each documented +# entity all documented functions referencing it will be listed. +# The default value is: NO. + +REFERENCED_BY_RELATION = NO + +# If the REFERENCES_RELATION tag is set to YES then for each documented function +# all documented entities called/used by that function will be listed. +# The default value is: NO. + +REFERENCES_RELATION = NO + +# If the REFERENCES_LINK_SOURCE tag is set to YES and SOURCE_BROWSER tag is set +# to YES then the hyperlinks from functions in REFERENCES_RELATION and +# REFERENCED_BY_RELATION lists will link to the source code. Otherwise they will +# link to the documentation. +# The default value is: YES. + +REFERENCES_LINK_SOURCE = YES + +# If SOURCE_TOOLTIPS is enabled (the default) then hovering a hyperlink in the +# source code will show a tooltip with additional information such as prototype, +# brief description and links to the definition and documentation. Since this +# will make the HTML file larger and loading of large files a bit slower, you +# can opt to disable this feature. +# The default value is: YES. +# This tag requires that the tag SOURCE_BROWSER is set to YES. + +SOURCE_TOOLTIPS = YES + +# If the USE_HTAGS tag is set to YES then the references to source code will +# point to the HTML generated by the htags(1) tool instead of doxygen built-in +# source browser. The htags tool is part of GNU's global source tagging system +# (see https://www.gnu.org/software/global/global.html). You will need version +# 4.8.6 or higher. +# +# To use it do the following: +# - Install the latest version of global +# - Enable SOURCE_BROWSER and USE_HTAGS in the configuration file +# - Make sure the INPUT points to the root of the source tree +# - Run doxygen as normal +# +# Doxygen will invoke htags (and that will in turn invoke gtags), so these +# tools must be available from the command line (i.e. in the search path). +# +# The result: instead of the source browser generated by doxygen, the links to +# source code will now point to the output of htags. +# The default value is: NO. +# This tag requires that the tag SOURCE_BROWSER is set to YES. + +USE_HTAGS = NO + +# If the VERBATIM_HEADERS tag is set the YES then doxygen will generate a +# verbatim copy of the header file for each class for which an include is +# specified. Set to NO to disable this. +# See also: Section \class. +# The default value is: YES. + +VERBATIM_HEADERS = YES + +# If the CLANG_ASSISTED_PARSING tag is set to YES then doxygen will use the +# clang parser (see: http://clang.llvm.org/) for more accurate parsing at the +# cost of reduced performance. This can be particularly helpful with template +# rich C++ code for which doxygen's built-in parser lacks the necessary type +# information. +# Note: The availability of this option depends on whether or not doxygen was +# generated with the -Duse_libclang=ON option for CMake. +# The default value is: NO. + +CLANG_ASSISTED_PARSING = NO + +# If clang assisted parsing is enabled you can provide the compiler with command +# line options that you would normally use when invoking the compiler. Note that +# the include paths will already be set by doxygen for the files and directories +# specified with INPUT and INCLUDE_PATH. +# This tag requires that the tag CLANG_ASSISTED_PARSING is set to YES. + +CLANG_OPTIONS = + +# If clang assisted parsing is enabled you can provide the clang parser with the +# path to the directory containing a file called compile_commands.json. This +# file is the compilation database (see: +# http://clang.llvm.org/docs/HowToSetupToolingForLLVM.html) containing the +# options used when the source files were built. This is equivalent to +# specifying the "-p" option to a clang tool, such as clang-check. These options +# will then be passed to the parser. Any options specified with CLANG_OPTIONS +# will be added as well. +# Note: The availability of this option depends on whether or not doxygen was +# generated with the -Duse_libclang=ON option for CMake. + +CLANG_DATABASE_PATH = + +#--------------------------------------------------------------------------- +# Configuration options related to the alphabetical class index +#--------------------------------------------------------------------------- + +# If the ALPHABETICAL_INDEX tag is set to YES, an alphabetical index of all +# compounds will be generated. Enable this if the project contains a lot of +# classes, structs, unions or interfaces. +# The default value is: YES. + +ALPHABETICAL_INDEX = YES + +# The COLS_IN_ALPHA_INDEX tag can be used to specify the number of columns in +# which the alphabetical index list will be split. +# Minimum value: 1, maximum value: 20, default value: 5. +# This tag requires that the tag ALPHABETICAL_INDEX is set to YES. + +COLS_IN_ALPHA_INDEX = 5 + +# In case all classes in a project start with a common prefix, all classes will +# be put under the same header in the alphabetical index. The IGNORE_PREFIX tag +# can be used to specify a prefix (or a list of prefixes) that should be ignored +# while generating the index headers. +# This tag requires that the tag ALPHABETICAL_INDEX is set to YES. + +IGNORE_PREFIX = + +#--------------------------------------------------------------------------- +# Configuration options related to the HTML output +#--------------------------------------------------------------------------- + +# If the GENERATE_HTML tag is set to YES, doxygen will generate HTML output +# The default value is: YES. + +GENERATE_HTML = YES + +# The HTML_OUTPUT tag is used to specify where the HTML docs will be put. If a +# relative path is entered the value of OUTPUT_DIRECTORY will be put in front of +# it. +# The default directory is: html. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_OUTPUT = html + +# The HTML_FILE_EXTENSION tag can be used to specify the file extension for each +# generated HTML page (for example: .htm, .php, .asp). +# The default value is: .html. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_FILE_EXTENSION = .html + +# The HTML_HEADER tag can be used to specify a user-defined HTML header file for +# each generated HTML page. If the tag is left blank doxygen will generate a +# standard header. +# +# To get valid HTML the header file that includes any scripts and style sheets +# that doxygen needs, which is dependent on the configuration options used (e.g. +# the setting GENERATE_TREEVIEW). It is highly recommended to start with a +# default header using +# doxygen -w html new_header.html new_footer.html new_stylesheet.css +# YourConfigFile +# and then modify the file new_header.html. See also section "Doxygen usage" +# for information on how to generate the default header that doxygen normally +# uses. +# Note: The header is subject to change so you typically have to regenerate the +# default header when upgrading to a newer version of doxygen. For a description +# of the possible markers and block names see the documentation. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_HEADER = + +# The HTML_FOOTER tag can be used to specify a user-defined HTML footer for each +# generated HTML page. If the tag is left blank doxygen will generate a standard +# footer. See HTML_HEADER for more information on how to generate a default +# footer and what special commands can be used inside the footer. See also +# section "Doxygen usage" for information on how to generate the default footer +# that doxygen normally uses. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_FOOTER = + +# The HTML_STYLESHEET tag can be used to specify a user-defined cascading style +# sheet that is used by each HTML page. It can be used to fine-tune the look of +# the HTML output. If left blank doxygen will generate a default style sheet. +# See also section "Doxygen usage" for information on how to generate the style +# sheet that doxygen normally uses. +# Note: It is recommended to use HTML_EXTRA_STYLESHEET instead of this tag, as +# it is more robust and this tag (HTML_STYLESHEET) will in the future become +# obsolete. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_STYLESHEET = + +# The HTML_EXTRA_STYLESHEET tag can be used to specify additional user-defined +# cascading style sheets that are included after the standard style sheets +# created by doxygen. Using this option one can overrule certain style aspects. +# This is preferred over using HTML_STYLESHEET since it does not replace the +# standard style sheet and is therefore more robust against future updates. +# Doxygen will copy the style sheet files to the output directory. +# Note: The order of the extra style sheet files is of importance (e.g. the last +# style sheet in the list overrules the setting of the previous ones in the +# list). For an example see the documentation. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_EXTRA_STYLESHEET = + +# The HTML_EXTRA_FILES tag can be used to specify one or more extra images or +# other source files which should be copied to the HTML output directory. Note +# that these files will be copied to the base HTML output directory. Use the +# $relpath^ marker in the HTML_HEADER and/or HTML_FOOTER files to load these +# files. In the HTML_STYLESHEET file, use the file name only. Also note that the +# files will be copied as-is; there are no commands or markers available. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_EXTRA_FILES = + +# The HTML_COLORSTYLE_HUE tag controls the color of the HTML output. Doxygen +# will adjust the colors in the style sheet and background images according to +# this color. Hue is specified as an angle on a colorwheel, see +# https://en.wikipedia.org/wiki/Hue for more information. For instance the value +# 0 represents red, 60 is yellow, 120 is green, 180 is cyan, 240 is blue, 300 +# purple, and 360 is red again. +# Minimum value: 0, maximum value: 359, default value: 220. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_COLORSTYLE_HUE = 220 + +# The HTML_COLORSTYLE_SAT tag controls the purity (or saturation) of the colors +# in the HTML output. For a value of 0 the output will use grayscales only. A +# value of 255 will produce the most vivid colors. +# Minimum value: 0, maximum value: 255, default value: 100. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_COLORSTYLE_SAT = 100 + +# The HTML_COLORSTYLE_GAMMA tag controls the gamma correction applied to the +# luminance component of the colors in the HTML output. Values below 100 +# gradually make the output lighter, whereas values above 100 make the output +# darker. The value divided by 100 is the actual gamma applied, so 80 represents +# a gamma of 0.8, The value 220 represents a gamma of 2.2, and 100 does not +# change the gamma. +# Minimum value: 40, maximum value: 240, default value: 80. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_COLORSTYLE_GAMMA = 80 + +# If the HTML_TIMESTAMP tag is set to YES then the footer of each generated HTML +# page will contain the date and time when the page was generated. Setting this +# to YES can help to show when doxygen was last run and thus if the +# documentation is up to date. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_TIMESTAMP = YES + +# If the HTML_DYNAMIC_MENUS tag is set to YES then the generated HTML +# documentation will contain a main index with vertical navigation menus that +# are dynamically created via JavaScript. If disabled, the navigation index will +# consists of multiple levels of tabs that are statically embedded in every HTML +# page. Disable this option to support browsers that do not have JavaScript, +# like the Qt help browser. +# The default value is: YES. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_DYNAMIC_MENUS = YES + +# If the HTML_DYNAMIC_SECTIONS tag is set to YES then the generated HTML +# documentation will contain sections that can be hidden and shown after the +# page has loaded. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_DYNAMIC_SECTIONS = NO + +# With HTML_INDEX_NUM_ENTRIES one can control the preferred number of entries +# shown in the various tree structured indices initially; the user can expand +# and collapse entries dynamically later on. Doxygen will expand the tree to +# such a level that at most the specified number of entries are visible (unless +# a fully collapsed tree already exceeds this amount). So setting the number of +# entries 1 will produce a full collapsed tree by default. 0 is a special value +# representing an infinite number of entries and will result in a full expanded +# tree by default. +# Minimum value: 0, maximum value: 9999, default value: 100. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_INDEX_NUM_ENTRIES = 100 + +# If the GENERATE_DOCSET tag is set to YES, additional index files will be +# generated that can be used as input for Apple's Xcode 3 integrated development +# environment (see: https://developer.apple.com/xcode/), introduced with OSX +# 10.5 (Leopard). To create a documentation set, doxygen will generate a +# Makefile in the HTML output directory. Running make will produce the docset in +# that directory and running make install will install the docset in +# ~/Library/Developer/Shared/Documentation/DocSets so that Xcode will find it at +# startup. See https://developer.apple.com/library/archive/featuredarticles/Doxy +# genXcode/_index.html for more information. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +GENERATE_DOCSET = NO + +# This tag determines the name of the docset feed. A documentation feed provides +# an umbrella under which multiple documentation sets from a single provider +# (such as a company or product suite) can be grouped. +# The default value is: Doxygen generated docs. +# This tag requires that the tag GENERATE_DOCSET is set to YES. + +DOCSET_FEEDNAME = "Doxygen generated docs" + +# This tag specifies a string that should uniquely identify the documentation +# set bundle. This should be a reverse domain-name style string, e.g. +# com.mycompany.MyDocSet. Doxygen will append .docset to the name. +# The default value is: org.doxygen.Project. +# This tag requires that the tag GENERATE_DOCSET is set to YES. + +DOCSET_BUNDLE_ID = org.doxygen.Project + +# The DOCSET_PUBLISHER_ID tag specifies a string that should uniquely identify +# the documentation publisher. This should be a reverse domain-name style +# string, e.g. com.mycompany.MyDocSet.documentation. +# The default value is: org.doxygen.Publisher. +# This tag requires that the tag GENERATE_DOCSET is set to YES. + +DOCSET_PUBLISHER_ID = org.doxygen.Publisher + +# The DOCSET_PUBLISHER_NAME tag identifies the documentation publisher. +# The default value is: Publisher. +# This tag requires that the tag GENERATE_DOCSET is set to YES. + +DOCSET_PUBLISHER_NAME = Publisher + +# If the GENERATE_HTMLHELP tag is set to YES then doxygen generates three +# additional HTML index files: index.hhp, index.hhc, and index.hhk. The +# index.hhp is a project file that can be read by Microsoft's HTML Help Workshop +# (see: https://www.microsoft.com/en-us/download/details.aspx?id=21138) on +# Windows. +# +# The HTML Help Workshop contains a compiler that can convert all HTML output +# generated by doxygen into a single compiled HTML file (.chm). Compiled HTML +# files are now used as the Windows 98 help format, and will replace the old +# Windows help format (.hlp) on all Windows platforms in the future. Compressed +# HTML files also contain an index, a table of contents, and you can search for +# words in the documentation. The HTML workshop also contains a viewer for +# compressed HTML files. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +GENERATE_HTMLHELP = NO + +# The CHM_FILE tag can be used to specify the file name of the resulting .chm +# file. You can add a path in front of the file if the result should not be +# written to the html output directory. +# This tag requires that the tag GENERATE_HTMLHELP is set to YES. + +CHM_FILE = + +# The HHC_LOCATION tag can be used to specify the location (absolute path +# including file name) of the HTML help compiler (hhc.exe). If non-empty, +# doxygen will try to run the HTML help compiler on the generated index.hhp. +# The file has to be specified with full path. +# This tag requires that the tag GENERATE_HTMLHELP is set to YES. + +HHC_LOCATION = + +# The GENERATE_CHI flag controls if a separate .chi index file is generated +# (YES) or that it should be included in the main .chm file (NO). +# The default value is: NO. +# This tag requires that the tag GENERATE_HTMLHELP is set to YES. + +GENERATE_CHI = NO + +# The CHM_INDEX_ENCODING is used to encode HtmlHelp index (hhk), content (hhc) +# and project file content. +# This tag requires that the tag GENERATE_HTMLHELP is set to YES. + +CHM_INDEX_ENCODING = + +# The BINARY_TOC flag controls whether a binary table of contents is generated +# (YES) or a normal table of contents (NO) in the .chm file. Furthermore it +# enables the Previous and Next buttons. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTMLHELP is set to YES. + +BINARY_TOC = NO + +# The TOC_EXPAND flag can be set to YES to add extra items for group members to +# the table of contents of the HTML help documentation and to the tree view. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTMLHELP is set to YES. + +TOC_EXPAND = NO + +# If the GENERATE_QHP tag is set to YES and both QHP_NAMESPACE and +# QHP_VIRTUAL_FOLDER are set, an additional index file will be generated that +# can be used as input for Qt's qhelpgenerator to generate a Qt Compressed Help +# (.qch) of the generated HTML documentation. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +GENERATE_QHP = NO + +# If the QHG_LOCATION tag is specified, the QCH_FILE tag can be used to specify +# the file name of the resulting .qch file. The path specified is relative to +# the HTML output folder. +# This tag requires that the tag GENERATE_QHP is set to YES. + +QCH_FILE = + +# The QHP_NAMESPACE tag specifies the namespace to use when generating Qt Help +# Project output. For more information please see Qt Help Project / Namespace +# (see: https://doc.qt.io/archives/qt-4.8/qthelpproject.html#namespace). +# The default value is: org.doxygen.Project. +# This tag requires that the tag GENERATE_QHP is set to YES. + +QHP_NAMESPACE = org.doxygen.Project + +# The QHP_VIRTUAL_FOLDER tag specifies the namespace to use when generating Qt +# Help Project output. For more information please see Qt Help Project / Virtual +# Folders (see: https://doc.qt.io/archives/qt-4.8/qthelpproject.html#virtual- +# folders). +# The default value is: doc. +# This tag requires that the tag GENERATE_QHP is set to YES. + +QHP_VIRTUAL_FOLDER = doc + +# If the QHP_CUST_FILTER_NAME tag is set, it specifies the name of a custom +# filter to add. For more information please see Qt Help Project / Custom +# Filters (see: https://doc.qt.io/archives/qt-4.8/qthelpproject.html#custom- +# filters). +# This tag requires that the tag GENERATE_QHP is set to YES. + +QHP_CUST_FILTER_NAME = + +# The QHP_CUST_FILTER_ATTRS tag specifies the list of the attributes of the +# custom filter to add. For more information please see Qt Help Project / Custom +# Filters (see: https://doc.qt.io/archives/qt-4.8/qthelpproject.html#custom- +# filters). +# This tag requires that the tag GENERATE_QHP is set to YES. + +QHP_CUST_FILTER_ATTRS = + +# The QHP_SECT_FILTER_ATTRS tag specifies the list of the attributes this +# project's filter section matches. Qt Help Project / Filter Attributes (see: +# https://doc.qt.io/archives/qt-4.8/qthelpproject.html#filter-attributes). +# This tag requires that the tag GENERATE_QHP is set to YES. + +QHP_SECT_FILTER_ATTRS = + +# The QHG_LOCATION tag can be used to specify the location of Qt's +# qhelpgenerator. If non-empty doxygen will try to run qhelpgenerator on the +# generated .qhp file. +# This tag requires that the tag GENERATE_QHP is set to YES. + +QHG_LOCATION = + +# If the GENERATE_ECLIPSEHELP tag is set to YES, additional index files will be +# generated, together with the HTML files, they form an Eclipse help plugin. To +# install this plugin and make it available under the help contents menu in +# Eclipse, the contents of the directory containing the HTML and XML files needs +# to be copied into the plugins directory of eclipse. The name of the directory +# within the plugins directory should be the same as the ECLIPSE_DOC_ID value. +# After copying Eclipse needs to be restarted before the help appears. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +GENERATE_ECLIPSEHELP = NO + +# A unique identifier for the Eclipse help plugin. When installing the plugin +# the directory name containing the HTML and XML files should also have this +# name. Each documentation set should have its own identifier. +# The default value is: org.doxygen.Project. +# This tag requires that the tag GENERATE_ECLIPSEHELP is set to YES. + +ECLIPSE_DOC_ID = org.doxygen.Project + +# If you want full control over the layout of the generated HTML pages it might +# be necessary to disable the index and replace it with your own. The +# DISABLE_INDEX tag can be used to turn on/off the condensed index (tabs) at top +# of each HTML page. A value of NO enables the index and the value YES disables +# it. Since the tabs in the index contain the same information as the navigation +# tree, you can set this option to YES if you also set GENERATE_TREEVIEW to YES. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +DISABLE_INDEX = NO + +# The GENERATE_TREEVIEW tag is used to specify whether a tree-like index +# structure should be generated to display hierarchical information. If the tag +# value is set to YES, a side panel will be generated containing a tree-like +# index structure (just like the one that is generated for HTML Help). For this +# to work a browser that supports JavaScript, DHTML, CSS and frames is required +# (i.e. any modern browser). Windows users are probably better off using the +# HTML help feature. Via custom style sheets (see HTML_EXTRA_STYLESHEET) one can +# further fine-tune the look of the index. As an example, the default style +# sheet generated by doxygen has an example that shows how to put an image at +# the root of the tree instead of the PROJECT_NAME. Since the tree basically has +# the same information as the tab index, you could consider setting +# DISABLE_INDEX to YES when enabling this option. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +GENERATE_TREEVIEW = NO + +# The ENUM_VALUES_PER_LINE tag can be used to set the number of enum values that +# doxygen will group on one line in the generated HTML documentation. +# +# Note that a value of 0 will completely suppress the enum values from appearing +# in the overview section. +# Minimum value: 0, maximum value: 20, default value: 4. +# This tag requires that the tag GENERATE_HTML is set to YES. + +ENUM_VALUES_PER_LINE = 4 + +# If the treeview is enabled (see GENERATE_TREEVIEW) then this tag can be used +# to set the initial width (in pixels) of the frame in which the tree is shown. +# Minimum value: 0, maximum value: 1500, default value: 250. +# This tag requires that the tag GENERATE_HTML is set to YES. + +TREEVIEW_WIDTH = 250 + +# If the EXT_LINKS_IN_WINDOW option is set to YES, doxygen will open links to +# external symbols imported via tag files in a separate window. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +EXT_LINKS_IN_WINDOW = NO + +# If the HTML_FORMULA_FORMAT option is set to svg, doxygen will use the pdf2svg +# tool (see https://github.com/dawbarton/pdf2svg) or inkscape (see +# https://inkscape.org) to generate formulas as SVG images instead of PNGs for +# the HTML output. These images will generally look nicer at scaled resolutions. +# Possible values are: png (the default) and svg (looks nicer but requires the +# pdf2svg or inkscape tool). +# The default value is: png. +# This tag requires that the tag GENERATE_HTML is set to YES. + +HTML_FORMULA_FORMAT = png + +# Use this tag to change the font size of LaTeX formulas included as images in +# the HTML documentation. When you change the font size after a successful +# doxygen run you need to manually remove any form_*.png images from the HTML +# output directory to force them to be regenerated. +# Minimum value: 8, maximum value: 50, default value: 10. +# This tag requires that the tag GENERATE_HTML is set to YES. + +FORMULA_FONTSIZE = 10 + +# Use the FORMULA_TRANSPARENT tag to determine whether or not the images +# generated for formulas are transparent PNGs. Transparent PNGs are not +# supported properly for IE 6.0, but are supported on all modern browsers. +# +# Note that when changing this option you need to delete any form_*.png files in +# the HTML output directory before the changes have effect. +# The default value is: YES. +# This tag requires that the tag GENERATE_HTML is set to YES. + +FORMULA_TRANSPARENT = YES + +# The FORMULA_MACROFILE can contain LaTeX \newcommand and \renewcommand commands +# to create new LaTeX commands to be used in formulas as building blocks. See +# the section "Including formulas" for details. + +FORMULA_MACROFILE = + +# Enable the USE_MATHJAX option to render LaTeX formulas using MathJax (see +# https://www.mathjax.org) which uses client side JavaScript for the rendering +# instead of using pre-rendered bitmaps. Use this if you do not have LaTeX +# installed or if you want to formulas look prettier in the HTML output. When +# enabled you may also need to install MathJax separately and configure the path +# to it using the MATHJAX_RELPATH option. +# The default value is: NO. +# This tag requires that the tag GENERATE_HTML is set to YES. + +USE_MATHJAX = NO + +# When MathJax is enabled you can set the default output format to be used for +# the MathJax output. See the MathJax site (see: +# http://docs.mathjax.org/en/latest/output.html) for more details. +# Possible values are: HTML-CSS (which is slower, but has the best +# compatibility), NativeMML (i.e. MathML) and SVG. +# The default value is: HTML-CSS. +# This tag requires that the tag USE_MATHJAX is set to YES. + +MATHJAX_FORMAT = HTML-CSS + +# When MathJax is enabled you need to specify the location relative to the HTML +# output directory using the MATHJAX_RELPATH option. The destination directory +# should contain the MathJax.js script. For instance, if the mathjax directory +# is located at the same level as the HTML output directory, then +# MATHJAX_RELPATH should be ../mathjax. The default value points to the MathJax +# Content Delivery Network so you can quickly see the result without installing +# MathJax. However, it is strongly recommended to install a local copy of +# MathJax from https://www.mathjax.org before deployment. +# The default value is: https://cdn.jsdelivr.net/npm/mathjax@2. +# This tag requires that the tag USE_MATHJAX is set to YES. + +MATHJAX_RELPATH = https://cdnjs.cloudflare.com/ajax/libs/mathjax/2.7.5/ + +# The MATHJAX_EXTENSIONS tag can be used to specify one or more MathJax +# extension names that should be enabled during MathJax rendering. For example +# MATHJAX_EXTENSIONS = TeX/AMSmath TeX/AMSsymbols +# This tag requires that the tag USE_MATHJAX is set to YES. + +MATHJAX_EXTENSIONS = + +# The MATHJAX_CODEFILE tag can be used to specify a file with javascript pieces +# of code that will be used on startup of the MathJax code. See the MathJax site +# (see: http://docs.mathjax.org/en/latest/output.html) for more details. For an +# example see the documentation. +# This tag requires that the tag USE_MATHJAX is set to YES. + +MATHJAX_CODEFILE = + +# When the SEARCHENGINE tag is enabled doxygen will generate a search box for +# the HTML output. The underlying search engine uses javascript and DHTML and +# should work on any modern browser. Note that when using HTML help +# (GENERATE_HTMLHELP), Qt help (GENERATE_QHP), or docsets (GENERATE_DOCSET) +# there is already a search function so this one should typically be disabled. +# For large projects the javascript based search engine can be slow, then +# enabling SERVER_BASED_SEARCH may provide a better solution. It is possible to +# search using the keyboard; to jump to the search box use + S +# (what the is depends on the OS and browser, but it is typically +# , /