From a328df25a131b5c1d30cbeadd4255ff39f19f977 Mon Sep 17 00:00:00 2001 From: Illia Silin <98187287+illsilin@users.noreply.github.com> Date: Mon, 8 Jul 2024 21:21:16 -0700 Subject: [PATCH 1/5] Fix the cmake logic when building with INSTANCES_ONLY=ON. (#1376) * fix the cmake logic when building for various targets * another minor fix --- example/CMakeLists.txt | 4 ++-- library/src/tensor_operation_instance/gpu/CMakeLists.txt | 8 ++++---- test/CMakeLists.txt | 4 ++-- 3 files changed, 8 insertions(+), 8 deletions(-) diff --git a/example/CMakeLists.txt b/example/CMakeLists.txt index 87c5a89f8a..45cfee4de9 100644 --- a/example/CMakeLists.txt +++ b/example/CMakeLists.txt @@ -67,7 +67,7 @@ function(add_example_executable EXAMPLE_NAME FILE_NAME) endforeach() #Do not build any WMMA examples if gfx11 targets are not on the list foreach(source IN LISTS FILE_NAME) - if(NOT GPU_TARGETS MATCHES "gfx11" AND NOT GPU_TARGETS MATCHES "gfx12" AND source MATCHES "_wmma") + if(NOT EX_TARGETS MATCHES "gfx11" AND NOT EX_TARGETS MATCHES "gfx12" AND source MATCHES "_wmma") message("removing wmma example ${source} ") list(REMOVE_ITEM FILE_NAME "${source}") endif() @@ -154,7 +154,7 @@ function(add_example_executable_no_testing EXAMPLE_NAME FILE_NAME) endforeach() #Do not build any WMMA examples if gfx11 targets are not on the list foreach(source IN LISTS FILE_NAME) - if(NOT GPU_TARGETS MATCHES "gfx11" AND NOT GPU_TARGETS MATCHES "gfx12" AND source MATCHES "_wmma") + if(NOT EX_TARGETS MATCHES "gfx11" AND NOT EX_TARGETS MATCHES "gfx12" AND source MATCHES "_wmma") message("removing wmma example ${source} ") list(REMOVE_ITEM FILE_NAME "${source}") endif() diff --git a/library/src/tensor_operation_instance/gpu/CMakeLists.txt b/library/src/tensor_operation_instance/gpu/CMakeLists.txt index 1bcc0f802b..2081422e3a 100644 --- a/library/src/tensor_operation_instance/gpu/CMakeLists.txt +++ b/library/src/tensor_operation_instance/gpu/CMakeLists.txt @@ -59,7 +59,7 @@ function(add_instance_library INSTANCE_NAME) endforeach() # Do not build WMMA instances if gfx11 targets are not on the target list foreach(source IN LISTS ARGN) - if(NOT GPU_TARGETS MATCHES "gfx11" AND NOT GPU_TARGETS MATCHES "gfx12" AND source MATCHES "_wmma") + if(NOT INST_TARGETS MATCHES "gfx11" AND NOT INST_TARGETS MATCHES "gfx12" AND source MATCHES "_wmma") message("removing wmma instance ${source} ") list(REMOVE_ITEM ARGN "${source}") endif() @@ -177,7 +177,7 @@ FOREACH(subdir_path ${dir_list}) message("Found only xdl instances, but gfx9 is not on the targets list. Skipping.") set(add_inst 0) endif() - if(("${cmake_instance}" MATCHES "ONLY WMMA_KERNELS") AND (NOT GPU_TARGETS MATCHES "gfx11") AND (NOT GPU_TARGETS MATCHES "gfx12")) + if(("${cmake_instance}" MATCHES "ONLY WMMA_KERNELS") AND (NOT INST_TARGETS MATCHES "gfx11") AND (NOT INST_TARGETS MATCHES "gfx12")) message("Found only wmma instances, but gfx11 is not on the targets list. Skipping.") set(add_inst 0) endif() @@ -185,11 +185,11 @@ FOREACH(subdir_path ${dir_list}) message("Found only xdl and dl instances, but gfx9 is not on the targets listand DL_KERNELS is not set. Skipping.") set(add_inst 0) endif() - if(("${cmake_instance}" MATCHES "ONLY XDL_AND_WMMA_KERNELS") AND (NOT GPU_TARGETS MATCHES "gfx11") AND (NOT GPU_TARGETS MATCHES "gfx12") AND (NOT GPU_TARGETS MATCHES "gfx9")) + if(("${cmake_instance}" MATCHES "ONLY XDL_AND_WMMA_KERNELS") AND (NOT INST_TARGETS MATCHES "gfx11") AND (NOT INST_TARGETS MATCHES "gfx12") AND (NOT INST_TARGETS MATCHES "gfx9")) message("Found only xdl and wmma instances, but gfx11 and gfx9 are not on the targets list. Skipping.") set(add_inst 0) endif() - if(("${cmake_instance}" MATCHES "XDL_DL_WMMA_KERNELS") AND (NOT GPU_TARGETS MATCHES "gfx11") AND (NOT GPU_TARGETS MATCHES "gfx12") AND (NOT GPU_TARGETS MATCHES "gfx9") AND (NOT DEFINED DL_KERNELS)) + if(("${cmake_instance}" MATCHES "XDL_DL_WMMA_KERNELS") AND (NOT INST_TARGETS MATCHES "gfx11") AND (NOT INST_TARGETS MATCHES "gfx12") AND (NOT INST_TARGETS MATCHES "gfx9") AND (NOT DEFINED DL_KERNELS)) message("Found xdl, dl, and wmma instances, but none of those meet the target list. Skipping.") set(add_inst 0) endif() diff --git a/test/CMakeLists.txt b/test/CMakeLists.txt index 3b121fc309..fc1bcfdb27 100644 --- a/test/CMakeLists.txt +++ b/test/CMakeLists.txt @@ -60,7 +60,7 @@ function(add_test_executable TEST_NAME) endif() endforeach() foreach(source IN LISTS ARGN) - if(NOT GPU_TARGETS MATCHES "gfx11" AND NOT GPU_TARGETS MATCHES "gfx12" AND source MATCHES "wmma") + if(NOT TEST_TARGETS MATCHES "gfx11" AND NOT TEST_TARGETS MATCHES "gfx12" AND source MATCHES "wmma") message("removing wmma test ${source} ") list(REMOVE_ITEM ARGN "${source}") endif() @@ -141,7 +141,7 @@ function(add_gtest_executable TEST_NAME) endif() endforeach() foreach(source IN LISTS ARGN) - if(NOT GPU_TARGETS MATCHES "gfx11" AND NOT GPU_TARGETS MATCHES "gfx12" AND source MATCHES "wmma") + if(NOT TEST_TARGETS MATCHES "gfx11" AND NOT TEST_TARGETS MATCHES "gfx12" AND source MATCHES "wmma") message("removing wmma test ${source} ") list(REMOVE_ITEM ARGN "${source}") endif() From ccfdc5302238198e0e0e5c0c3f05f41b79fcebb8 Mon Sep 17 00:00:00 2001 From: carlushuang Date: Tue, 9 Jul 2024 20:30:07 +0800 Subject: [PATCH 2/5] update owner (#1377) * remove zjing14, add poyenc * remove yigex --- .github/CODEOWNERS | 12 ++++++------ 1 file changed, 6 insertions(+), 6 deletions(-) diff --git a/.github/CODEOWNERS b/.github/CODEOWNERS index bc49ac1669..de17acb9cd 100644 --- a/.github/CODEOWNERS +++ b/.github/CODEOWNERS @@ -1,8 +1,8 @@ -* @zjing14 @junliume @illsilin @carlushuang @aosewski @yigex +* @junliume @illsilin @carlushuang @aosewski @poyenc # Documentation files -docs/* @ROCm/rocm-documentation @zjing14 @junliume @illsilin @carlushuang @aosewski @yigex -*.md @ROCm/rocm-documentation @zjing14 @junliume @illsilin @carlushuang @aosewski @yigex -*.rst @ROCm/rocm-documentation @zjing14 @junliume @illsilin @carlushuang @aosewski @yigex -.readthedocs.yaml @ROCm/rocm-documentation @zjing14 @junliume @illsilin @carlushuang @aosewski @yigex +docs/* @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc +*.md @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc +*.rst @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc +.readthedocs.yaml @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc # Header directory for Doxygen documentation -library/include/* @ROCm/rocm-documentation @zjing14 @junliume @illsilin @carlushuang @aosewski @yigex +library/include/* @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc From da42a889645c03d80e61423531fecfdc188c2ab9 Mon Sep 17 00:00:00 2001 From: "dependabot[bot]" <49699333+dependabot[bot]@users.noreply.github.com> Date: Tue, 9 Jul 2024 12:48:23 -0700 Subject: [PATCH 3/5] Bump rocm-docs-core from 1.4.1 to 1.5.0 in /docs/sphinx (#1374) Bumps [rocm-docs-core](https://github.com/ROCm/rocm-docs-core) from 1.4.1 to 1.5.0. - [Release notes](https://github.com/ROCm/rocm-docs-core/releases) - [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md) - [Commits](https://github.com/ROCm/rocm-docs-core/compare/v1.4.1...v1.5.0) --- updated-dependencies: - dependency-name: rocm-docs-core dependency-type: direct:production update-type: version-update:semver-minor ... Signed-off-by: dependabot[bot] Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com> Co-authored-by: Sam Wu <22262939+samjwu@users.noreply.github.com> --- docs/sphinx/requirements.in | 2 +- docs/sphinx/requirements.txt | 2 +- 2 files changed, 2 insertions(+), 2 deletions(-) diff --git a/docs/sphinx/requirements.in b/docs/sphinx/requirements.in index 6605380a51..51bfef2898 100644 --- a/docs/sphinx/requirements.in +++ b/docs/sphinx/requirements.in @@ -1,2 +1,2 @@ -rocm-docs-core==1.4.1 +rocm-docs-core==1.5.0 sphinxcontrib-bibtex==2.6.2 diff --git a/docs/sphinx/requirements.txt b/docs/sphinx/requirements.txt index a3566090ef..6d2fe6ca57 100644 --- a/docs/sphinx/requirements.txt +++ b/docs/sphinx/requirements.txt @@ -103,7 +103,7 @@ requests==2.31.0 # via # pygithub # sphinx -rocm-docs-core==1.4.1 +rocm-docs-core==1.5.0 # via -r requirements.in six==1.16.0 # via From 860f957c22b99b9b6add0b1b9132fd78079f3d10 Mon Sep 17 00:00:00 2001 From: Sam Wu <22262939+samjwu@users.noreply.github.com> Date: Wed, 10 Jul 2024 09:36:10 -0600 Subject: [PATCH 4/5] Update changelog release headers (#1378) * Update doc codeowner syntax * Add doc link to changelog * Update changelog formatting for markdownlint Also change headings for releases --- .github/CODEOWNERS | 4 ++-- CHANGELOG.md | 27 ++++++++++++++++++++------- 2 files changed, 22 insertions(+), 9 deletions(-) diff --git a/.github/CODEOWNERS b/.github/CODEOWNERS index de17acb9cd..1809abebba 100644 --- a/.github/CODEOWNERS +++ b/.github/CODEOWNERS @@ -1,8 +1,8 @@ * @junliume @illsilin @carlushuang @aosewski @poyenc # Documentation files -docs/* @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc +docs/ @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc *.md @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc *.rst @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc .readthedocs.yaml @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc # Header directory for Doxygen documentation -library/include/* @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc +library/include/ @ROCm/rocm-documentation @junliume @illsilin @carlushuang @aosewski @poyenc diff --git a/CHANGELOG.md b/CHANGELOG.md index fb2ba1975f..dec6334cf5 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -1,38 +1,46 @@ # Changelog for Composable Kernel -Full documentation for Composable Kernel is not yet available. +Documentation for Composable Kernel available at [https://rocm.docs.amd.com/projects/composable_kernel/en/latest/](https://rocm.docs.amd.com/projects/composable_kernel/en/latest/). -## CK for ROCm 6.1.0 +## Composable Kernel 1.1.0 for ROCm 6.1.0 ### Additions + * Added generic instances for GEMM XDL operations (#1161) * Added gamma and beta parameters for the layernorm and groupnorm bwd operations (#1133) * Introduced wrapper sublibrary (limited functionality). (#1071, #1098, #1108, #1126) * Added an option to vary the number of warm-up cycles and iterations for ckProfiler (#1124) ### Optimizations + * New performance optimizations for GEMM operations on MI200 and MI300 architectures (#1135) ### Fixes + * Reduced the build time for most GPU architectures (#1084) * Fixed some conversion issues for fp8 data type (#1099) ### Changes + None ### Known issues + None -## CK for ROCm 6.0.0 +## Composable Kernel 1.1.0 for ROCm 6.0.0 ### Fixes - * Fixed a hazard associated with inline v_dot (#808) - * Fixed two bugs in grouped convolution backward data without K padding (#848 #876) + +* Fixed a hazard associated with inline v_dot (#808) +* Fixed two bugs in grouped convolution backward data without K padding (#848 #876) ### Optimizations + None ### Additions + * Added an image to a column kernel (#867) * Added a column to an image kernel (#930) * Support for 3D grouped convolution on RDNA 3 GPUs (#935, #950, #985) @@ -42,18 +50,22 @@ None * Support for Batched GEMM DL (#732) ### Changes - * Changed the grouped convolution API to maintain consistency with other convolution kernels (#817) -## CK 0.2.0 for ROCm 5.7.0 +* Changed the grouped convolution API to maintain consistency with other convolution kernels (#817) + +## Composable Kernel 0.2.0 for ROCm 5.7.0 ### Fixes + * Fixed a bug in 6-dimensional kernels (#555) * Fixed a test case failure with grouped convolution backward weight (#524) ### Optimizations + * Improved the performance of the normalization kernel ### Additions + * New CMake flags: * "DL_KERNELS"-* Must be set to "ON" in order to build the GEMM DL and batched_gemm_multi_d_dl instances * "DTYPES" -- Can be set to any subset of "fp64;fp32;fp16;fp8;bf16;int8" to build an instance of the specified data types @@ -71,4 +83,5 @@ None * MaxPool and AvgPool forward (#815); MaxPool backward (#750) ### Changes + None From a8eb872055f1f741fad8033611ca6c8aacdd10a8 Mon Sep 17 00:00:00 2001 From: Illia Silin <98187287+illsilin@users.noreply.github.com> Date: Wed, 10 Jul 2024 14:54:04 -0700 Subject: [PATCH 5/5] [gfx12] add gfx12 to the default target list (#1379) --- CMakeLists.txt | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index b27e6ab4fb..fc0cc4ddba 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -112,7 +112,7 @@ message("checking which targets are supported") #Setting GPU_TARGETS on command line will override this list if(NOT PROFILER_ONLY) rocm_check_target_ids(DEFAULT_GPU_TARGETS - TARGETS "gfx908;gfx90a;gfx940;gfx941;gfx942;gfx1030;gfx1100;gfx1101;gfx1102") + TARGETS "gfx908;gfx90a;gfx940;gfx941;gfx942;gfx1030;gfx1100;gfx1101;gfx1102;gfx1200;gfx1201") else() add_definitions(-DPROFILER_ONLY) set(GPU_TARGETS "" CACHE STRING "" FORCE) @@ -148,7 +148,7 @@ if (GPU_TARGETS) add_definitions(-DCK_USE_XDL) set(CK_USE_XDL "ON") endif() - if (GPU_TARGETS MATCHES "gfx11") + if (GPU_TARGETS MATCHES "gfx11" OR GPU_TARGETS MATCHES "gfx12") add_definitions(-DCK_USE_WMMA) set(CK_USE_WMMA "ON") endif()