CMakeLists.txt 17.7 KB
Newer Older
1
cmake_minimum_required(VERSION 3.14)
JD's avatar
JD committed
2

3
set(version 1.1.0)
JD's avatar
JD committed
4
# Check support for CUDA/HIP in Cmake
5
project(composable_kernel VERSION ${version})
Chao Liu's avatar
Chao Liu committed
6

7
8
list(APPEND CMAKE_MODULE_PATH "${PROJECT_SOURCE_DIR}/cmake")

9
if (DTYPES)
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
    add_definitions(-DDTYPES)
    if (DTYPES MATCHES "int8")
        add_definitions(-DCK_ENABLE_INT8)
        set(CK_ENABLE_INT8 "ON")
    endif()
    if (DTYPES MATCHES "fp8")
        add_definitions(-DCK_ENABLE_FP8)
        set(CK_ENABLE_FP8 "ON")
    endif()
    if (DTYPES MATCHES "fp16")
        add_definitions(-DCK_ENABLE_FP16)
        set(CK_ENABLE_FP16 "ON")
    endif()
    if (DTYPES MATCHES "fp32")
        add_definitions(-DCK_ENABLE_FP32)
        set(CK_ENABLE_FP32 "ON")
    endif()
    if (DTYPES MATCHES "fp64")
        add_definitions(-DCK_ENABLE_FP64)
        set(CK_ENABLE_FP64 "ON")
    endif()
    if (DTYPES MATCHES "bf16")
        add_definitions(-DCK_ENABLE_BF16)
        set(CK_ENABLE_BF16 "ON")
    endif()
    message("DTYPES macro set to ${DTYPES}")
36
else()
37
38
    add_definitions(-DCK_ENABLE_INT8 -DCK_ENABLE_FP8 -DCK_ENABLE_FP16 -DCK_ENABLE_FP32 -DCK_ENABLE_FP64 -DCK_ENABLE_BF16)
    set(CK_ENABLE_ALL_DTYPES "ON")
39
40
endif()

41
42
if(DL_KERNELS)
    add_definitions(-DDL_KERNELS)
43
    set(CK_ENABLE_DL_KERNELS "ON")
44
45
46
47
endif()

if(INSTANCES_ONLY)
    add_definitions(-DINSTANCES_ONLY)
48
    set(CK_ENABLE_INSTANCES_ONLY "ON")
49
50
endif()

51
52
53
54
55
56
57
58
# CK config file to record supported datatypes, etc.
configure_file("${PROJECT_SOURCE_DIR}/include/ck/config.h.in" "${PROJECT_BINARY_DIR}/include/ck/config.h")

# CK version file to record release version as well as git commit hash
find_package(Git REQUIRED)
execute_process(COMMAND "${GIT_EXECUTABLE}" rev-parse HEAD OUTPUT_VARIABLE COMMIT_ID OUTPUT_STRIP_TRAILING_WHITESPACE)
configure_file("${PROJECT_SOURCE_DIR}/include/ck/version.h.in" "${PROJECT_BINARY_DIR}/include/ck/version.h")

JD's avatar
JD committed
59
60
enable_testing()

61
set(ROCM_SYMLINK_LIBS OFF)
Anthony Chang's avatar
Anthony Chang committed
62
find_package(ROCM REQUIRED PATHS /opt/rocm)
JD's avatar
JD committed
63
64
65
66
67
68

include(ROCMInstallTargets)
include(ROCMPackageConfigHelpers)
include(ROCMSetupVersion)
include(ROCMInstallSymlinks)
include(ROCMCreatePackage)
Chao Liu's avatar
Chao Liu committed
69
include(CheckCXXCompilerFlag)
70
include(ROCMCheckTargetIds)
JD's avatar
JD committed
71
include(TargetFlags)
72
73
74

rocm_setup_version(VERSION ${version})

JD's avatar
JD committed
75
76
list(APPEND CMAKE_PREFIX_PATH ${CMAKE_INSTALL_PREFIX} ${CMAKE_INSTALL_PREFIX}/llvm ${CMAKE_INSTALL_PREFIX}/hip /opt/rocm /opt/rocm/llvm /opt/rocm/hip)

77
78
79
80
81
82
message("GPU_TARGETS= ${GPU_TARGETS}")

