sci-libs/composable-kernel: add 7.0.2

Bug: https://bugs.gentoo.org/964799

Signed-off-by: Sv. Lockal <lockalsash@gmail.com>
Part-of: https://github.com/gentoo/gentoo/pull/44165
Signed-off-by: Sam James <sam@gentoo.org>
This commit is contained in:
Sv. Lockal
2025-10-12 07:24:21 +00:00
committed by Sam James
parent d885903e2c
commit 9355a33a68
4 changed files with 595 additions and 0 deletions

View File

@@ -1,3 +1,4 @@
DIST composable-kernel-6.3.0.tar.gz 3376552 BLAKE2B ad3dde27d0134fa8bf6b5935603f7cac17c959ab7b123e8d69cfe82bee188ade0fe9d3666d46d9819de479e875e46c0fdad2574f68a3bcf0e9acbb871f163b24 SHA512 4fa7be3f54d06d894dca7ef9fa8e3db24971fb857fee96e7ccff89844fa466789756d5f140d2791a984311d59a044b5351d0358e676c218ea0b0aadf60954d8e
DIST composable-kernel-6.4.3.tar.gz 4373178 BLAKE2B 1128f5de53e31365ff44d2c0b540ab662cd5c1659018d15f35e4c7036c8cf613ba5b95f910b7b3af94673a08ee36917fc38f9f19140140f2779252abda976fa4 SHA512 7073889f9af312a366eae9924771336e40b378255150a48308ce3975e254378da07bb49d615fbda7cc89dab35a14269df776ca88f483b803663cce325e463bb9
DIST composable-kernel-7.0.2.tar.gz 5171930 BLAKE2B 2f54c21888ab8a0e76c0b4e5734b8e0c556b313e36ededed0610ea00ea8f0e4e72a757392cc2f47c84fa7fa7da0a72c41bd18b74b3a88b40e53ab7dd6ff2efbd SHA512 84c85ecad99d94a6c0bb159c1853d81e39fcb4d02c63a109cada84297f62b90aa9eb3d9102cb8e50b0c4f84982d8e5c8c8fdc5377acd014493621c9aec5439e0
DIST gtest-1.14.0_p20220421.tar.gz 849107 BLAKE2B e9c32d9c1d98959583b696430a6be411f5d69e3db96669b16cffc0f1dcad42512391c0a4733f6e829c82953ff33fa243ee782f3cf6e86436f0d290f45d1a3db7 SHA512 967aac7d85da0e216fe13b17d10f894a31d763d9b88201bac7d3dfc37600552b5472d30a166a6ef27f2778677e73fd3e43d082695a48f47824262eb9cd4fcd2f

View File

