Skip to content
Draft
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
3 changes: 3 additions & 0 deletions .github/scripts/therock_configure_ci.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

import fnmatch
import json
import os
Expand Down
3 changes: 3 additions & 0 deletions codegen/test/batched_gemm_softmax_gemm.cpp
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include "ck/host/device_batched_gemm_softmax_gemm/problem.hpp"
#include "ck/host/stringutils.hpp"
#include "ck/host/utils.hpp"
Expand Down
3 changes: 3 additions & 0 deletions codegen/test/include/test.hpp
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

/*
* The MIT License (MIT)
*
Expand Down
3 changes: 3 additions & 0 deletions docs/conf.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# Configuration file for the Sphinx documentation builder.
#
# This file only contains a selection of the most common options. For a full
Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#pragma once

bool run_gemm_add_add_fastgelu(const ProblemSize& problem_size, const ExecutionConfig& config)
Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include <iostream>
#include <numeric>
#include <initializer_list>
Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include <iostream>
#include <numeric>
#include <initializer_list>
Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include <iostream>
#include <numeric>
#include <initializer_list>
Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include <iostream>
#include <numeric>
#include <initializer_list>
Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#pragma once
#include <type_traits>

Expand Down
3 changes: 3 additions & 0 deletions example/68_gemm_add/run_gemm_add_example_wmma.inc
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#pragma once

bool run_gemm_add(const ProblemSize& problem_size, const ExecutionConfig& config)
Expand Down
3 changes: 3 additions & 0 deletions example/68_gemm_add/run_gemm_add_example_xdl.inc
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#pragma once

bool run_gemm_add(const ProblemSize& problem_size, const ExecutionConfig& config)
Expand Down
3 changes: 3 additions & 0 deletions example/69_gemm_add_relu/run_gemm_add_relu_example_wmma.inc
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#pragma once

bool run_gemm_add_relu(const ProblemSize& problem_size, const ExecutionConfig& config)
Expand Down
3 changes: 3 additions & 0 deletions example/69_gemm_add_relu/run_gemm_add_relu_example_xdl.inc
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#pragma once

bool run_gemm_add_relu(const ProblemSize& problem_size, const ExecutionConfig& config)
Expand Down
2 changes: 2 additions & 0 deletions example/ck_tile/01_fmha/codegen/__init__.py
Original file line number Diff line number Diff line change
@@ -0,0 +1,2 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/codegen/arch.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# SPDX-License-Identifier: MIT
# Copyright (c) 2025, Advanced Micro Devices, Inc. All rights reserved.

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/codegen/cmake_config.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# SPDX-License-Identifier: MIT
# Copyright (c) 2018-2024, Advanced Micro Devices, Inc. All rights reserved.
# generate kernel instances to speed up compilation
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/codegen/cpp_symbol_map.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# SPDX-License-Identifier: MIT
# Copyright (c) 2018-2025, Advanced Micro Devices, Inc. All rights reserved.
# generate kernel instances to speed up compilation
Expand Down
2 changes: 2 additions & 0 deletions example/ck_tile/01_fmha/codegen/ops/__init__.py
Original file line number Diff line number Diff line change
@@ -0,0 +1,2 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/codegen/ops/fmha_fwd_appendkv.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# SPDX-License-Identifier: MIT
# Copyright (c) 2018-2025, Advanced Micro Devices, Inc. All rights reserved.
# generate kernel instances to speed up compilation
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/codegen/ops/fmha_fwd_splitkv.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# SPDX-License-Identifier: MIT
# Copyright (c) 2018-2025, Advanced Micro Devices, Inc. All rights reserved.
# generate kernel instances to speed up compilation
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/codegen/ops/fmha_pagedkv_prefill.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# SPDX-License-Identifier: MIT
# Copyright (c) 2018-2025, Advanced Micro Devices, Inc. All rights reserved.
# generate kernel instances to speed up compilation
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/codegen/utils.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# SPDX-License-Identifier: MIT
# Copyright (c) 2025, Advanced Micro Devices, Inc. All rights reserved.
# generate kernel instances to speed up compilation
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/generate.py
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# SPDX-License-Identifier: MIT
# Copyright (c) 2018-2025, Advanced Micro Devices, Inc. All rights reserved.

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/script/benchmark_bwd.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# TODO: run this script from CK root or build directory
EXE="$(find . -name tile_example_fmha_bwd -type f | head -n 1)"
VALID=0
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/script/benchmark_fwd.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# TODO: run this script from CK root or build directory
EXE="$(find . -name tile_example_fmha_fwd -type f | head -n 1)"
VALID=0
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/script/benchmark_fwd_v3.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# TODO: run this script from CK root or build directory
EXE="$(find . -name tile_example_fmha_fwd_v3 -type f | head -n 1)"
VALID=0
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/script/run_full_test.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/bash
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

#
# in order to run this script you'd first need to build the tile_example_fmha_fwd and tile_eaxmple_fmha_bwd executables in ../build/bin/
#
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/script/smoke_test_bwd.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/bash
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# TODO: run this script from CK root or build directory
set -euo pipefail

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/01_fmha/script/smoke_test_fwd.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/bash
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# TODO: run this script from CK root or build directory
set -euo pipefail

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/02_layernorm2d/layernorm2d_fwd.cpp
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include "ck_tile/host.hpp"
#include "layernorm2d_fwd.hpp"
#include "ck_tile/utility/json_dump.hpp"
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/02_layernorm2d/script/perf_test.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_layernorm2d_fwd -type f | head -n 1)"

$EXE -m=1 -n=1 -e=1e-12 -v=1 -prec_i=bf16 -repeat=1000
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/02_layernorm2d/script/smoke_test.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_layernorm2d_fwd -type f | head -n 1)"

for fquant in "" "-fquant=1 -prec_o=int8" "-fquant=1 -prec_o=fp8"; do
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/03_gemm/script/benchmark_basic_bf16.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_basic -type f | head -n 1)"
VALID=1

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/03_gemm/script/benchmark_basic_bf8.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_basic -type f | head -n 1)"
VALID=1

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/03_gemm/script/benchmark_basic_fp16.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_basic -type f | head -n 1)"
VALID=1

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/03_gemm/script/benchmark_basic_fp8.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_basic -type f | head -n 1)"
VALID=1

Expand Down
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_universal -type f | head -n 1)"
VALID=1

Expand Down
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_universal -type f | head -n 1)"
VALID=1

Expand Down
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_universal -type f | head -n 1)"
VALID=1

Expand Down
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_universal -type f | head -n 1)"
VALID=1

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/03_gemm/script/run_full_test.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/bash
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

#
# in order to run this script you'd first need to build the tile_example_gemm executables in ../build/bin/
#
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/03_gemm/script/smoke_test_basic.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/bash
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_basic -type f | head -n 1)"
KNAME=1

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/03_gemm/script/smoke_test_mem_pipeline.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/bash
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_example_gemm_universal -type f | head -n 1)"
KNAME=1

Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include "matrix_core_swizzle.hpp"
#include "matrix_core_swizzle_kernel.hpp"

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/06_permute/script/smoke_test.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

# TODO: run this script from CK root
BUILD=build
EXE=$BUILD/bin/tile_example_permute
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/09_topk_softmax/script/smoke_test.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT


EXE=./build/bin/tile_example_topk_softmax

Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/10_rmsnorm2d/example_rmsnorm2d_fwd.cpp
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include "ck_tile/host.hpp"
#include "ck_tile/core.hpp"
#include "ck_tile/host/kernel_launch.hpp"
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/10_rmsnorm2d/rmsnorm2d_fwd.cpp
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include "ck_tile/host.hpp"
#include "rmsnorm2d_fwd.hpp"
#include <cstring>
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/10_rmsnorm2d/script/perf_test.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_rmsnorm2d_fwd -type f | head -n 1)"

# 0: for no specific RMSNorm
Expand Down
3 changes: 3 additions & 0 deletions example/ck_tile/10_rmsnorm2d/script/smoke_test.sh
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/bash
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT


EXE="$(find . -name tile_rmsnorm2d_fwd -type f | head -n 1)"

Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include "ck_tile/host.hpp"
#include "add_rmsnorm2d_rdquant_fwd.hpp"
#include <cstring>
Expand Down
Original file line number Diff line number Diff line change
@@ -1,3 +1,6 @@
// Copyright © Advanced Micro Devices, Inc. or its affiliates.
// SPDX-License-Identifier: MIT

#include "ck_tile/host.hpp"
#include "ck_tile/core.hpp"
#include "ck_tile/host/kernel_launch.hpp"
Expand Down
Original file line number Diff line number Diff line change
@@ -1,4 +1,7 @@
#!/bin/sh
## Copyright © Advanced Micro Devices, Inc. or its affiliates.
## SPDX-License-Identifier: MIT

EXE="$(find . -name tile_add_rmsnorm2d_rdquant_fwd -type f | head -n 1)"

$EXE -m=1 -n=1 -e=1e-12 -v=1 -prec=bf16 -repeat=1000
Expand Down
Loading
Loading