ROCm/composable_kernel
1
0
Fork 0
mirror of https://github.com/ROCm/composable_kernel.git synced 2026-05-16 02:54:21 +00:00
Code Issues Packages Projects Releases Wiki Activity

Temporarily close the instance for hdim64 and hdim256 to save compiling time

Browse Source
This commit is contained in:
Qianfeng Zhang
2025-04-30 01:54:43 +00:00
parent d63dab90da
commit 2972de4c88
70 changed files with 3 additions and 1923 deletions
Download Patch File Download Diff File Expand all files Collapse all files

2
example/ck_tile/18_hstu_attention/generate_instances.py
View File

@@ -161,7 +161,7 @@ def create_forward_instances_ref(instance_dir: Path, headdims: List) -> None:
file.write(forward_instance)
if __name__ == "__main__":
headdims_fwd = [64, 128, 256]
headdims_fwd = [128]
this_dir = os.path.dirname(__file__)
output_dir = Path(this_dir) / "instances"

4
example/ck_tile/18_hstu_attention/hstu_attention_hdim_switch.hpp
View File

@@ -10,7 +10,7 @@
[&] { \
if(HDIM_1 <= 64 && HDIM_2 <= 64) \
{ \
constexpr ck_tile::index_t CONST_NAME = 64; \
constexpr ck_tile::index_t CONST_NAME = 128; \
__VA_ARGS__(); \
} \
else if(HDIM_1 <= 128 && HDIM_2 <= 128) \
@@ -20,7 +20,7 @@
} \
else if(HDIM_1 <= 256 && HDIM_2 <= 256) \
{ \
constexpr ck_tile::index_t CONST_NAME = 256; \
constexpr ck_tile::index_t CONST_NAME = 128; \
__VA_ARGS__(); \
} \
else \

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_has_causal_has_local_has_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_has_causal_has_local_has_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_has_causal_has_local_has_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_has_causal_has_local_has_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_has_causal_has_local_no_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_has_causal_has_local_no_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_has_causal_has_local_no_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_has_causal_has_local_no_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

128
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_instances_ref.hpp
View File

@@ -13,70 +13,6 @@
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
@@ -140,67 +76,3 @@ extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_no_causal_no_local_has_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_no_causal_no_local_has_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_no_causal_no_local_has_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_no_causal_no_local_has_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_no_causal_no_local_no_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_no_causal_no_local_no_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_no_causal_no_local_no_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_bf16_no_causal_no_local_no_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_has_causal_has_local_has_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_has_causal_has_local_has_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_has_causal_has_local_has_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_has_causal_has_local_has_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_has_causal_has_local_no_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_has_causal_has_local_no_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_has_causal_has_local_no_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_has_causal_has_local_no_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

128
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_instances_ref.hpp
View File

@@ -13,70 +13,6 @@
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
@@ -140,67 +76,3 @@ extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_no_causal_no_local_has_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_no_causal_no_local_has_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_no_causal_no_local_has_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_no_causal_no_local_has_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_no_causal_no_local_no_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_no_causal_no_local_no_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_no_causal_no_local_no_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_batched_forward_fp16_no_causal_no_local_no_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_batched_forward_dispatch.hpp"
template void run_batched_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_has_causal_has_local_has_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_has_causal_has_local_has_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_has_causal_has_local_has_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_has_causal_has_local_has_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_has_causal_has_local_no_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_has_causal_has_local_no_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_has_causal_has_local_no_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_has_causal_has_local_no_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

128
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_instances_ref.hpp
View File

@@ -13,70 +13,6 @@
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
@@ -140,67 +76,3 @@ extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
true,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_no_causal_no_local_has_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_no_causal_no_local_has_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_no_causal_no_local_has_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_no_causal_no_local_has_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_no_causal_no_local_no_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_no_causal_no_local_no_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_no_causal_no_local_no_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_bf16_no_causal_no_local_no_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/half.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::bf16_t,
false,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_has_causal_has_local_has_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_has_causal_has_local_has_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_has_causal_has_local_has_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_has_causal_has_local_has_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_has_causal_has_local_no_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_has_causal_has_local_no_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_has_causal_has_local_no_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_has_causal_has_local_no_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

128
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_instances_ref.hpp
View File

@@ -13,70 +13,6 @@
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
@@ -140,67 +76,3 @@ extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
true,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);
extern template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_no_causal_no_local_has_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_no_causal_no_local_has_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_no_causal_no_local_has_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_no_causal_no_local_has_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_no_causal_no_local_no_bias_has_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_no_causal_no_local_no_bias_has_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_no_causal_no_local_no_bias_no_dropout_maxk_256.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

22
example/ck_tile/18_hstu_attention/instances/hstu_attention_jagged_forward_fp16_no_causal_no_local_no_bias_no_dropout_maxk_64.cpp
View File

@@ -1,22 +0,0 @@
/*
Copyright (c) 2024, Advanced Micro Devices, Inc. All rights reserved.
*
* This source code is licensed under the BSD-style license found in the
* LICENSE file in the root directory of this source tree.
*
* The file is automatically generated, don't modify!
* See the generator script
* `composable_kernel/example/ck_tile/18_hstu_attention/generate_instances.py`
*/
#include <ck_tile/core/numeric/bfloat16.hpp>
#include "hstu_attention_jagged_forward_dispatch.hpp"
template void run_jagged_forward_causal_local_bias_dropout_dispatch<
ck_tile::fp16_t,
false,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);