@@ -0,0 +1,162 @@
# Copyright 1999-2025 Gentoo Authors
# Distributed under the terms of the GNU General Public License v2
EAPI=8
ROCM_VERSION=${PV}
PYTHON_COMPAT=( python3_{10..14} python3_13t )
inherit check-reqs cmake flag-o-matic multiprocessing python-r1 rocm
GTEST_COMMIT="b85864c64758dec007208e56af933fc3f52044ee"
GTEST_FILE="gtest-1.14.0_p20220421.tar.gz"
DESCRIPTION="High Performance Composable Kernel for AMD GPUs"
HOMEPAGE="https://github.com/ROCm/composable_kernel"
SRC_URI="https://github.com/ROCm/composable_kernel/archive/rocm-${PV}.tar.gz -> ${P}.tar.gz
test? ( https://github.com/google/googletest/archive/${GTEST_COMMIT}.tar.gz -> ${GTEST_FILE} )"
S="${WORKDIR}/composable_kernel-rocm-${PV}"
LICENSE="MIT"
SLOT="0/$(ver_cut 1-2)"
KEYWORDS="~amd64"
IUSE="debug profiler test"
REQUIRED_USE="${ROCM_REQUIRED_USE} ${PYTHON_REQUIRED_USE}"
RESTRICT="!test? ( test )"
RDEPEND="
dev-util/hip:${SLOT}
${PYTHON_DEPS}
"
DEPEND="${RDEPEND}"
BDEPEND="
dev-build/rocm-cmake
"
PATCHES=(
"${FILESDIR}"/${PN}-6.1.1-no-git-no-hash.patch
"${FILESDIR}"/${PN}-6.3.0-conditional-kernels.patch
"${FILESDIR}"/${PN}-7.0.1-conditional-ckprofiler.patch
"${FILESDIR}"/${PN}-6.4.1-expand-isa.patch
"${FILESDIR}"/${PN}-7.0.1-libcxx-includes.patch
)
ck_check-reqs() {
[[ ${MERGE_TYPE} == binary ]] && return
targets=($AMDGPU_TARGETS)
if [[ ${#targets[@]} -gt 1 ]]; then
ewarn "composable-kernel will be compiled for multiple GPU architectures,"
ewarn "which will take a significant amount of time."
ewarn "Please consider setting AMDGPU_TARGETS USE_EXPAND variable to a single architecture."
fi
# It takes ~2Gb of RAM per build thread
local user_jobs=$(makeopts_jobs)
local available_memory_mb=$(free -m | awk '/Mem:/ {print $7}')
local max_jobs=$(( available_memory_mb / 2048 ))
max_jobs=$(( max_jobs < 1 ? 1 : max_jobs ))
local limited_jobs=$(( user_jobs < max_jobs ? user_jobs : max_jobs ))
if [[ "${max_jobs}" -lt "${user_jobs}" ]]; then
ewarn "${available_memory_mb} MB of free RAM is not enough for ${user_jobs} parallel build jobs (~2Gb per job)."
ewarn "Please consider setting MAKEOPTS=\"-j${limited_jobs}\" for this package."
fi
local CHECKREQS_MEMORY=$((user_jobs*2048))M
check-reqs_${EBUILD_PHASE_FUNC}
}
pkg_pretend() {
ck_check-reqs
}
pkg_setup() {
ck_check-reqs
}
src_prepare() {
sed -e '/-Werror/d' -i cmake/EnableCompilerWarnings.cmake || die
# don't build examples
sed -e "/add_subdirectory(example)/d" -i CMakeLists.txt || die
# Flag -amdgpu-early-inline-all explodes memory consumption
# https://github.com/llvm/llvm-project/issues/86332
sed -e "/-amdgpu-early-inline-all/d" -e "/-amdgpu-function-calls/d" -i CMakeLists.txt || die
cmake_src_prepare
}
src_configure() {
rocm_use_hipcc
if ! use debug; then
append-cflags "-DNDEBUG"
append-cxxflags "-DNDEBUG"
CMAKE_BUILD_TYPE="Release"
else
CMAKE_BUILD_TYPE="Debug"
fi
local mycmakeargs=(
-DCMAKE_SKIP_RPATH=ON
-DBUILD_DEV=OFF
-DGPU_TARGETS="$(get_amdgpu_flags)"
-DCMAKE_INSTALL_PREFIX="${EPREFIX}/usr"
-DBUILD_TESTING=$(usex test ON OFF)
-DCK_USE_PROFILER=$(usex profiler ON OFF)
-Wno-dev
)
# Since 6.4.1 "fallback" DL kernels should be enabled manually...
if use amdgpu_targets_gfx1010 || use amdgpu_targets_gfx1011 || use amdgpu_targets_gfx1012 \
|| use amdgpu_targets_gfx1030 || use amdgpu_targets_gfx1031 ; then
mycmakeargs+=(-DDL_KERNELS=ON)
fi
if use test; then
mycmakeargs+=(
-DFETCHCONTENT_SOURCE_DIR_GTEST="${WORKDIR}/googletest-${GTEST_COMMIT}"
)
fi
# rocminfo call during configuration; should not happen
# Bug: https://github.com/ROCm/composable_kernel/issues/2994
check_amdgpu
addpredict /dev/random
cmake_src_configure
}
src_install() {
cmake_src_install
# shellcheck disable=SC2329
installation() {
python_domodule python/ck4inductor
# install package-data manually, as there is no PEP517 compliance
shopt -s globstar
package_data=(
include/ck/**/*.hpp
library/src/tensor_operation_instance/gpu/gemm_universal/**/*.hpp
)
shopt -u globstar
inst_path="${D}$(python_get_sitedir)/ck4inductor"
for file in "${package_data[@]}"; do
location="${inst_path}/$(dirname "$file")"
mkdir -p "${location}"
cp "${file}" "${location}"
done
}
python_foreach_impl installation
}
src_test() {
check_amdgpu
LD_LIBRARY_PATH="${BUILD_DIR}"/lib cmake_src_test -j1
}

View File

@@ -0,0 +1,12 @@
Add a flag to build without ckprofiler, which takes few GB of space and not used in many cases.
--- a/CMakeLists.txt
+++ b/CMakeLists.txt
@@ -645,7 +645,7 @@ if(NOT GPU_ARCHS AND USER_GPU_TARGETS AND NOT MIOPEN_REQ_LIBS_ONLY)
endif()
endif()
-if (NOT MIOPEN_REQ_LIBS_ONLY)
+if (NOT MIOPEN_REQ_LIBS_ONLY OR CK_USE_PROFILER)
rocm_package_setup_component(profiler
LIBRARY_NAME composablekernel
PACKAGE_NAME ckprofiler

View File

@@ -0,0 +1,420 @@
--- a/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_wmma_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_wmma_cshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_multiple_d_xdl_cshuffle_v3.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_multiple_d_xdl_cshuffle_v3.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_wmma_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_wmma_cshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include <numeric>
#include <initializer_list>
--- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_wmma_cshuffle_v3.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_wmma_cshuffle_v3.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_xdl.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_xdl.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_xdl_fpAintB_b_scale.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_xdl_fpAintB_b_scale.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_contraction_multiple_abd_xdl_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_contraction_multiple_abd_xdl_cshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include <vector>
--- a/include/ck/tensor_operation/gpu/device/impl/device_fpAintB_gemm_wmma.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_fpAintB_gemm_wmma.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_dpp.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_dpp.hpp
@@ -3,6 +3,7 @@
#pragma once
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_abd_xdl_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_abd_xdl_cshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_layernorm_xdl_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_layernorm_xdl_cshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_wmma_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_wmma_cshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle.hpp
@@ -5,6 +5,7 @@
#ifndef __HIPCC_RTC__
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/host_utility/device_prop.hpp"
#include "ck/host_utility/kernel_launch.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_lds_direct_load.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_lds_direct_load.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_v3.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_v3.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_v3_ab_scale.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_v3_ab_scale.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_v3_b_preshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_v3_b_preshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_v3_blockscale_bpreshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle_v3_blockscale_bpreshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_wmma.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_wmma.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_wmma_cshuffle_v3.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_wmma_cshuffle_v3.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_lds_direct_load.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_lds_direct_load.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_streamk_v3.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_streamk_v3.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v2.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v2.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3_b_preshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3_b_preshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3_b_scale.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3_b_scale.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3_mx.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3_mx.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3r1.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_cshuffle_v3r1.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include <typeinfo>
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_splitk_c_shuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_splitk_c_shuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_splitk_c_shuffle_lds_direct_load.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_splitk_c_shuffle_lds_direct_load.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_bwd_weight_two_stage_xdl_cshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_bwd_weight_two_stage_xdl_cshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <numeric>
#include <sstream>
--- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle_v3.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle_v3.hpp
@@ -6,6 +6,7 @@
#include <functional>
#include <iostream>
#include <iterator>
+#include <map>
#include <numeric>
#include <sstream>
--- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_gemm_multiple_d_xdl_cshuffle_tile_loop.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_gemm_multiple_d_xdl_cshuffle_tile_loop.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include <tuple>
--- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_query_attention_forward_wmma.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_query_attention_forward_wmma.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include <numeric>
#include <initializer_list>
--- a/include/ck/tensor_operation/gpu/device/impl/device_moe_gemm.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_moe_gemm.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_moe_gemm_blockscale.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_moe_gemm_blockscale.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include <hip/hip_runtime.h>
--- a/include/ck/tensor_operation/gpu/device/impl/device_moe_mx_gemm.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_moe_mx_gemm.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_moe_mx_gemm_bns.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_moe_mx_gemm_bns.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_moe_mx_gemm_bpreshuffle.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_moe_mx_gemm_bpreshuffle.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include "ck/utility/common_header.hpp"
--- a/include/ck/tensor_operation/gpu/device/impl/device_multi_query_attention_forward_wmma.hpp
+++ b/include/ck/tensor_operation/gpu/device/impl/device_multi_query_attention_forward_wmma.hpp
@@ -4,6 +4,7 @@
#pragma once
#include <iostream>
+#include <map>
#include <sstream>
#include <numeric>
#include <initializer_list>