message("checking which targets are supported")
#This is the list of targets to be used in case GPU_TARGETS is not set on command line
#These targets will be filtered and only supported ones will be used
#Setting GPU_TARGETS on command line will override this list
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97
98
99
100
101
if(NOT PROFILER_ONLY)
    rocm_check_target_ids(DEFAULT_GPU_TARGETS
        TARGETS "gfx900;gfx906;gfx908;gfx90a;gfx940;gfx941;gfx942;gfx1030;gfx1100;gfx1101;gfx1102")
else()
    add_definitions(-DPROFILER_ONLY)
    if(GPU_TARGETS)
        message(FATAL_ERROR "For PROFILE_ONLY build, please do not set GPU_TARGETS, use GPU_ARCH = gfx9, gfx10, or gfx11")
    endif()
    if(GPU_ARCH MATCHES "gfx9")
        rocm_check_target_ids(DEFAULT_GPU_TARGETS TARGETS "gfx900;gfx906;gfx908;gfx90a;gfx940;gfx941;gfx942")
    elseif(GPU_ARCH MATCHES "gfx10")
        rocm_check_target_ids(DEFAULT_GPU_TARGETS TARGETS "gfx1030")
    elseif(GPU_ARCH MATCHES "gfx11")
        rocm_check_target_ids(DEFAULT_GPU_TARGETS TARGETS "gfx1100;gfx1101;gfx1102")
    else()
        message(FATAL_ERROR "For PROFILE_ONLY build, please specify GPU_ARCH as gfx9, gfx10, or gfx11")
    endif()
endif()

102
message("Supported GPU_TARGETS= ${DEFAULT_GPU_TARGETS}")
103

104
set(AMDGPU_TARGETS "${DEFAULT_GPU_TARGETS}" CACHE STRING " ")
105
106
107
108
109
110

if(GPU_TARGETS)
    message("Building CK for the following targets: ${GPU_TARGETS}")
else()
    message("Building CK for the following targets: ${AMDGPU_TARGETS}")
endif()
111
find_package(hip)
112
113
114
115
# No assumption that HIP kernels are launched with uniform block size for backward compatibility
# SWDEV-413293 and https://reviews.llvm.org/D155213
math(EXPR hip_VERSION_FLAT "(${hip_VERSION_MAJOR} * 1000 + ${hip_VERSION_MINOR}) * 100000 + ${hip_VERSION_PATCH}")
message("hip_version_flat=${hip_VERSION_FLAT}")
116
if(${hip_VERSION_FLAT} GREATER 500723302)
117
118
119
   message("Adding the fno-offload-uniform-block compiler flag")
   add_compile_options(-fno-offload-uniform-block)
endif()
120

Adam Osewski's avatar
Adam Osewski committed
121
option(USE_BITINT_EXTENSION_INT4, "Whether to enable clang's BitInt extension to provide int4 data type." OFF)
122
option(USE_OPT_NAVI3X, "Whether to enable LDS cumode and Wavefront32 mode for NAVI3X silicons." OFF)
Adam Osewski's avatar
Adam Osewski committed
123
124
125
126
127
128
129

if(USE_BITINT_EXTENSION_INT4)
    add_compile_definitions(CK_EXPERIMENTAL_BIT_INT_EXTENSION_INT4)
    add_compile_options(-Wno-bit-int-extension)
    message("CK compiled with USE_BITINT_EXTENSION_INT4 set to ${USE_BITINT_EXTENSION_INT4}")
endif()

130
131
132
133
134
135
if(USE_OPT_NAVI3X)
    add_compile_options(-mcumode)
    add_compile_options(-mno-wavefrontsize64)
    message("CK compiled with USE_OPT_NAVI3X set to ${USE_OPT_NAVI3X}")
endif()

136
137
138
139
140
## Threads
set(THREADS_PREFER_PTHREAD_FLAG ON)
find_package(Threads REQUIRED)
link_libraries(Threads::Threads)

