ROCm/composable_kernel
1
0
Fork 0
mirror of https://github.com/ROCm/composable_kernel.git synced 2026-06-29 19:28:33 +00:00
Code Issues Packages Projects Releases Wiki Activity
Files
felix/flatmm
composable_kernel/library/src/tensor_operation_instance/gpu
History
Andriy Roshchenko 25b359d630 MX GEMM - Add FP6 GEMM Test (#2488) 
* Add F6 GEMM MX Test

* Add BF6 GEMM MX Test
2025-07-11 15:32:12 -06:00
..
avg_pool2d_bwd
Pool2d max/avg kernel in the BWD version (#1494)
2024-09-12 11:47:52 +02:00
avg_pool3d_bwd
Refactoring cmake files to build data types separately. (#932)
2023-09-20 22:15:56 -07:00
batched_gemm
Implement batched gemm wmma (RDNA batched gemm) based on wmma cshuffle v3 (#2319)
2025-06-24 07:28:13 -07:00
batched_gemm_add_relu_gemm_add
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
batched_gemm_b_scale
Added Int4 mixed batch gemm support (#1839)
2025-02-10 11:17:02 +08:00
batched_gemm_bias_permute
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
batched_gemm_gemm
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
batched_gemm_multi_d
Refactoring cmake files to build data types separately. (#932)
2023-09-20 22:15:56 -07:00
batched_gemm_reduce
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
batched_gemm_softmax_gemm
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
batched_gemm_softmax_gemm_permute
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
batchnorm
Refactor elementwise kernels (#1222)
2024-04-19 13:31:17 +02:00
column_to_image
Add support for groups in Img2Col/Col2Img (#1007)
2023-10-31 10:46:32 +01:00
contraction_bilinear
Extend support for contraction 6D (#1207)
2024-04-09 23:46:21 +02:00
contraction_scale
Extend support for contraction 6D (#1207)
2024-04-09 23:46:21 +02:00
conv1d_bwd_data
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
conv2d_bwd_data
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
conv2d_fwd
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
conv2d_fwd_bias_relu
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
conv2d_fwd_bias_relu_add
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
conv3d_bwd_data
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
elementwise
Refactor elementwise kernels (#1222)
2024-04-19 13:31:17 +02:00
elementwise_normalization
Refactoring cmake files to build data types separately. (#932)
2023-09-20 22:15:56 -07:00
gemm
Optimized GEMMs for MX FP4/8 (#2294)
2025-06-05 13:54:15 -06:00
gemm_ab_scale
[Block Scale GEMM] Optimized block scale gemm (#1950)
2025-03-11 10:11:21 -07:00
gemm_add
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_add_add_fastgelu
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
gemm_add_fastgelu
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
gemm_add_multiply
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_add_relu
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_add_relu_add_layernorm
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
gemm_add_silu
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_b_scale
Added Int4 mixed batch gemm support (#1839)
2025-02-10 11:17:02 +08:00
gemm_bias_add_reduce
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_bilinear
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_blockscale_wp
check for if misched-bottomup flag is valid (#2341)
2025-06-13 13:34:22 -07:00
gemm_fastgelu
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
gemm_multi_abd
Split the gemm_multi_abd instances. (#1306)
2024-05-23 09:17:02 -07:00
gemm_multiply_add
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_multiply_multiply
[Block Scale GEMM] Optimized block scale gemm (#1950)
2025-03-11 10:11:21 -07:00
gemm_multiply_multiply_wp
Convert CK (GeMM MulMul Weight Preshuffle) instances to use 16x16 xdl tile (#2229)
2025-06-10 09:37:14 -07:00
gemm_mx
MX GEMM - Add FP6 GEMM Test (#2488)
2025-07-11 15:32:12 -06:00
gemm_reduce
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_splitk
Optimized GEMMs for MX FP4/8 (#2294)
2025-06-05 13:54:15 -06:00
gemm_streamk
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
gemm_universal
WMMA GEMM universal pipeline v1, mixed precision and paddings, examples (#2230)
2025-06-04 12:22:33 +06:00
gemm_universal_batched
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
gemm_universal_preshuffle
[ckProfiler] Add infrastructure and instances to profile gemm_universal with B preshuffle (#2427)
2025-07-01 18:34:52 -07:00
gemm_universal_reduce
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
gemm_universal_streamk
chore: unset executable permission (#2303)
2025-06-10 09:13:59 -07:00
grouped_conv1d_bwd_weight
Enable grouped conv bwd wei bf16 NGCHW (#1589)
2024-10-22 16:18:28 +02:00
grouped_conv1d_fwd
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
grouped_conv2d_bwd_data
Integrate universal gemm with conv bwd data and add SplitK (#1315)
2025-04-28 23:54:49 +02:00
grouped_conv2d_bwd_weight
Grouped conv bwd wei add for larger filter and Merge Groupes optimization (#2197)
2025-05-21 22:47:34 +02:00
grouped_conv2d_fwd
[CK][CONV] Support NCHW in class DeviceGroupedConvFwdMultipleABD_Xdl_CShuffle (#2375)
2025-06-26 08:32:39 +08:00
grouped_conv2d_fwd_bias_clamp
Grouped conv bias clamp fp32/fp16 support (#2366)
2025-06-20 11:41:04 +02:00
grouped_conv2d_fwd_clamp
Grouped conv bias clamp fp32/fp16 support (#2366)
2025-06-20 11:41:04 +02:00
grouped_conv2d_fwd_dynamic_op
Temporary disable part of dynamic op conv instances (#1630)
2024-11-04 13:34:17 -08:00
grouped_conv3d_bwd_data
Integrate universal gemm with conv bwd data and add SplitK (#1315)
2025-04-28 23:54:49 +02:00
grouped_conv3d_bwd_data_bilinear
Integrate universal gemm with conv bwd data and add SplitK (#1315)
2025-04-28 23:54:49 +02:00
grouped_conv3d_bwd_data_scale
Integrate universal gemm with conv bwd data and add SplitK (#1315)
2025-04-28 23:54:49 +02:00
grouped_conv3d_bwd_weight
Grouped conv bwd wei add for larger filter and Merge Groupes optimization (#2197)
2025-05-21 22:47:34 +02:00
grouped_conv3d_bwd_weight_bilinear
Re-enable fp8 types for all architectures. (#1470)
2024-08-16 16:07:52 -06:00
grouped_conv3d_bwd_weight_scale
Re-enable fp8 types for all architectures. (#1470)
2024-08-16 16:07:52 -06:00
grouped_conv3d_fwd
[CK][CONV] Support NCHW in class DeviceGroupedConvFwdMultipleABD_Xdl_CShuffle (#2375)
2025-06-26 08:32:39 +08:00
grouped_conv3d_fwd_bias_clamp
Grouped conv bias clamp fp32/fp16 support (#2366)
2025-06-20 11:41:04 +02:00
grouped_conv3d_fwd_bilinear
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
grouped_conv3d_fwd_clamp
Grouped conv bias clamp fp32/fp16 support (#2366)
2025-06-20 11:41:04 +02:00
grouped_conv3d_fwd_convinvscale
Add a convinvscale op, related instances and examples (#1307)
2024-06-10 14:48:49 -05:00
grouped_conv3d_fwd_convscale
Adding Instances and Examples for FP8-based Scaled Convolution and AMAX Reduction. (#1473)
2024-08-21 15:22:41 -07:00
grouped_conv3d_fwd_convscale_add
Adding more instances of grouped convolution 3d forward for FP8 with ConvScale+Bias element-wise operation. (#1412)
2024-07-24 15:49:55 -05:00
grouped_conv3d_fwd_convscale_relu
Adding Instances and Examples for FP8-based Scaled Convolution and AMAX Reduction. (#1473)
2024-08-21 15:22:41 -07:00
grouped_conv3d_fwd_dynamic_op
Temporary disable part of dynamic op conv instances (#1630)
2024-11-04 13:34:17 -08:00
grouped_conv3d_fwd_scale
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
grouped_conv3d_fwd_scaleadd_ab
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
grouped_conv3d_fwd_scaleadd_scaleadd_relu
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
grouped_convnd_bwd_weight
Grouped conv bwd wei explicit GEMM for odd C/K (#2306)
2025-06-10 11:17:12 +02:00
grouped_gemm
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
grouped_gemm_bias
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
grouped_gemm_fastgelu
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
grouped_gemm_fixed_nk
Added bf16 instances grouped gemm fixed nk (#1825)
2025-01-20 09:13:09 -08:00
grouped_gemm_fixed_nk_multi_abd
bf16A_Int8B with fastgelu/bias (#1264)
2024-04-26 07:26:30 -05:00
grouped_gemm_tile_loop
Remove CK_USE_AMD_MFMA_GFX950 (#1935)
2025-03-04 10:32:25 -08:00
image_to_column
Add support for groups in Img2Col/Col2Img (#1007)
2023-10-31 10:46:32 +01:00
max_pool_bwd
This commit contains implementation of max pool2d for f8 type (#1506)
2024-09-16 10:15:06 +02:00
mha
Label CMakeLists message() as DEBUG or STATUS for clean build output (#2301)
2025-06-10 10:46:47 -07:00
normalization_bwd_data
layernorm and groupnorm backward data (#1083)
2023-12-19 04:23:11 +08:00
normalization_bwd_gamma_beta
layernorm & groupnorm bwd gamma beta (#1133)
2024-01-25 19:53:15 +08:00
normalization_fwd
layernorm and groupnorm backward data (#1083)
2023-12-19 04:23:11 +08:00
permute_scale
Adding Instances and Examples for FP8-based Scaled Convolution with ReLU Activation and AMAX Reduction. (#1469)
2024-08-20 10:30:56 -05:00
pool2d_fwd
Add pool2d int8 and fp8 instances (#1508)
2024-09-13 10:18:21 -07:00
pool3d_fwd
OCP FP8 support for gfx12. (#1710)
2024-12-03 08:42:55 -08:00
quantization
Split the instances by architecture. (#1223)
2024-04-02 09:42:17 -07:00
reduce
Adding Instances and Examples for FP8-based Scaled Convolution with ReLU Activation and AMAX Reduction. (#1469)
2024-08-20 10:30:56 -05:00
softmax
enable compilation of INSTANCES_ONLY for Windows (#1082)
2023-12-20 14:34:53 -08:00
transpose
Hip tensor permute unit test (#1068)
2023-12-18 21:35:00 -06:00
CMakeLists.txt
[ckProfiler] Add infrastructure and instances to profile gemm_universal with B preshuffle (#2427)
2025-07-01 18:34:52 -07:00