Fix bug in generate_instances.py and re-generate the instances

This commit is contained in:
Qianfeng Zhang
2025-07-17 08:37:09 +00:00
parent 0306a1e6ab
commit f0c8dca9bf
97 changed files with 2114 additions and 2 deletions

View File

@@ -88,7 +88,7 @@ MODE_NAME_MAP = {
def create_forward_instances(instance_dir: Path, headdims: List) -> None:
for mode in ["batched", "jagged"]:
for dtype in ["fp16", "bf16"]:
for has_causal, has_local in zip([True, False],[True, False]):
for has_causal, has_local in ([True, True], [True, False], [False, True], [False, False]):
for has_bias in [True, False]:
for has_dropout in [True, False]:
for max_k in headdims:
@@ -112,7 +112,7 @@ def create_forward_instances(instance_dir: Path, headdims: List) -> None:
mode=mode,
dtype=TYPE_CTYPE_MAP[dtype],
has_causal=BOOL_MAP[has_causal],
has_local=BOOL_MAP[has_causal],
has_local=BOOL_MAP[has_local],
has_bias=BOOL_MAP[has_bias],
has_dropout=BOOL_MAP[has_dropout],
max_k=max_k,

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
false,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
true,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
true,
64>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
128>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
256>(HstuAttentionFwdParams& param, hipStream_t stream);

View File

@@ -0,0 +1,22 @@
/*
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,
true,
false,
false,
64>(HstuAttentionFwdParams& param, hipStream_t stream);