141
## C++
Chao Liu's avatar
Chao Liu committed
142
enable_language(CXX)
Chao Liu's avatar
Chao Liu committed
143
set(CMAKE_CXX_STANDARD 17)
Chao Liu's avatar
Chao Liu committed
144
set(CMAKE_CXX_STANDARD_REQUIRED ON)
Chao Liu's avatar
Chao Liu committed
145
set(CMAKE_CXX_EXTENSIONS OFF)
Chao Liu's avatar
Chao Liu committed
146
147
message("CMAKE_CXX_COMPILER_ID: ${CMAKE_CXX_COMPILER_ID}")

Alan Turner's avatar
Alan Turner committed
148
option(CK_BUILD_JIT_LIB, "Only build the CK JIT Helper Library" OFF)
Alan Turner's avatar
Alan Turner committed
149
150
if (NOT CK_BUILD_JIT_LIB)
    option(USE_BITINT_EXTENSION_INT4, "Whether to enable clang's BitInt extension to provide int4 data type." OFF)
Alan Turner's avatar
Alan Turner committed
151
    option(USE_OPT_NAVI3X, "Whether to enable LDS cumode and Wavefront32 mode for NAVI3X silicons." OFF)
Chao Liu's avatar
Chao Liu committed
152

Alan Turner's avatar
Alan Turner committed
153
154
155
156
157
    if(USE_BITINT_EXTENSION_INT4)
        add_compile_definitions(CK_EXPERIMENTAL_BIT_INT_EXTENSION_INT4)
        add_compile_options(-Wno-bit-int-extension)
        message("CK compiled with USE_BITINT_EXTENSION_INT4 set to ${USE_BITINT_EXTENSION_INT4}")
    endif()
Chao Liu's avatar
Chao Liu committed
158

Alan Turner's avatar
Alan Turner committed
159
160
161
162
163
    if(USE_OPT_NAVI3X)
        add_compile_options(-mcumode)
        add_compile_options(-mno-wavefrontsize64)
        message("CK compiled with USE_OPT_NAVI3X set to ${USE_OPT_NAVI3X}")
    endif()
Chao Liu's avatar
Chao Liu committed
164

Alan Turner's avatar
Alan Turner committed
165
166
167
168
169
170
171
172
173
174
175
176
177
178
179
180
181
    ## Threads
    set(THREADS_PREFER_PTHREAD_FLAG ON)
    find_package(Threads REQUIRED)
    link_libraries(Threads::Threads)
    
    ## OpenMP
    if(CMAKE_CXX_COMPILER_ID MATCHES "Clang")
        # workaround issue hipcc in rocm3.5 cannot find openmp
        set(OpenMP_CXX "${CMAKE_CXX_COMPILER}")
        set(OpenMP_CXX_FLAGS "-fopenmp=libomp -Wno-unused-command-line-argument")
        set(OpenMP_CXX_LIB_NAMES "libomp" "libgomp" "libiomp5")
        set(OpenMP_libomp_LIBRARY ${OpenMP_CXX_LIB_NAMES})
        set(OpenMP_libgomp_LIBRARY ${OpenMP_CXX_LIB_NAMES})
        set(OpenMP_libiomp5_LIBRARY ${OpenMP_CXX_LIB_NAMES})
    else()
        find_package(OpenMP REQUIRED)
    endif()
Chao Liu's avatar
Chao Liu committed
182

