ROCm/composable_kernel
1
0
Fork 0
mirror of https://github.com/ROCm/composable_kernel.git synced 2026-06-30 03:37:38 +00:00
Code Issues Packages Projects Releases Wiki Activity

fix backward compatibility

Browse Source
This commit is contained in:
Feng Shijie
2025-09-09 04:27:33 +00:00
parent b09b6cdce9
commit 08bae35abd
1 changed files with 46 additions and 0 deletions
Download Patch File Download Diff File Expand all files Collapse all files

46
include/ck_tile/ops/moe_flatmm/kernel/moe_flatmm_kernel.hpp
View File

@@ -31,6 +31,52 @@ struct MoeFlatmmHostArgs : ScaleFlatmmHostArgs<ScaleM, ScaleN, 0>
ExpertBias exp_bias;
CK_TILE_HOST MoeFlatmmHostArgs() noexcept = default;
CK_TILE_HOST MoeFlatmmHostArgs(const ck_tile::index_t* p_sorted_token_ids_,
const void* p_sorted_expert_weights_,
const ck_tile::index_t* p_sorted_expert_ids_,
const ck_tile::index_t* p_max_token_id_,
const void* a_ptr_,
const void* b_ptr_,
void* c_ptr_,
ck_tile::index_t NumTokens_,
ck_tile::index_t NumExperts_,
ck_tile::index_t TopK_,
ck_tile::index_t k_batch_,
ck_tile::index_t M_,
ck_tile::index_t N_,
ck_tile::index_t K_,
ck_tile::index_t stride_A_,
ck_tile::index_t stride_B_,
ck_tile::index_t stride_C_,
ScaleM scale_m_ = {},
ScaleN scale_n_ = {},
ExpertBias exp_bias_ = {})
: MoeFlatmmHostArgs(p_sorted_token_ids_,
p_sorted_expert_weights_,
p_sorted_expert_ids_,
p_max_token_id_,
a_ptr_,
b_ptr_,
c_ptr_,
NumTokens_,
NumExperts_,
TopK_,
k_batch_,
M_,
N_,
K_,
stride_A_,
stride_B_,
stride_C_,
0, // n_padded_zeros_
0, // k_padded_zeros_
scale_m_,
scale_n_,
exp_bias_)
{
}
CK_TILE_HOST MoeFlatmmHostArgs(const ck_tile::index_t* p_sorted_token_ids_,
const void* p_sorted_expert_weights_,
const ck_tile::index_t* p_sorted_expert_ids_,