Skip to content

Enable building on Windows with necessary modifications #2167

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Open
amd-mtrifuno wants to merge 4 commits into
base: develop
Choose a base branch
from
Open

Enable building on Windows with necessary modifications #2167

Show file tree
Hide file tree
Changes from 1 commit
Commits
Show all changes
4 commits
Select commit Hold shift + click to select a range
c21cde0
Enable building on Windows with necessary modifications
amd-mtrifuno Jun 3, 2025
88ee306
Add flush after write in clang_args file
amd-mtrifuno Jun 16, 2025
ceec597
Fix copying *.co files
amd-mtrifuno Jun 19, 2025
02ba28b
Add print messages
amd-mtrifuno Jun 19, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
38 changes: 29 additions & 9 deletions CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -116,9 +116,13 @@ option(Tensile_SEPARATE_ARCHITECTURES "Tensile to use GPU architecture specific
option(Tensile_NO_LAZY_LIBRARY_LOADING "Diasble loading kernels on demand?" OFF)
# For roctx
include(CMakeDependentOption)
cmake_dependent_option(HIPBLASLT_ENABLE_MARKER "Enable roctx marker in hipBLASLt" ON "BUILD_SHARED_LIBS" OFF)
# For rocRoller
option(HIPBLASLT_USE_ROCROLLER "Build with the rocRoller library" ON)
if(WIN32)
cmake_dependent_option(HIPBLASLT_ENABLE_MARKER "Disable roctx marker in hipBLASLt - roctracer does not support on Windows" OFF "BUILD_SHARED_LIBS" OFF)
option(USE_ROCROLLER "Build without the rocRoller library" OFF)
else()
cmake_dependent_option(HIPBLASLT_ENABLE_MARKER "Enable roctx marker in hipBLASLt" ON "BUILD_SHARED_LIBS" OFF)
option(USE_ROCROLLER "Build with the rocRoller library" ON)
endif()

if(BUILD_CODE_COVERAGE)
add_compile_options(-fprofile-arcs -ftest-coverage)
Expand Down Expand Up @@ -266,10 +270,12 @@ endif()
cmake_host_system_information(RESULT OS_PLATFORM QUERY DISTRIB_ID)
cmake_host_system_information(RESULT OS_INFO QUERY DISTRIB_INFO)

if( ${OS_PLATFORM} STREQUAL "rhel")
if( ${OS_INFO_VERSION_ID} VERSION_LESS "9.5")
message(WARNING "RHEL version ${OS_INFO_VERSION_ID} too old, not building RocRoller")
set(HIPBLASLT_USE_ROCROLLER OFF)
if(NOT WIN32)
if( ${OS_PLATFORM} STREQUAL "rhel")
if( ${OS_INFO_VERSION_ID} VERSION_LESS "9.5")
message(WARNING "RHEL version ${OS_INFO_VERSION_ID} too old, not building RocRoller")
set(USE_ROCROLLER OFF)
endif()
endif()
endif()

Expand Down Expand Up @@ -378,12 +384,26 @@ if(BUILD_DOCS)
add_subdirectory(docs)
endif()

if( WIN32 )
set( CPACK_SOURCE_GENERATOR "ZIP" )
set( CPACK_GENERATOR "ZIP" )
endif( )

# Package specific CPACK vars
set( CPACK_RESOURCE_FILE_LICENSE "${CMAKE_CURRENT_SOURCE_DIR}/LICENSE.md" )
set( CPACK_RPM_PACKAGE_LICENSE "MIT")

if( NOT CPACK_PACKAGING_INSTALL_PREFIX )
set( CPACK_PACKAGING_INSTALL_PREFIX "${CMAKE_INSTALL_PREFIX}" )
if (WIN32)
SET( CMAKE_INSTALL_PREFIX "C:/hipSDK" CACHE PATH "Install path" FORCE )
SET( INSTALL_PREFIX "C:/hipSDK" )
SET( CPACK_SET_DESTDIR FALSE )
SET( CPACK_PACKAGE_INSTALL_DIRECTORY "C:/hipSDK" )
SET( CPACK_PACKAGING_INSTALL_PREFIX "" )
set( CPACK_INCLUDE_TOPLEVEL_DIRECTORY OFF )
else()
if( NOT CPACK_PACKAGING_INSTALL_PREFIX )
set( CPACK_PACKAGING_INSTALL_PREFIX "${CMAKE_INSTALL_PREFIX}" )
endif()
endif()

set( CPACK_RPM_EXCLUDE_FROM_AUTO_FILELIST_ADDITION "\${CPACK_PACKAGING_INSTALL_PREFIX}" "\${CPACK_PACKAGING_INSTALL_PREFIX}/include" "\${CPACK_PACKAGING_INSTALL_PREFIX}/lib" )
Expand Down
23 changes: 18 additions & 5 deletions clients/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -87,12 +87,25 @@ endif( )
if( BUILD_CLIENTS_BENCHMARKS OR BUILD_CLIENTS_TESTS)

# Linking lapack library requires fortran flags
find_package( cblas REQUIRED CONFIG )
if(${BLIS_FOUND})
set( BLAS_LIBRARY ${BLIS_LIB} )
set( BLIS_CPP ../common/blis_interface.cpp )
if ( NOT WIN32 )
find_package( cblas REQUIRED CONFIG )
if(${BLIS_FOUND})
set( BLAS_LIBRARY ${BLIS_LIB} )
set( BLIS_CPP ../common/blis_interface.cpp )
else()
set( BLAS_LIBRARY "blas" )
endif()
else()
set( BLAS_LIBRARY "blas" )
set( BLAS_INCLUDE_DIR ${OPENBLAS_DIR}/include CACHE PATH "OpenBLAS library include path" )
find_library( BLAS_LIBRARY libopenblas
PATHS ${OPENBLAS_DIR}/lib
NO_DEFAULT_PATH
)
if (NOT BLAS_LIBRARY)
find_package( OPENBLAS CONFIG REQUIRED )
set( BLAS_LIBRARY OpenBLAS::OpenBLAS )
set( BLAS_INCLUDE_DIR "" )
endif()
endif()

# Find the package ROCmSMI
Expand Down
2 changes: 2 additions & 0 deletions clients/benchmarks/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -92,6 +92,8 @@ if (NOT WIN32)
list( APPEND COMMON_LINK_LIBS "-lflang -lflangrti") # for lapack
endif()
else()
find_package(lapack REQUIRED)
target_link_libraries(hipblaslt-bench PRIVATE ${LAPACK_LIBRARIES})
list( APPEND COMMON_LINK_LIBS "libomp")
endif()
target_link_libraries( hipblaslt-bench PRIVATE ${COMMON_LINK_LIBS} )
Expand Down
53 changes: 35 additions & 18 deletions clients/gtest/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -67,6 +67,11 @@ target_include_directories( hipblaslt-test
message("BLIS_INCLUDE_DIR=" ${BLIS_INCLUDE_DIR})
target_link_libraries( hipblaslt-test PRIVATE ${BLAS_LIBRARY} ${GTEST_BOTH_LIBRARIES} roc::hipblaslt )

if(WIN32)
find_package(lapack REQUIRED)
target_link_libraries( hipblaslt-test PRIVATE ${LAPACK_LIBRARIES})
endif()

if( NOT BUILD_CUDA )
target_link_libraries( hipblaslt-test PRIVATE hip::host hip::device )
else()
Expand All @@ -89,32 +94,44 @@ target_compile_definitions( hipblaslt-test PRIVATE ROCM_USE_FLOAT16 HIPBLASLT_IN
target_compile_options(hipblaslt-test PRIVATE $<$<COMPILE_LANGUAGE:CXX>:${COMMON_CXX_OPTIONS}>)
# target_compile_options does not go to linker like CMAKE_CXX_FLAGS does, so manually add

if (BUILD_CUDA)
target_link_libraries( hipblaslt-test PRIVATE -llapack -lcblas )
else()
target_link_libraries( hipblaslt-test PRIVATE lapack cblas )
endif()
if( NOT WIN32)
if (BUILD_CUDA)
target_link_libraries( hipblaslt-test PRIVATE -llapack -lcblas )
else()
target_link_libraries( hipblaslt-test PRIVATE lapack cblas )
endif()

list( APPEND COMMON_LINK_LIBS "-lm -lstdc++fs")
list( APPEND COMMON_LINK_LIBS "-lm -lstdc++fs")

if (CMAKE_Fortran_COMPILER_ID MATCHES "GNU")
list( APPEND COMMON_LINK_LIBS "-lgfortran") # for lapack
if (CMAKE_Fortran_COMPILER_ID MATCHES "GNU")
list( APPEND COMMON_LINK_LIBS "-lgfortran") # for lapack
else()
list( APPEND COMMON_LINK_LIBS "-lflang -lflangrti") # for lapack
endif()
else()
list( APPEND COMMON_LINK_LIBS "-lflang -lflangrti") # for lapack
list( APPEND COMMON_LINK_LIBS "libomp")
endif()

#if (NOT WIN32)
# target_link_libraries( hipblaslt-test PRIVATE lapack cblas )
# list( APPEND COMMON_LINK_LIBS "-lm -lstdc++fs")
# if (NOT BUILD_FORTRAN_CLIENTS)
# list( APPEND COMMON_LINK_LIBS "-lgfortran") # for lapack
# endif()
#else()
# list( APPEND COMMON_LINK_LIBS "libomp")
#endif()

target_link_libraries( hipblaslt-test PRIVATE ${COMMON_LINK_LIBS} )

if (WIN32)

@stellaraccident stellaraccident Jun 3, 2025

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Sorry, we can't land changes like this anymore. It is never appropriate to be copying files from system32 or using a platform gate to move files around by absolute path. Getting these things lined up is a job for whatever is building this project, not just something to be done unconditionally here.

# for now adding in all .dll as dependency chain is not cmake based on win32
file( GLOB third_party_dlls
LIST_DIRECTORIES OFF
CONFIGURE_DEPENDS
${OPENBLAS_DIR}/bin/*.dll
${HIP_DIR}/bin/amd*.dll
${HIP_DIR}/bin/hiprt*.dll
${HIP_DIR}/bin/hipinfo.exe
${CMAKE_SOURCE_DIR}/rtest.*
C:/Windows/System32/libomp140*.dll
)
foreach( file_i ${third_party_dlls})
add_custom_command( TARGET hipblaslt-test POST_BUILD COMMAND ${CMAKE_COMMAND} ARGS -E copy ${file_i} ${PROJECT_BINARY_DIR}/staging/ )
endforeach( file_i )
endif()

set_target_properties( hipblaslt-test PROPERTIES
IMPORT_PREFIX ""
IMPORT_SUFFIX ".lib"
Expand Down
9 changes: 8 additions & 1 deletion cmake/virtualenv.cmake
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -8,7 +8,6 @@ find_package(Python REQUIRED COMPONENTS Interpreter)

set(VIRTUALENV_PYTHON_EXE ${Python_EXECUTABLE})

get_filename_component(VIRTUALENV_PYTHON_EXENAME ${VIRTUALENV_PYTHON_EXE} NAME CACHE)

set(VIRTUALENV_HOME_DIR ${CMAKE_BINARY_DIR}/virtualenv CACHE PATH "Path to virtual environment")

Expand All @@ -23,6 +22,14 @@ function(virtualenv_create)
else()
set(VIRTUALENV_BIN_DIR ${VIRTUALENV_HOME_DIR}/bin CACHE PATH "Path to virtualenv bin directory")
endif()

# verify python executable name inside virtualenv as may be python3 or python (even if installed by python3)
find_program(VIRTUALENV_INST_PYTHON_EXE python3 PATHS ${VIRTUALENV_BIN_DIR} NO_DEFAULT_PATH)
if(NOT VIRTUALENV_INST_PYTHON_EXE)
find_program(VIRTUALENV_INST_PYTHON_EXE python PATHS ${VIRTUALENV_BIN_DIR} NO_DEFAULT_PATH)
endif()

get_filename_component(VIRTUALENV_PYTHON_EXENAME ${VIRTUALENV_INST_PYTHON_EXE} NAME CACHE)
endfunction()

function(virtualenv_install)
Expand Down
22 changes: 21 additions & 1 deletion library/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -78,6 +78,22 @@ endif()

# Target compile definitions
if(NOT BUILD_CUDA)
if (WIN32)

@stellaraccident stellaraccident Jun 3, 2025

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

No - I don't see why this is win32 guarded.

find_package(msgpack-cxx REQUIRED)

target_compile_definitions(hipblaslt PUBLIC -DTENSILE_MSGPACK=1)
get_target_property(msgpack_inc msgpack-cxx INTERFACE_INCLUDE_DIRECTORIES)

if(DEFINED msgpack_inc)
# include C++ headers manually
# External header includes included as system files
target_include_directories(hipblaslt
SYSTEM PRIVATE $<BUILD_INTERFACE:${msgpack_inc}>
)

target_link_libraries(hipblaslt PRIVATE TensileHost shlwapi)
endif()
endif()

if( BUILD_SHARED_LIBS )
target_link_libraries( hipblaslt PRIVATE TensileHost )
Expand Down Expand Up @@ -207,6 +223,10 @@ set_target_properties(hipblaslt PROPERTIES CXX_VISIBILITY_PRESET "hidden" VISIBI
set_target_properties(hipblaslt PROPERTIES RUNTIME_OUTPUT_DIRECTORY "${PROJECT_BINARY_DIR}/staging")
set_target_propertieS(hipblaslt PROPERTIES DEBUG_POSTFIX "-d")

if (WIN32 AND BUILD_CLIENTS)

@stellaraccident stellaraccident Jun 3, 2025

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Cannot land things like this anymore. This is not win32 generally but related to the expectations of your build system.

add_custom_command( TARGET hipblaslt POST_BUILD COMMAND ${CMAKE_COMMAND} -E copy ${PROJECT_BINARY_DIR}/staging/$<TARGET_FILE_NAME:hipblaslt> ${PROJECT_BINARY_DIR}/clients/staging/$<TARGET_FILE_NAME:hipblaslt> )
endif()

# TODO ??
# Following boost conventions of prefixing 'lib' on static built libraries
if(NOT BUILD_SHARED_LIBS)
Expand Down Expand Up @@ -239,7 +259,7 @@ install(

if ( NOT BUILD_CUDA )
if (WIN32)
set( HIPBLASLT_TENSILE_LIBRARY_DIR "\${CPACK_PACKAGING_INSTALL_PREFIX}hipblaslt/bin" CACHE PATH "path to tensile library" )
set( HIPBLASLT_TENSILE_LIBRARY_DIR "\${CPACK_PACKAGING_INSTALL_PREFIX}/bin/hipblaslt" CACHE PATH "path to tensile library" )

@stellaraccident stellaraccident Jun 3, 2025

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Fyi - we're aligning these libraries to ask be based on the posix paths. The hipBLASLt host code has already been updated to use lib as the location for all platforms.

else()
set( HIPBLASLT_TENSILE_LIBRARY_DIR "\${CPACK_PACKAGING_INSTALL_PREFIX}${CMAKE_INSTALL_LIBDIR}/hipblaslt" CACHE PATH "path to tensile library" )
endif()
Expand Down
62 changes: 36 additions & 26 deletions library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -80,17 +80,17 @@ foreach(arch IN LISTS archs)
${CMAKE_CURRENT_BINARY_DIR}/A_H_H_256_4_${arch}.s
${CMAKE_CURRENT_BINARY_DIR}/A_H_S_256_4_${arch}.s
${CMAKE_CURRENT_BINARY_DIR}/A_S_H_256_4_${arch}.s
COMMAND ${python_launch_prefix} LayerNormGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/L_256_4_1_${arch}.s -w 256 -c 4 --sweep-once 1 --arch ${arch}
COMMAND ${python_launch_prefix} LayerNormGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/L_256_4_0_${arch}.s -w 256 -c 4 --sweep-once 0 --arch ${arch}
COMMAND ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_8_32_${arch}.s -m 8 -n 32 --arch ${arch}
COMMAND ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_16_16_${arch}.s -m 16 -n 16 --arch ${arch}
COMMAND ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_4_64_${arch}.s -m 4 -n 64 --arch ${arch}
COMMAND ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_2_128_${arch}.s -m 2 -n 128 --arch ${arch}
COMMAND ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_1_256_${arch}.s -m 1 -n 256 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_256_4_${arch}.s -t S -d S -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/A_H_H_256_4_${arch}.s -t H -d H -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/A_H_S_256_4_${arch}.s -t H -d S -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_256_4_${arch}.s -t S -d H -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} LayerNormGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/L_256_4_1_${arch}.s -w 256 -c 4 --sweep-once 1 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} LayerNormGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/L_256_4_0_${arch}.s -w 256 -c 4 --sweep-once 0 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_8_32_${arch}.s -m 8 -n 32 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_16_16_${arch}.s -m 16 -n 16 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_4_64_${arch}.s -m 4 -n 64 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_2_128_${arch}.s -m 2 -n 128 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} SoftmaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/S_1_256_${arch}.s -m 1 -n 256 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_256_4_${arch}.s -t S -d S -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/A_H_H_256_4_${arch}.s -t H -d H -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/A_H_S_256_4_${arch}.s -t H -d S -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_256_4_${arch}.s -t S -d H -w 256 -c 4 --arch ${arch}
COMMENT "Creating Layer Norm, Softmax and Amax Assembly for ${arch}"
WORKING_DIRECTORY ${ops_path}
)
Expand All @@ -101,10 +101,10 @@ foreach(arch IN LISTS archs)
${CMAKE_CURRENT_BINARY_DIR}/A_S_S_B8N_256_4_${arch}.s
${CMAKE_CURRENT_BINARY_DIR}/A_S_H_F8N_256_4_${arch}.s
${CMAKE_CURRENT_BINARY_DIR}/A_S_H_B8N_256_4_${arch}.s
COMMAND ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_F8N_256_4_${arch}.s -t S -d S -s F8N -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_B8N_256_4_${arch}.s -t S -d S -s B8N -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_F8N_256_4_${arch}.s -t S -d H -s F8N -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_B8N_256_4_${arch}.s -t S -d H -s B8N -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_F8N_256_4_${arch}.s -t S -d S -s F8N -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_B8N_256_4_${arch}.s -t S -d S -s B8N -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_F8N_256_4_${arch}.s -t S -d H -s F8N -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_B8N_256_4_${arch}.s -t S -d H -s B8N -w 256 -c 4 --arch ${arch}
COMMENT "Creating Extra Amax Assembly for gfx942"
WORKING_DIRECTORY ${ops_path}
)
Expand All @@ -123,10 +123,10 @@ foreach(arch IN LISTS archs)
${CMAKE_CURRENT_BINARY_DIR}/A_S_S_B8_256_4_${arch}.s
${CMAKE_CURRENT_BINARY_DIR}/A_S_H_F8_256_4_${arch}.s
${CMAKE_CURRENT_BINARY_DIR}/A_S_H_B8_256_4_${arch}.s
COMMAND ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_F8_256_4_${arch}.s -t S -d S -s F8 -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_B8_256_4_${arch}.s -t S -d S -s B8 -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_F8_256_4_${arch}.s -t S -d H -s F8 -w 256 -c 4 --arch ${arch}
COMMAND ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_B8_256_4_${arch}.s -t S -d H -s B8 -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_F8_256_4_${arch}.s -t S -d S -s F8 -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_S_B8_256_4_${arch}.s -t S -d S -s B8 -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_F8_256_4_${arch}.s -t S -d H -s F8 -w 256 -c 4 --arch ${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} AMaxGenerator.py --is-scale -o ${CMAKE_CURRENT_BINARY_DIR}/A_S_H_B8_256_4_${arch}.s -t S -d H -s B8 -w 256 -c 4 --arch ${arch}
COMMENT "Creating Extra Amax Assembly for gfx950"
WORKING_DIRECTORY ${ops_path}
)
Expand All @@ -142,21 +142,31 @@ foreach(arch IN LISTS archs)
DEPENDS ExtOpObj_${arch}
OUTPUT ${CMAKE_CURRENT_BINARY_DIR}/extop_${arch}.co
COMMAND ${CMAKE_CXX_COMPILER};-target;amdgcn-amdhsa;-Xlinker;$<TARGET_OBJECTS:ExtOpObj_${arch}>;-o;${CMAKE_CURRENT_BINARY_DIR}/extop_${arch}.co
COMMAND ${CMAKE_COMMAND} -E copy "${CMAKE_CURRENT_BINARY_DIR}/extop_${arch}.co" "${PROJECT_BINARY_DIR}/Tensile/library/"
COMMENT "Creating extop_${arch}"
COMMAND_EXPAND_LISTS
)
add_custom_target(ExtOpLibrary_${arch} ALL
DEPENDS ${dat_depends} ${CMAKE_CURRENT_BINARY_DIR}/extop_${arch}.co
COMMAND ${python_launch_prefix} ExtOpCreateLibrary.py --src=${CMAKE_CURRENT_BINARY_DIR} --co=${CMAKE_CURRENT_BINARY_DIR}/extop_${arch}.co --output=${CMAKE_CURRENT_BINARY_DIR} --arch=${arch}
COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} ExtOpCreateLibrary.py --src=${CMAKE_CURRENT_BINARY_DIR} --co=${CMAKE_CURRENT_BINARY_DIR}/extop_${arch}.co --output=${CMAKE_CURRENT_BINARY_DIR} --arch=${arch}
COMMAND ${CMAKE_COMMAND} -E copy "${CMAKE_CURRENT_BINARY_DIR}/hipblasltExtOpLibrary.dat" "${PROJECT_BINARY_DIR}/Tensile/library/"
COMMENT "Creating hipblasltExtOpLibrary.dat"
WORKING_DIRECTORY ${ops_path}
)
list(APPEND dat_depends "ExtOpLibrary_${arch}")
endforeach()

add_custom_target(ExtOpCp ALL
DEPENDS ${dat_depends} TENSILE_LIBRARY_TARGET
COMMAND ${CMAKE_COMMAND} -E copy "${CMAKE_CURRENT_BINARY_DIR}/*.co" ${PROJECT_BINARY_DIR}/Tensile/library
COMMAND ${CMAKE_COMMAND} -E copy "${CMAKE_CURRENT_BINARY_DIR}/*.dat" ${PROJECT_BINARY_DIR}/Tensile/library
COMMENT "Copying ExtOp Library"
)
add_custom_target(ExtOpCp ALL
DEPENDS ${dat_depends} TENSILE_LIBRARY_TARGET
COMMENT "Copying .co and .dat files"
)

file(GLOB FILES_TO_COPY "${CMAKE_CURRENT_BINARY_DIR}/*.co" "${CMAKE_CURRENT_BINARY_DIR}/*.dat")

@stellaraccident stellaraccident Jun 3, 2025

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I don't understand how this can work: the glob will be evaluated at configure time but it is trying to find files that are generated at build time. I think this will only work after the first build, which would be wrong.


foreach(file ${FILES_TO_COPY})
add_custom_command(
TARGET ExtOpCp POST_BUILD
COMMAND ${CMAKE_COMMAND} -E copy "${file}" "${PROJECT_BINARY_DIR}/Tensile/library/"
COMMENT "Copying ${file} to ${PROJECT_BINARY_DIR}/Tensile/library/"
)
endforeach()
3 changes: 3 additions & 0 deletions library/src/amd_detail/rocblaslt/src/rocblaslt_auxiliary.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -2256,11 +2256,14 @@ std::optional<std::filesystem::path> rocblaslt_find_library_relative_path(
// {lib_dir}/hipblaslt/library
// Legacy:
// {lib_dir}/../Tensile/library
// {lib_dir}/../../Tensile/library
// {lib_dir}/library
if(auto p = pathIfExists(lib_dir / "hipblaslt" / "library"))
return *p;
if(auto p = pathIfExists(lib_dir.parent_path() / "Tensile" / "library"))
return *p;
if(auto p = pathIfExists(lib_dir.parent_path().parent_path() / "Tensile" / "library"))
return *p;
if(auto p = pathIfExists(lib_dir / "library"))
return *p;
return std::nullopt;
Expand Down
Loading