Alan Turner's avatar
Alan Turner committed
183
184
185
186
187
188
189
190
191
192
193
194
195
196
197
198
199
200
201
202
203
204
205
206
207
208
209
210
211
212
213
    message("OpenMP_CXX_LIB_NAMES: ${OpenMP_CXX_LIB_NAMES}")
    message("OpenMP_gomp_LIBRARY: ${OpenMP_gomp_LIBRARY}")
    message("OpenMP_pthread_LIBRARY: ${OpenMP_pthread_LIBRARY}")
    message("OpenMP_CXX_FLAGS: ${OpenMP_CXX_FLAGS}")

    link_libraries(${OpenMP_gomp_LIBRARY})
    link_libraries(${OpenMP_pthread_LIBRARY})

    ## HIP
    find_package(HIP REQUIRED)
    # Override HIP version in config.h, if necessary.
    # The variables set by find_package() can't be overwritten,
    # therefore let's use intermediate variables.
    set(CK_HIP_VERSION_MAJOR "${HIP_VERSION_MAJOR}")
    set(CK_HIP_VERSION_MINOR "${HIP_VERSION_MINOR}")
    set(CK_HIP_VERSION_PATCH "${HIP_VERSION_PATCH}")
    if( DEFINED CK_OVERRIDE_HIP_VERSION_MAJOR )
        set(CK_HIP_VERSION_MAJOR "${CK_OVERRIDE_HIP_VERSION_MAJOR}")
        message(STATUS "CK_HIP_VERSION_MAJOR overriden with ${CK_OVERRIDE_HIP_VERSION_MAJOR}")
    endif()
    if( DEFINED CK_OVERRIDE_HIP_VERSION_MINOR )
        set(CK_HIP_VERSION_MINOR "${CK_OVERRIDE_HIP_VERSION_MINOR}")
        message(STATUS "CK_HIP_VERSION_MINOR overriden with ${CK_OVERRIDE_HIP_VERSION_MINOR}")
    endif()
    if( DEFINED CK_OVERRIDE_HIP_VERSION_PATCH )
        set(CK_HIP_VERSION_PATCH "${CK_OVERRIDE_HIP_VERSION_PATCH}")
        message(STATUS "CK_HIP_VERSION_PATCH overriden with ${CK_OVERRIDE_HIP_VERSION_PATCH}")
    endif()
    message(STATUS "Build with HIP ${HIP_VERSION}")
    link_libraries(hip::device)
    add_compile_definitions(__HIP_PLATFORM_HCC__=1)
JD's avatar
JD committed
214
endif()
215

Chao Liu's avatar
Chao Liu committed
216
217
## tidy
include(EnableCompilerWarnings)
JD's avatar
JD committed
218
set(CK_TIDY_ERRORS ERRORS * -readability-inconsistent-declaration-parameter-name)
Chao Liu's avatar
Chao Liu committed
219
if(CMAKE_CXX_COMPILER MATCHES ".*hcc" OR CMAKE_CXX_COMPILER MATCHES ".*clang\\+\\+")
JD's avatar
JD committed
220
    set(CK_TIDY_CHECKS -modernize-use-override -readability-non-const-parameter)
Chao Liu's avatar
Chao Liu committed
221
# Enable tidy on hip
JD's avatar
JD committed
222
223
elseif(CK_BACKEND STREQUAL "HIP" OR CK_BACKEND STREQUAL "HIPNOGPU")
    set(CK_TIDY_ERRORS ALL)
Chao Liu's avatar
Chao Liu committed
224
225
endif()

JD's avatar
JD committed
226

