ROCm/composable_kernel
1
0
Fork 0
mirror of https://github.com/ROCm/composable_kernel.git synced 2026-05-13 17:55:48 +00:00
Code Issues Packages Projects Releases Wiki Activity

[CK][CK_Tile] Ensure CK Tile engine benchmarking targets are excluded from default build. (#6135)

Browse Source 
## Motivation

Ensuring that tile engine benchmarking does not build by default and
slow other developers.

## Technical Details

- Added EXCLUDE_FROM_ALL to all add_subdirectory calls in
tile_engine/CMakeLists.txt and ops/gemm/CMakeLists.txt, so none of the
tile engine ops targets are part of the default all build.

- Added missing EXCLUDE_FROM_ALL to add_executable in
ops/pooling/CMakeLists.txt and ops/reduce/CMakeLists.txt (the GEMM
variants already had it).

- Downgraded message(STATUS ...) to message(VERBOSE ...) (or DEBUG for
per-target creation) in ops/pooling/, ops/gemm_streamk/, and ops/reduce/
CMakeLists. The other four GEMM variants (gemm_universal, gemm_multi_d,
gemm_preshuffle, grouped_gemm) already used VERBOSE.

- Targets can still be built on demand via their aggregate names (e.g.
make benchmark_pooling_all, make benchmark_gemm_streamk_all).


## Test Plan

Tile engine benchmark testing stage should be unaffected. 

## Test Result

N/A

## Submission Checklist

- [x] Look over the contributing guidelines at
https://github.com/ROCm/ROCm/blob/develop/CONTRIBUTING.md#pull-requests.
This commit is contained in:
Vidyasagar Ananthan
2026-04-03 15:07:58 -07:00
committed by GitHub
parent 12fe2c3de4
commit 9478c6c69f
5 changed files with 42 additions and 41 deletions
Download Patch File Download Diff File Expand all files Collapse all files

8
tile_engine/CMakeLists.txt
View File

@@ -5,8 +5,8 @@ include_directories(BEFORE
${CMAKE_CURRENT_LIST_DIR}/include
)
add_subdirectory(ops/gemm)
add_subdirectory(ops/gemm_streamk)
add_subdirectory(ops/pooling)
add_subdirectory(ops/reduce)
add_subdirectory(ops/gemm EXCLUDE_FROM_ALL)
add_subdirectory(ops/gemm_streamk EXCLUDE_FROM_ALL)
add_subdirectory(ops/pooling EXCLUDE_FROM_ALL)
add_subdirectory(ops/reduce EXCLUDE_FROM_ALL)

8
tile_engine/ops/gemm/CMakeLists.txt
View File

@@ -1,7 +1,7 @@
# Copyright (c) Advanced Micro Devices, Inc., or its affiliates.
# SPDX-License-Identifier: MIT
add_subdirectory(gemm_universal)
add_subdirectory(gemm_multi_d)
add_subdirectory(gemm_preshuffle)
add_subdirectory(grouped_gemm)
add_subdirectory(gemm_universal EXCLUDE_FROM_ALL)
add_subdirectory(gemm_multi_d EXCLUDE_FROM_ALL)
add_subdirectory(gemm_preshuffle EXCLUDE_FROM_ALL)
add_subdirectory(grouped_gemm EXCLUDE_FROM_ALL)

36
tile_engine/ops/gemm_streamk/CMakeLists.txt
View File

@@ -124,15 +124,15 @@ function(build_individual_gemm_targets datatype layout)
if(DEFINED ENV{GEMM_STREAMK_CONFIG_FILE} AND NOT "$ENV{GEMM_STREAMK_CONFIG_FILE}" STREQUAL "")
set(config_filename "$ENV{GEMM_STREAMK_CONFIG_FILE}")
set(json_blob "${CMAKE_CURRENT_LIST_DIR}/configs/${config_filename}")
message(STATUS " Using config from environment variable: ${config_filename}")
message(VERBOSE " Using config from environment variable: ${config_filename}")
elseif(NOT "${GEMM_STREAMK_CONFIG_FILE}" STREQUAL "")
# Use CMake variable if set
set(json_blob "${CMAKE_CURRENT_LIST_DIR}/configs/${GEMM_STREAMK_CONFIG_FILE}")
message(STATUS " Using custom config: ${GEMM_STREAMK_CONFIG_FILE}")
message(VERBOSE " Using custom config: ${GEMM_STREAMK_CONFIG_FILE}")
else()
# Use default config for all layouts
set(json_blob "${CMAKE_CURRENT_LIST_DIR}/configs/default_config.json")
message(STATUS " Using default config for layout ${layout}")
message(VERBOSE " Using default config for layout ${layout}")
endif()
# Check if config file exists
@@ -153,17 +153,17 @@ function(build_individual_gemm_targets datatype layout)
endif()
# Generate individual kernel files using parallel version
message(STATUS "Generating individual kernels for ${datatype} ${layout} using ${num_workers} workers...")
message(STATUS " Working path: ${working_path}")
message(STATUS " Config file: ${json_blob}")
message(STATUS " Python executable: ${Python3_EXECUTABLE}")
message(STATUS " Script path: ${CMAKE_CURRENT_LIST_DIR}/gemm_streamk_instance_builder.py")
message(VERBOSE "Generating individual kernels for ${datatype} ${layout} using ${num_workers} workers...")
message(VERBOSE " Working path: ${working_path}")
message(VERBOSE " Config file: ${json_blob}")
message(VERBOSE " Python executable: ${Python3_EXECUTABLE}")
message(VERBOSE " Script path: ${CMAKE_CURRENT_LIST_DIR}/gemm_streamk_instance_builder.py")
# Create working directory first
file(MAKE_DIRECTORY ${working_path})
# First, just list the kernels (fast operation)
message(STATUS " Listing kernel configurations...")
message(VERBOSE " Listing kernel configurations...")
execute_process(
COMMAND ${Python3_EXECUTABLE} -u ${CMAKE_CURRENT_LIST_DIR}/gemm_streamk_instance_builder.py
--working_path ${working_path}
@@ -185,7 +185,7 @@ function(build_individual_gemm_targets datatype layout)
if(EXISTS ${working_path}/gemm_kernel_count.txt)
file(READ ${working_path}/gemm_kernel_count.txt kernel_count)
string(STRIP "${kernel_count}" kernel_count)
message(STATUS " Found ${kernel_count} kernel configurations")
message(VERBOSE " Found ${kernel_count} kernel configurations")
else()
message(FATAL_ERROR "Kernel count file not found")
endif()
@@ -209,10 +209,10 @@ function(build_individual_gemm_targets datatype layout)
endfunction()
# Main build logic - Only individual builds supported
message(STATUS "=== Starting Tile Engine StreamK GEMM Configuration ===")
message(STATUS "GEMM_STREAMK_DATATYPE: ${GEMM_STREAMK_DATATYPE}")
message(STATUS "GEMM_STREAMK_LAYOUT: ${GEMM_STREAMK_LAYOUT}")
message(STATUS "SUPPORTED_GPU_TARGETS: ${SUPPORTED_GPU_TARGETS}")
message(VERBOSE "=== Starting Tile Engine StreamK GEMM Configuration ===")
message(VERBOSE "GEMM_STREAMK_DATATYPE: ${GEMM_STREAMK_DATATYPE}")
message(VERBOSE "GEMM_STREAMK_LAYOUT: ${GEMM_STREAMK_LAYOUT}")
message(VERBOSE "SUPPORTED_GPU_TARGETS: ${SUPPORTED_GPU_TARGETS}")
# Filter GPU targets to only gfx90a, gfx942
set(GEMM_GPU_TARGETS_INDIVIDUAL "")
@@ -221,7 +221,7 @@ set(DESIRED_TARGETS "gfx90a;gfx942") # TODO: Add gfx950 when supported
foreach(target IN LISTS SUPPORTED_GPU_TARGETS)
if(target IN_LIST DESIRED_TARGETS)
list(APPEND GEMM_GPU_TARGETS_INDIVIDUAL ${target})
message(STATUS " Adding GPU target: ${target}")
message(VERBOSE " Adding GPU target: ${target}")
endif()
endforeach()
@@ -229,7 +229,7 @@ endforeach()
if(NOT GEMM_GPU_TARGETS_INDIVIDUAL)
message(WARNING "Skipping Tile Engine GEMM build: No supported GPU targets (gfx90a, gfx942) found in SUPPORTED_GPU_TARGETS: ${SUPPORTED_GPU_TARGETS}")
else()
message(STATUS "Building individual GEMM targets for GPU targets: ${GEMM_GPU_TARGETS_INDIVIDUAL}")
message(VERBOSE "Building individual GEMM targets for GPU targets: ${GEMM_GPU_TARGETS_INDIVIDUAL}")
# Enable parallel compilation optimizations
# Set up job pools for better parallel compilation control
@@ -244,12 +244,12 @@ else()
find_program(CCACHE_PROGRAM ccache)
if(CCACHE_PROGRAM)
set(CMAKE_CXX_COMPILER_LAUNCHER ${CCACHE_PROGRAM})
message(STATUS "Using ccache for faster compilation")
message(VERBOSE "Using ccache for faster compilation")
else()
message(WARNING "ccache requested but not found")
endif()
else()
message(STATUS "ccache disabled for GEMM ops (use -DENABLE_CCACHE_GEMM=ON to enable)")
message(VERBOSE "ccache disabled for GEMM ops (use -DENABLE_CCACHE_GEMM=ON to enable)")
endif()
# Create master collection targets

25
tile_engine/ops/pooling/CMakeLists.txt
View File

@@ -55,6 +55,7 @@ function(create_individual_pool_target datatype kernel_name trait tile_config co
# Create the executable
add_executable(${target_name}
EXCLUDE_FROM_ALL
${target_source}
${instance_header}
)
@@ -90,7 +91,7 @@ function(create_individual_pool_target datatype kernel_name trait tile_config co
add_dependencies(benchmark_pooling_all ${target_name})
add_dependencies(benchmark_pooling_${datatype} ${target_name})
message(STATUS " Created pooling benchmark target: ${target_name}")
message(DEBUG " Created pooling benchmark target: ${target_name}")
endfunction()
# ============================================================================
@@ -105,13 +106,13 @@ function(build_individual_pool_targets datatype)
if(DEFINED ENV{POOLING_CONFIG_FILE} AND NOT "$ENV{POOLING_CONFIG_FILE}" STREQUAL "")
set(config_filename "$ENV{POOLING_CONFIG_FILE}")
set(json_blob "${CMAKE_CURRENT_LIST_DIR}/configs/${config_filename}")
message(STATUS " Using config from environment variable: ${config_filename}")
message(VERBOSE " Using config from environment variable: ${config_filename}")
elseif(NOT "${POOLING_CONFIG_FILE}" STREQUAL "")
set(json_blob "${CMAKE_CURRENT_LIST_DIR}/configs/${POOLING_CONFIG_FILE}")
message(STATUS " Using custom config: ${POOLING_CONFIG_FILE}")
message(VERBOSE " Using custom config: ${POOLING_CONFIG_FILE}")
else()
set(json_blob "${CMAKE_CURRENT_LIST_DIR}/configs/default_config.json")
message(STATUS " Using default config for pooling")
message(VERBOSE " Using default config for pooling")
endif()
if(NOT EXISTS ${json_blob})
@@ -121,7 +122,7 @@ function(build_individual_pool_targets datatype)
file(MAKE_DIRECTORY ${working_path})
# Step 1: List kernels
message(STATUS " Listing pooling kernel configurations for ${datatype}...")
message(VERBOSE " Listing pooling kernel configurations for ${datatype}...")
execute_process(
COMMAND ${Python3_EXECUTABLE} -u ${CMAKE_CURRENT_LIST_DIR}/pooling_instance_builder.py
--working_path ${working_path}
@@ -142,7 +143,7 @@ function(build_individual_pool_targets datatype)
if(EXISTS ${working_path}/pool_kernel_count.txt)
file(READ ${working_path}/pool_kernel_count.txt kernel_count)
string(STRIP "${kernel_count}" kernel_count)
message(STATUS " Found ${kernel_count} pooling kernel configurations")
message(VERBOSE " Found ${kernel_count} pooling kernel configurations")
else()
message(FATAL_ERROR "Pooling kernel count file not found")
endif()
@@ -169,9 +170,9 @@ endfunction()
# MAIN EXECUTION
# ============================================================================
message(STATUS "=== Starting Tile Engine Pooling Configuration ===")
message(STATUS "POOLING_DATATYPE: ${POOLING_DATATYPE}")
message(STATUS "SUPPORTED_GPU_TARGETS: ${SUPPORTED_GPU_TARGETS}")
message(VERBOSE "=== Starting Tile Engine Pooling Configuration ===")
message(VERBOSE "POOLING_DATATYPE: ${POOLING_DATATYPE}")
message(VERBOSE "SUPPORTED_GPU_TARGETS: ${SUPPORTED_GPU_TARGETS}")
# Filter GPU targets
set(POOLING_GPU_TARGETS "")
@@ -180,21 +181,21 @@ set(DESIRED_TARGETS "gfx90a;gfx942;gfx950;gfx1201")
foreach(target IN LISTS SUPPORTED_GPU_TARGETS)
if(target IN_LIST DESIRED_TARGETS)
list(APPEND POOLING_GPU_TARGETS ${target})
message(STATUS " Adding GPU target for pooling: ${target}")
message(VERBOSE " Adding GPU target for pooling: ${target}")
endif()
endforeach()
if(NOT POOLING_GPU_TARGETS)
message(WARNING "Skipping Tile Engine Pooling build: No supported GPU targets (gfx90a, gfx942, gfx950, gfx1201) found in SUPPORTED_GPU_TARGETS: ${SUPPORTED_GPU_TARGETS}")
else()
message(STATUS "Building pooling targets for GPU targets: ${POOLING_GPU_TARGETS}")
message(VERBOSE "Building pooling targets for GPU targets: ${POOLING_GPU_TARGETS}")
# Enable ccache if requested
if(ENABLE_CCACHE_POOLING)
find_program(CCACHE_PROGRAM ccache)
if(CCACHE_PROGRAM)
set(CMAKE_CXX_COMPILER_LAUNCHER ${CCACHE_PROGRAM})
message(STATUS "Using ccache for pooling compilation")
message(VERBOSE "Using ccache for pooling compilation")
endif()
endif()

6
tile_engine/ops/reduce/CMakeLists.txt
View File

@@ -26,7 +26,7 @@ function(build_multi_reduce_for_datatype datatype variant)
return()
endif()
message(STATUS "Building Reduction for GPU targets: ${GPU_TARGETS}")
message(VERBOSE "Building Reduction for GPU targets: ${GPU_TARGETS}")
set(working_path "${CMAKE_CURRENT_BINARY_DIR}/${datatype}/${variant}")
file(MAKE_DIRECTORY "${working_path}")
@@ -75,7 +75,7 @@ function(build_multi_reduce_for_datatype datatype variant)
message(FATAL_ERROR "Failed to generate kernels for ${datatype} ${variant}: ${ret}")
endif()
message(STATUS "Generated ${datatype} ${variant} reduction kernel blobs at: ${working_path}")
message(VERBOSE "Generated ${datatype} ${variant} reduction kernel blobs at: ${working_path}")
# # Add test executables for each generated test
file(STRINGS "${working_path}/reduce_${variant}_blobs_list.txt" test_basenames)
@@ -85,7 +85,7 @@ function(build_multi_reduce_for_datatype datatype variant)
set(test_src "${working_path}/${test_base}.cpp")
set(test_target "${test_base}")
add_executable(${test_target} ${test_src})
add_executable(${test_target} EXCLUDE_FROM_ALL ${test_src})
target_include_directories(${test_target} PRIVATE
"${CMAKE_SOURCE_DIR}/test/ck_tile/reduce/"
${working_path}