Chao Liu's avatar
Chao Liu committed
227
228
229
230
231
232
233
234
235
236
237
238
239
240
241
242
243
244
245
246
247
248
249
250
251
252
253
254
255
256
257
258
259
260
261
262
263
264
265
266
267
268
269
270
271
272
273
274
275
276
277
278
279
280
281
282
283
284
285
286
287
288
289
290
291
292
293
294
295
296
297
298
299
300
301
302
303
304
305
306
307
308
309
310
311
312
313
314
315
316
317
include(ClangTidy)
enable_clang_tidy(
    CHECKS
        *
        -abseil-*
        -android-cloexec-fopen
        # Yea we shouldn't be using rand()
        -cert-msc30-c
        -bugprone-exception-escape
        -bugprone-macro-parentheses
        -cert-env33-c
        -cert-msc32-c
        -cert-msc50-cpp
        -cert-msc51-cpp
        -cert-dcl37-c
        -cert-dcl51-cpp
        -clang-analyzer-alpha.core.CastToStruct
        -clang-analyzer-optin.performance.Padding
        -clang-diagnostic-deprecated-declarations
        -clang-diagnostic-extern-c-compat
        -clang-diagnostic-unused-command-line-argument
        -cppcoreguidelines-avoid-c-arrays
        -cppcoreguidelines-avoid-magic-numbers
        -cppcoreguidelines-explicit-virtual-functions
        -cppcoreguidelines-init-variables
        -cppcoreguidelines-macro-usage
        -cppcoreguidelines-non-private-member-variables-in-classes
        -cppcoreguidelines-pro-bounds-array-to-pointer-decay
        -cppcoreguidelines-pro-bounds-constant-array-index
        -cppcoreguidelines-pro-bounds-pointer-arithmetic
        -cppcoreguidelines-pro-type-member-init
        -cppcoreguidelines-pro-type-reinterpret-cast
        -cppcoreguidelines-pro-type-union-access
        -cppcoreguidelines-pro-type-vararg
        -cppcoreguidelines-special-member-functions
        -fuchsia-*
        -google-explicit-constructor
        -google-readability-braces-around-statements
        -google-readability-todo
        -google-runtime-int
        -google-runtime-references
        -hicpp-vararg
        -hicpp-braces-around-statements
        -hicpp-explicit-conversions
        -hicpp-named-parameter
        -hicpp-no-array-decay
        # We really shouldn't use bitwise operators with signed integers, but
        # opencl leaves us no choice
        -hicpp-avoid-c-arrays
        -hicpp-signed-bitwise
        -hicpp-special-member-functions
        -hicpp-uppercase-literal-suffix
        -hicpp-use-auto
        -hicpp-use-equals-default
        -hicpp-use-override
        -llvm-header-guard
        -llvm-include-order
        #-llvmlibc-*
        -llvmlibc-restrict-system-libc-headers
        -llvmlibc-callee-namespace
        -llvmlibc-implementation-in-namespace
        -llvm-else-after-return
        -llvm-qualified-auto
        -misc-misplaced-const
        -misc-non-private-member-variables-in-classes
        -misc-no-recursion
        -modernize-avoid-bind
        -modernize-avoid-c-arrays
        -modernize-pass-by-value
        -modernize-use-auto
        -modernize-use-default-member-init
        -modernize-use-equals-default
        -modernize-use-trailing-return-type
        -modernize-use-transparent-functors
        -performance-unnecessary-value-param
        -readability-braces-around-statements
        -readability-else-after-return
        # we are not ready to use it, but very useful
        -readability-function-cognitive-complexity
        -readability-isolate-declaration
        -readability-magic-numbers
        -readability-named-parameter
        -readability-uppercase-literal-suffix
        -readability-convert-member-functions-to-static
        -readability-qualified-auto
        -readability-redundant-string-init
        # too many narrowing conversions in our code
        -bugprone-narrowing-conversions
        -cppcoreguidelines-narrowing-conversions
        -altera-struct-pack-align
        -cppcoreguidelines-prefer-member-initializer
JD's avatar
JD committed
318
319
        ${CK_TIDY_CHECKS}
        ${CK_TIDY_ERRORS}
Chao Liu's avatar
Chao Liu committed
320
321
322
    HEADER_FILTER
        "\.hpp$"
    EXTRA_ARGS
JD's avatar
JD committed
323
        -DCK_USE_CLANG_TIDY
Chao Liu's avatar
Chao Liu committed
324
325
)

Alan Turner's avatar
Alan Turner committed
326

Chao Liu's avatar
Chao Liu committed
327
328
329
330
331
332
333
334
335
336
337
338
339
include(CppCheck)
enable_cppcheck(
    CHECKS
        warning
        style
        performance
        portability
    SUPPRESS
        ConfigurationNotChecked
        constStatement
        duplicateCondition
        noExplicitConstructor
        passedByValue
Chao Liu's avatar
Chao Liu committed
340
        preprocessorErrorDirective
Chao Liu's avatar
Chao Liu committed
341
342
343
344
345
346
347
        shadowVariable
        unusedFunction
        unusedPrivateFunction
        unusedStructMember
        unmatchedSuppression
    FORCE
    SOURCES
Chao Liu's avatar
Chao Liu committed
348
        library/src
Chao Liu's avatar
Chao Liu committed
349
350
351
    INCLUDE
        ${CMAKE_CURRENT_SOURCE_DIR}/include
        ${CMAKE_CURRENT_BINARY_DIR}/include
Chao Liu's avatar
Chao Liu committed
352
        ${CMAKE_CURRENT_SOURCE_DIR}/library/include
Chao Liu's avatar
Chao Liu committed
353
354
355
356
    DEFINE
        CPPCHECK=1
        __linux__=1
)
Chao Liu's avatar
Chao Liu committed
357

JD's avatar
JD committed
358
359
360
361
set(CMAKE_LIBRARY_OUTPUT_DIRECTORY ${CMAKE_CURRENT_BINARY_DIR}/lib)
set(CMAKE_ARCHIVE_OUTPUT_DIRECTORY ${CMAKE_CURRENT_BINARY_DIR}/lib)
set(CMAKE_RUNTIME_OUTPUT_DIRECTORY ${CMAKE_CURRENT_BINARY_DIR}/bin)

362
# set CK project include directories
Chao Liu's avatar
Chao Liu committed
363
include_directories(BEFORE
364
    ${PROJECT_BINARY_DIR}/include
Chao Liu's avatar
Chao Liu committed
365
366
    ${PROJECT_SOURCE_DIR}/include
    ${PROJECT_SOURCE_DIR}/library/include
367
    ${HIP_INCLUDE_DIRS}
JD's avatar
JD committed
368
)
Chao Liu's avatar
Chao Liu committed
369

JD's avatar
JD committed
370
371
SET(BUILD_DEV ON CACHE BOOL "BUILD_DEV")
if(BUILD_DEV)
Chao Liu's avatar
Chao Liu committed
372
373
    add_compile_options(-Werror)
    add_compile_options(-Weverything)
JD's avatar
JD committed
374
375
376
endif()
message("CMAKE_CXX_FLAGS: ${CMAKE_CXX_FLAGS}")

Alan Turner's avatar
Alan Turner committed
377
378
add_custom_target(check COMMAND ${CMAKE_CTEST_COMMAND} --output-on-failure -C ${CMAKE_CFG_INTDIR})

Alan Turner's avatar
Alan Turner committed
379
380
381
382
383
if (NOT CK_BUILD_JIT_LIB)
    file(GLOB_RECURSE INSTANCE_FILES "${PROJECT_SOURCE_DIR}/*/device_*_instance.cpp")
    file(GLOB dir_list RELATIVE ${PROJECT_SOURCE_DIR}/library/src/tensor_operation_instance/gpu ${PROJECT_SOURCE_DIR}/library/src/tensor_operation_instance/gpu/*)
    set(CK_DEVICE_INSTANCES)
    FOREACH(subdir_path ${dir_list})
384
385
386
387
388
389
390
391
392
393
394
395
396
397
398
399
400
401
402
403
404
405
406
407
408
409
410
411
412
413
414
415
416
417
418
419
420
    set(target_dir)
    IF(IS_DIRECTORY "${PROJECT_SOURCE_DIR}/library/src/tensor_operation_instance/gpu/${subdir_path}")
        set(cmake_instance)
        file(READ "${PROJECT_SOURCE_DIR}/library/src/tensor_operation_instance/gpu/${subdir_path}/CMakeLists.txt" cmake_instance)
        set(add_inst 0)
        if("${cmake_instance}" MATCHES "DTYPES MATCHES \"fp8\" " AND DTYPES MATCHES "fp8")
            #message("fp8 instance found!")
            set(add_inst 1)
        endif()
        if("${cmake_instance}" MATCHES "DTYPES MATCHES \"fp16\"" AND DTYPES MATCHES "fp16")
            #message("fp16 instance found!")
            set(add_inst 1)
        endif()
        if("${cmake_instance}" MATCHES "DTYPES MATCHES \"fp32\"" AND DTYPES MATCHES "fp32")
            #message("fp32 instance found!")
            set(add_inst 1)
        endif()
        if("${cmake_instance}" MATCHES "DTYPES MATCHES \"fp64\"" AND DTYPES MATCHES "fp64")
            #message("fp64 instance found!")
            set(add_inst 1)
        endif()
        if("${cmake_instance}" MATCHES "DTYPES MATCHES \"bf16\"" AND DTYPES MATCHES "bf16")
            #message("bf16 instance found!")
            set(add_inst 1)
        endif()
        if("${cmake_instance}" MATCHES "DTYPES MATCHES \"int8\"" AND DTYPES MATCHES "int8")
            #message("int8 instance found!")
            set(add_inst 1)
        endif()
        if(NOT "${cmake_instance}" MATCHES "DTYPES")
            #message("instance should be built for all types!")
            set(add_inst 1)
        endif()
        if(add_inst EQUAL 1 OR NOT DEFINED DTYPES)
            list(APPEND CK_DEVICE_INSTANCES device_${subdir_path}_instance)
        endif()
    ENDIF()
Alan Turner's avatar
Alan Turner committed
421
    ENDFOREACH()
422

Alan Turner's avatar
Alan Turner committed
423
    add_custom_target(instances DEPENDS utility;${CK_DEVICE_INSTANCES}  SOURCES ${INSTANCE_FILES})
424
    add_subdirectory(library)
Anthony Chang's avatar
Anthony Chang committed
425

426
427
    if(NOT DEFINED INSTANCES_ONLY)
    if(NOT DEFINED PROFILER_ONLY)
Alan Turner's avatar
Alan Turner committed
428
429
430
431
    rocm_package_setup_component(tests
            LIBRARY_NAME composablekernel
            PACKAGE_NAME tests # Prevent -static suffix on package name
    )
432

Alan Turner's avatar
Alan Turner committed
433
434
435
436
    rocm_package_setup_component(examples
            LIBRARY_NAME composablekernel
            PACKAGE_NAME examples
    )
437
438
    add_subdirectory(example)
    add_subdirectory(test)
439

Alan Turner's avatar
Alan Turner committed
440
441
442
443
444
    rocm_package_setup_component(profiler
            LIBRARY_NAME composablekernel
            PACKAGE_NAME ckProfiler
    )
    add_subdirectory(profiler)
445
446
447
448
449
450
451
452
453
    else()
        #When building PROFILER_ONLY, label the package with GPU_ARCH
        rocm_package_setup_component(profiler
        LIBRARY_NAME composablekernel
        PACKAGE_NAME ckProfiler_${GPU_ARCH}
        )
        add_subdirectory(profiler)
    endif()
    endif()
454
else()
Alan Turner's avatar
Alan Turner committed
455
    rocm_package_setup_component(jit_library
456
457
        LIBRARY_NAME composablekernel
        PACKAGE_NAME jit_library
Alan Turner's avatar
Alan Turner committed
458
    )
459
460
    add_subdirectory(library)
    add_subdirectory(test)
Alan Turner's avatar
Alan Turner committed
461
endif()
JD's avatar
JD committed
462
463
464
465
466
467
468
469
470
471

#Create an interface target for the include only files and call it "composablekernels"
include(CMakePackageConfigHelpers)

write_basic_package_version_file(
    "${CMAKE_CURRENT_BINARY_DIR}/composable_kernelConfigVersion.cmake"
    VERSION "${version}"
    COMPATIBILITY AnyNewerVersion
)

Anthony Chang's avatar
Anthony Chang committed
472
configure_package_config_file(${CMAKE_CURRENT_SOURCE_DIR}/Config.cmake.in
473
474
475
    "${CMAKE_CURRENT_BINARY_DIR}/composable_kernelConfig.cmake"
    INSTALL_DESTINATION ${CMAKE_INSTALL_LIBDIR}/cmake/composable_kernel
    NO_CHECK_REQUIRED_COMPONENTS_MACRO
JD's avatar
JD committed
476
477
)

478
rocm_install(FILES
JD's avatar
JD committed
479
480
    "${CMAKE_CURRENT_BINARY_DIR}/composable_kernelConfig.cmake"
    "${CMAKE_CURRENT_BINARY_DIR}/composable_kernelConfigVersion.cmake"
Anthony Chang's avatar
Anthony Chang committed
481
    DESTINATION ${CMAKE_INSTALL_LIBDIR}/cmake/composable_kernel
JD's avatar
JD committed
482
)
483

484
485
486
487
488
489
490
# Install CK version and configuration files
install(FILES
    ${PROJECT_BINARY_DIR}/include/ck/version.h
    ${PROJECT_BINARY_DIR}/include/ck/config.h
    DESTINATION ${CMAKE_INSTALL_INCLUDEDIR}/ck/
)

491
492
493
494
495
496
497
498
499
500
set(CPACK_RESOURCE_FILE_LICENSE "${CMAKE_CURRENT_SOURCE_DIR}/LICENSE")
set(CPACK_RPM_PACKAGE_LICENSE "MIT")

rocm_create_package(
    NAME composablekernel
    DESCRIPTION "High Performance Composable Kernel for AMD GPUs"
    MAINTAINER "MIOpen Kernels Dev Team <dl.MIOpen@amd.com>"
    LDCONFIG
    HEADER_ONLY
)