From c21cde00aa046376a7c7c02953166fa6e994c44a Mon Sep 17 00:00:00 2001 From: Milica Trifunovic Date: Tue, 3 Jun 2025 11:00:08 -0400 Subject: [PATCH 1/4] Enable building on Windows with necessary modifications --- CMakeLists.txt | 38 +- clients/CMakeLists.txt | 23 +- clients/benchmarks/CMakeLists.txt | 2 + clients/gtest/CMakeLists.txt | 53 ++- cmake/virtualenv.cmake | 9 +- library/CMakeLists.txt | 22 +- .../rocblaslt/src/extops/CMakeLists.txt | 62 +-- .../rocblaslt/src/rocblaslt_auxiliary.cpp | 3 + rdeps.py | 226 +++++++++++ rdeps.xml | 11 + rmake.py | 357 ++++++++++++++++++ tensilelite/Tensile/Source/lib/CMakeLists.txt | 10 +- tensilelite/Tensile/Toolchain/Component.py | 11 +- toolchain-windows.cmake | 73 ++++ 14 files changed, 833 insertions(+), 67 deletions(-) create mode 100644 rdeps.py create mode 100644 rdeps.xml create mode 100644 rmake.py create mode 100644 toolchain-windows.cmake diff --git a/CMakeLists.txt b/CMakeLists.txt index cd0c83142a..04bbba2050 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -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) @@ -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() @@ -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" ) diff --git a/clients/CMakeLists.txt b/clients/CMakeLists.txt index 05e6c98901..305f78062d 100755 --- a/clients/CMakeLists.txt +++ b/clients/CMakeLists.txt @@ -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 diff --git a/clients/benchmarks/CMakeLists.txt b/clients/benchmarks/CMakeLists.txt index b9e8db1187..de9a4c0c34 100644 --- a/clients/benchmarks/CMakeLists.txt +++ b/clients/benchmarks/CMakeLists.txt @@ -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} ) diff --git a/clients/gtest/CMakeLists.txt b/clients/gtest/CMakeLists.txt index ac7df61cdd..e8edc0c2fe 100644 --- a/clients/gtest/CMakeLists.txt +++ b/clients/gtest/CMakeLists.txt @@ -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() @@ -89,32 +94,44 @@ target_compile_definitions( hipblaslt-test PRIVATE ROCM_USE_FLOAT16 HIPBLASLT_IN target_compile_options(hipblaslt-test PRIVATE $<$:${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) + # 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" diff --git a/cmake/virtualenv.cmake b/cmake/virtualenv.cmake index 7b4a78fe40..a201b4f07b 100644 --- a/cmake/virtualenv.cmake +++ b/cmake/virtualenv.cmake @@ -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") @@ -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) diff --git a/library/CMakeLists.txt b/library/CMakeLists.txt index 7ff210af2b..2ffaca95b6 100644 --- a/library/CMakeLists.txt +++ b/library/CMakeLists.txt @@ -78,6 +78,22 @@ endif() # Target compile definitions if(NOT BUILD_CUDA) + if (WIN32) + 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 $ + ) + + target_link_libraries(hipblaslt PRIVATE TensileHost shlwapi) + endif() + endif() if( BUILD_SHARED_LIBS ) target_link_libraries( hipblaslt PRIVATE TensileHost ) @@ -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) + add_custom_command( TARGET hipblaslt POST_BUILD COMMAND ${CMAKE_COMMAND} -E copy ${PROJECT_BINARY_DIR}/staging/$ ${PROJECT_BINARY_DIR}/clients/staging/$ ) +endif() + # TODO ?? # Following boost conventions of prefixing 'lib' on static built libraries if(NOT BUILD_SHARED_LIBS) @@ -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" ) else() set( HIPBLASLT_TENSILE_LIBRARY_DIR "\${CPACK_PACKAGING_INSTALL_PREFIX}${CMAKE_INSTALL_LIBDIR}/hipblaslt" CACHE PATH "path to tensile library" ) endif() diff --git a/library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt b/library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt index e08b14eb52..721d49a395 100644 --- a/library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt +++ b/library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt @@ -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} ) @@ -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} ) @@ -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} ) @@ -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;$;-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") + +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() diff --git a/library/src/amd_detail/rocblaslt/src/rocblaslt_auxiliary.cpp b/library/src/amd_detail/rocblaslt/src/rocblaslt_auxiliary.cpp index d30ec35507..0263eae846 100644 --- a/library/src/amd_detail/rocblaslt/src/rocblaslt_auxiliary.cpp +++ b/library/src/amd_detail/rocblaslt/src/rocblaslt_auxiliary.cpp @@ -2256,11 +2256,14 @@ std::optional 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; diff --git a/rdeps.py b/rdeps.py new file mode 100644 index 0000000000..0ca6d288ed --- /dev/null +++ b/rdeps.py @@ -0,0 +1,226 @@ +#!/usr/bin/python3 + +"""Copyright (C) 2021-2023 Advanced Micro Devices, Inc. All rights reserved. + Permission is hereby granted, free of charge, to any person obtaining a copy + of this software and associated documentation files (the "Software"), to deal + in the Software without restriction, including without limitation the rights + to use, copy, modify, merge, publish, distribute, sublicense, and/or sell cop- + ies of the Software, and to permit persons to whom the Software is furnished + to do so, subject to the following conditions: + The above copyright notice and this permission notice shall be included in all + copies or substantial portions of the Software. + THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR IM- + PLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, FITNESS + FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE AUTHORS OR + COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER LIABILITY, WHETHER + IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR IN CONNE- + CTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. +""" + +import os +import platform +import subprocess +import argparse +import pathlib +from xml.dom import minidom +import shutil + +SCRIPT_VERSION = 0.1 + +args = {} +param = {} +OS_info = {} +var_subs = {} + +vcpkg_script = ['tdir %IDIR%', + 'git clone -b 2024.03.25 https://github.com/microsoft/vcpkg %IDIR%', 'cd %IDIR%', 'bootstrap-vcpkg.bat -disableMetrics' ] + +xml_script = [ '%XML%' ] + + +def parse_args(): + """Parse command-line arguments""" + parser = argparse.ArgumentParser(description=""" + Checks build arguments + """) + # parser.add_argument('--install', required=False, default = True, action='store_true', + # help='Install dependencies (optional, default: True)') + parser.add_argument('-i', '--install_dir', type=str, required=False, default = ("" if os.name == "nt" else "./build/deps"), + + help='Install directory path (optional, windows default: C:\\github\\vcpkg, linux default: ./build/deps)') + # parser.add_argument('-v', '--verbose', required=False, default = False, action='store_true', + # help='Verbose install (optional, default: False)') + return parser.parse_args() + +def os_detect(): + global OS_info + OS_info["ID"] = platform.system() + OS_info["NUM_PROC"] = os.cpu_count() + print(OS_info) + +def create_dir(dir_path): + if os.path.isabs(dir_path): + full_path = dir_path + else: + full_path = os.path.join( os.getcwd(), dir_path ) + return pathlib.Path(full_path).mkdir(parents=True, exist_ok=True) + +def delete_dir(dir_path) : + if (not os.path.exists(dir_path)): + return + return shutil.rmtree(dir_path, True) + +def run_cmd(cmd): + global args + if (cmd.startswith('cd ')): + return os.chdir(cmd[3:]) + if (cmd.startswith('mkdir ')): + return create_dir(cmd[6:]) + cmdline = f"{cmd}" + print(cmdline) + proc = subprocess.run(cmdline, check=True, stderr=subprocess.STDOUT, shell=True) + return proc.returncode + + +def install_deps( os_node ): + global var_subs + + cwd = pathlib.Path.absolute(pathlib.Path(os.curdir)) + + vc_node = os_node.getElementsByTagName('vcpkg') + if vc_node: + cmdline = "cd %IDIR%" + cd_vcpkg = cmdline.replace('%IDIR%', args.install_dir) + run_cmd(cd_vcpkg) + for p in vc_node[0].getElementsByTagName('pkg'): + name = p.getAttribute('name') + package = p.firstChild.data + if name: + print( f'***\n*** VCPKG Installing: {name}\n***' ) + raw_cmd = p.firstChild.data + var_cmd = raw_cmd.format_map(var_subs) + error = run_cmd( f'vcpkg.exe install {var_cmd}') + os.chdir(cwd) + + pip_node = os_node.getElementsByTagName('pip') + if pip_node: + for p in pip_node[0].getElementsByTagName('pkg'): + name = p.getAttribute('name') + package = p.firstChild.data + if name: + print( f'***\n*** Pip Installing: {name}\n***' ) + raw_cmd = p.firstChild.data + var_cmd = raw_cmd.format_map(var_subs) + error = run_cmd( f'pip install {var_cmd}') + for p in pip_node[0].getElementsByTagName('req'): + name = p.getAttribute('name') + package = p.firstChild.data + if name: + print( f'***\n*** Pip Requirements: {name}\n***' ) + raw_cmd = p.firstChild.data + var_cmd = raw_cmd.format_map(var_subs) + requirements_file = os.path.abspath( os.path.join( cwd, var_cmd ) ) + error = run_cmd( f'pip install -r {requirements_file}') + +def run_install_script(script, xml): + '''executes a simple batch style install script, the scripts are defined at top of file''' + global OS_info + global args + global var_subs + # + cwd = pathlib.Path.absolute(pathlib.Path(os.getcwd())) + + fail = False + last_cmd_index = 0 + for i in range(len(script)): + last_cmd_index = i + cmdline = script[i] + cmd = cmdline.replace('%IDIR%', args.install_dir) + if cmd.startswith('tdir '): + if pathlib.Path(cmd[5:]).exists(): + break # all further cmds skipped + else: + continue + error = False + if cmd.startswith('%XML%'): + fileversion = xml.getElementsByTagName('fileversion') + if len(fileversion) == 0: + print("WARNING: Could not find the version of this xml configuration file.") + elif len(fileversion) > 1: + print("WARNING: Multiple version tags found.") + else: + version = float(fileversion[0].firstChild.data) + if version > SCRIPT_VERSION: + print(f"ERROR: This xml requires script version >= {version}, running script version {SCRIPT_VERSION}") + exit(1) + + for var in xml.getElementsByTagName('var'): + name = var.getAttribute('name') + if var.hasAttribute('value'): + val = var.getAttribute('value') + elif var.firstChild is not None: + val = var.firstChild.data + else: + val = "" + var_subs[name] = val + + for os_node in xml.getElementsByTagName('os'): + os_names = os_node.getAttribute('names') + os_list = os_names.split(',') + if (OS_info['ID'] in os_list) or ("all" in os_list): + error = install_deps( os_node ) + else: + error = run_cmd(cmd) + fail = fail or error + if fail: + break + + os.chdir( cwd ) + if (fail): + if (script[last_cmd_index] == "%XML%"): + print(f"FAILED xml dependency installation!") + else: + print(f"ERROR running: {script[last_cmd_index]}") + return 1 + else: + return 0 + +def installation(): + global vcpkg_script + global xml_script + global xmlDoc + + # install + cwd = os.getcwd() + + xmlPath = os.path.join( cwd, 'rdeps.xml') + xmlDoc = minidom.parse( xmlPath ) + + scripts = [] + + if xmlDoc.getElementsByTagName('vcpkg'): + scripts.append( vcpkg_script ) + scripts.append( xml_script ) + + for i in scripts: + if (run_install_script(i, xmlDoc)): + #print("Failure in script. ABORTING") + os.chdir( cwd ) + return 1 + os.chdir( cwd ) + return 0 + +def main(): + global args + + os_detect() + args = parse_args() + + if not args.install_dir: + vcpkg_root = os.getenv( 'VCPKG_PATH', "C:\\github\\vcpkg") + args.install_dir = vcpkg_root + + installation() + +if __name__ == '__main__': + main() \ No newline at end of file diff --git a/rdeps.xml b/rdeps.xml new file mode 100644 index 0000000000..d444f3921c --- /dev/null +++ b/rdeps.xml @@ -0,0 +1,11 @@ + + 0.1 + + + gtest:x64-windows + msgpack:x64-windows + openblas:x64-windows + lapack:x64-windows + + + \ No newline at end of file diff --git a/rmake.py b/rmake.py new file mode 100644 index 0000000000..e7c871c71a --- /dev/null +++ b/rmake.py @@ -0,0 +1,357 @@ +#!/usr/bin/python3 +"""Copyright (C) 2020-2023 Advanced Micro Devices, Inc. All rights reserved. + Permission is hereby granted, free of charge, to any person obtaining a copy + of this software and associated documentation files (the "Software"), to deal + in the Software without restriction, including without limitation the rights + to use, copy, modify, merge, publish, distribute, sublicense, and/or sell cop- + ies of the Software, and to permit persons to whom the Software is furnished + to do so, subject to the following conditions: + The above copyright notice and this permission notice shall be included in all + copies or substantial portions of the Software. + THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR IM- + PLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, FITNESS + FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE AUTHORS OR + COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER LIABILITY, WHETHER + IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR IN CONNE- + CTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. +""" + +import os +import platform +import subprocess +import shutil +import argparse +import pathlib + +try: + import psutil + psutil_imported = True +except ImportError: + psutil_imported = False + +args = {} +OS_info = {} + + +# yapf: disable +def parse_args(): + """Parse command-line arguments""" + global OS_info + + parser = argparse.ArgumentParser(description="""Checks build arguments""") + + parser.add_argument('-a', '--architecture', dest='gpu_architecture', required=False, default="all", + help='Set GPU architectures, e.g. all, auto, "gfx803;gfx906:xnack-", gfx1030, gfx1101 (optional, default: all)') + + parser.add_argument( '--address-sanitizer', dest='address_sanitizer', required=False, default=False, action='store_true', + help='Build with address sanitizer enabled. (optional, default: False') + + parser.add_argument( '--build_dir', type=str, required=False, default = "build", + help='Configure & build process output directory.(optional, default: ./build)') + + parser.add_argument('-b', '--branch', dest='tensile_tag', type=str, required=False, default="", + help='Specify the Tensile repository branch or tag to use. (eg. develop, mybranch or )') + + parser.add_argument('-c', '--clients', dest='build_clients', required=False, default=False, action='store_true', + help='Build the library clients benchmark and gtest (optional, default: False, Generated binaries will be located at /clients/staging)') + + parser.add_argument( '--codecoverage', required=False, default=False, action='store_true', + help='Code coverage build. Requires Debug (-g|--debug) or RelWithDebInfo mode (-k|--relwithdebinfo), (optional, default: False)') + + parser.add_argument( '-d', '--dependencies', required=False, default=False, action='store_true', + help='Build and install external dependencies. (Handled by install.sh and on Windows rdeps.py') + + parser.add_argument('-f', '--fork', dest='tensile_fork', type=str, required=False, default="", + help='Specify the username to fork the Tensile GitHub repository (e.g., ROCmSoftwarePlatform or MyUserName)') + + parser.add_argument('-g', '--debug', required=False, default=False, action='store_true', + help='Build in Debug mode (optional, default: False)') + + parser.add_argument('-i', '--install', required=False, default=False, dest='install', action='store_true', + help='Generate and install library package after build. Windows only. Linux use install.sh (optional, default: False)') + + parser.add_argument('-j', '--jobs', type=int, required=False, default=0, + help='Specify number of parallel jobs to launch, increases memory usage (default: heuristic around logical core count)') + + parser.add_argument('-k', '--relwithdebinfo', required=False, default=False, action='store_true', + help='Build in Release with Debug Info (optional, default: False)') + + parser.add_argument('-l', '--logic', dest='tensile_logic', type=str, required=False, default="asm_full", + help='Specify the Tensile logic target, e.g., asm_full, asm_lite, etc. (optional, default: asm_full)') + + parser.add_argument('-n', '--no_tensile', dest='build_tensile', required=False, default=True, action='store_false', + help='Build a subset of hipblaslt library which does not require Tensile.') + + parser.add_argument( '--msgpack', dest='tensile_msgpack_backend', required=False, default=True, action='store_true', + help='Build Tensile backend to use MessagePack (optional, default: True)') + + parser.add_argument( '--no-msgpack', dest='tensile_msgpack_backend', required=False, default=True, action='store_false', + help='Build Tensile backend not to use MessagePack and so use YAML (optional)') + + parser.add_argument('-s', '--static', required=False, default=False, dest='static_lib', action='store_true', + help='Build hipblaslt as a static library. (optional, default: False)') + + parser.add_argument('-t', '--test_local_path', dest='tensile_test_local_path', type=str, required=False, default="", + help='Use a local path for Tensile instead of remote GIT repo (optional)') + + parser.add_argument('-u', '--use-custom-version', dest='tensile_version', type=str, required=False, default="", + help='Ignore Tensile version and just use the Tensile tag (optional)') + + parser.add_argument('-v', '--verbose', required=False, default=False, action='store_true', + help='Verbose build (optional, default: False)') + + return parser.parse_args() +# yapf: enable + +def get_ram_GB(): + """ + Total amount of GB RAM available or zero if unknown + """ + if psutil_imported: + gb = round(psutil.virtual_memory().total / pow(1024, 3)) + else: + gb = 0 + return gb + +def strip_ECC(token): + return token.replace(':sramecc+', '').replace(':sramecc-', '').strip() + +def gpu_detect(): + global OS_info + OS_info["GPU"] = "" + cmd = "hipinfo.exe" + + process = subprocess.run([cmd], stdout=subprocess.PIPE) + for line_in in process.stdout.decode().splitlines(): + if 'gcnArchName' in line_in: + OS_info["GPU"] = strip_ECC( line_in.split(":")[1] ) + break + +def os_detect(): + global OS_info + OS_info["ID"] = platform.system() + OS_info["NUM_PROC"] = os.cpu_count() + OS_info["RAM_GB"] = get_ram_GB() + +def jobs_heuristic(): + # auto jobs heuristics + jobs = min(OS_info["NUM_PROC"], 128) # disk limiter + ram = OS_info["RAM_GB"] + if (ram >= 16): # don't apply if below minimum RAM + jobs = min(round(ram/2), jobs) # RAM limiter + hipcc_flags = os.getenv('HIPCC_COMPILE_FLAGS_APPEND', "") + pjstr = hipcc_flags.split("parallel-jobs=") + if (len(pjstr) > 1): + pjobs = int(pjstr[1][0]) + if (pjobs > 1 and pjobs < jobs): + jobs = round(jobs / pjobs) + jobs = min(61, jobs) # multiprocessing limit (used by tensile) + return int(jobs) + +def create_dir(dir_path): + full_path = "" + if os.path.isabs(dir_path): + full_path = dir_path + else: + full_path = os.path.join(os.getcwd(), dir_path) + pathlib.Path(full_path).mkdir(parents=True, exist_ok=True) + return + + +def delete_dir(dir_path): + if (not os.path.exists(dir_path)): + return + shutil.rmtree(dir_path) + +def cmake_path(os_path): + return os_path.replace("\\", "/") + +def fatal(msg, code=1): + print(msg) + exit(code) + + +def deps_cmd(): + exe = f"python rdeps.py" + all_args = "" + return exe, all_args + + +def config_cmd(): + global args + global OS_info + cwd_path = os.getcwd() + cmake_executable = "cmake" + cmake_options = [] + src_path = cmake_path(cwd_path) + cmake_platform_opts = [] + generator = f"-G Ninja" + cmake_options.append(generator) + + # CMAKE_PREFIX_PATH set to rocm_path and HIP_PATH set BY SDK Installer + raw_rocm_path = cmake_path(os.getenv('HIP_PATH', "C:/hip")) + + if raw_rocm_path: + os.environ["HIP_PATH"] = raw_rocm_path + + + rocm_path = f'"{raw_rocm_path}"' # guard against spaces in path + # CPACK_PACKAGING_INSTALL_PREFIX= defined as blank as it is appended to end of path for archive creation + #cmake_platform_opts.append(f"-DCPACK_PACKAGING_INSTALL_PREFIX=") + #cmake_platform_opts.append(f'-DCMAKE_INSTALL_PREFIX="C:/hipSDK"') + cmake_platform_opts.append( f"-DCPACK_PACKAGING_INSTALL_PREFIX=" ) + cmake_platform_opts.append( f"-DCMAKE_INSTALL_PREFIX=\"C:/hipSDK\"" ) + toolchain = os.path.join(src_path, "toolchain-windows.cmake") + print(f"Build source path: {src_path}") + + tools = f"-DCMAKE_TOOLCHAIN_FILE={toolchain}" + cmake_options.append(tools) + + cmake_options.extend(cmake_platform_opts) + + cmake_base_options = f"-DROCM_PATH={rocm_path} -DCMAKE_PREFIX_PATH:PATH={rocm_path}" + cmake_options.append(cmake_base_options) + + # packaging options + cmake_pack_options = f"-DCPACK_SET_DESTDIR=OFF" + cmake_options.append(cmake_pack_options) + + if os.getenv('CMAKE_CXX_COMPILER_LAUNCHER'): + cmake_options.append(f'-DCMAKE_CXX_COMPILER_LAUNCHER={os.getenv("CMAKE_CXX_COMPILER_LAUNCHER")}') + + # build type + cmake_config = "" + build_dir = os.path.realpath(args.build_dir) + if args.debug: + build_path = os.path.join(build_dir, "debug") + cmake_config = "Debug" + elif args.relwithdebinfo: + build_path = os.path.join(build_dir, "release-debug") + cmake_config = "RelWithDebInfo" + else: + build_path = os.path.join(build_dir, "release") + cmake_config = "Release" + + cmake_options.append(f"-DCMAKE_BUILD_TYPE={cmake_config}") + + if args.codecoverage: + if args.debug or args.relwithdebinfo: + cmake_options.append(f"-DBUILD_CODE_COVERAGE=ON") + else: + fatal("*** Code coverage is not supported for Release build! Aborting. ***") + + if args.address_sanitizer: + cmake_options.append(f"-DBUILD_ADDRESS_SANITIZER=ON") + + # clean + delete_dir(build_path) + + create_dir(os.path.join(build_path, "clients")) + os.chdir(build_path) + + if args.static_lib: + cmake_options.append(f"-DBUILD_SHARED_LIBS=OFF") + + if args.build_clients: + cmake_build_dir = cmake_path(build_dir) + cmake_options.append( + f"-DBUILD_CLIENTS_TESTS=ON -DBUILD_CLIENTS_BENCHMARKS=ON -DBUILD_CLIENTS_SAMPLES=ON -DBUILD_DIR={cmake_build_dir}" + ) + + if args.gpu_architecture == "auto": + gpu_detect() + if len(OS_info["GPU"]): + args.gpu_architecture = OS_info["GPU"] + else: + fatal("Could not detect GPU as requested. Not continuing.") + # not just for tensile + cmake_options.append(f'-DAMDGPU_TARGETS=\"{args.gpu_architecture}\"') + + if not args.build_tensile: + cmake_options.append(f"-DBUILD_WITH_TENSILE=OFF") + else: + cmake_options.append(f"-DTensile_CODE_OBJECT_VERSION=4") + if args.tensile_logic: + cmake_options.append(f"-DTensile_LOGIC={args.tensile_logic}") + if args.tensile_fork: + cmake_options.append(f"-Dtensile_fork={args.tensile_fork}") + if args.tensile_tag: + cmake_options.append(f"-Dtensile_tag={args.tensile_tag}") + if args.tensile_test_local_path: + cmake_options.append(f"-DTensile_TEST_LOCAL_PATH={args.tensile_test_local_path}") + if args.tensile_version: + cmake_options.append(f"-DTENSILE_VERSION={args.tensile_version}") + if args.tensile_msgpack_backend: + cmake_options.append(f"-DTensile_LIBRARY_FORMAT=msgpack") + else: + cmake_options.append(f"-DTensile_LIBRARY_FORMAT=yaml") + if args.jobs != OS_info["NUM_PROC"]: + cmake_options.append(f"-DTensile_CPU_THREADS={str(args.jobs)}") + + cmake_options.append(f"{src_path}") + cmd_opts = " ".join(cmake_options) + + return cmake_executable, cmd_opts + + +def make_cmd(): + global args + global OS_info + + make_options = [] + + # the CMAKE_BUILD_PARALLEL_LEVEL currently doesn't work for windows build, so using -j + # make_executable = f"cmake.exe -DCMAKE_BUILD_PARALLEL_LEVEL=4 --build . " # ninja + make_executable = f"ninja.exe -j {args.jobs}" + if args.verbose: + make_options.append("--verbose") + make_options.append("all") # for cmake "--target all" ) + if args.install: + make_options.append("package install") # for cmake "--target package --target install" ) + cmd_opts = " ".join(make_options) + + return make_executable, cmd_opts + + +def run_cmd(exe, opts): + program = f"{exe} {opts}" + print(program) + proc = subprocess.run(program, check=True, stderr=subprocess.STDOUT, shell=True) + return proc.returncode + + +def main(): + global args + os_detect() + args = parse_args() + + if args.jobs == 0: + args.jobs = jobs_heuristic() + if args.jobs > 61: + print( f"WARNING: jobs > 61 may fail on windows python multiprocessing (jobs = {args.jobs}).") + + print(OS_info) + + root_dir = os.curdir + + # depdendency install + if (args.dependencies): + exe, opts = deps_cmd() + if run_cmd(exe, opts): + fatal("Dependency install failed. Not continuing.") + + # configure + exe, opts = config_cmd() + if run_cmd(exe, opts): + fatal("Configuration failed. Not continuing.") + + # make + exe, opts = make_cmd() + if run_cmd(exe, opts): + fatal("Build failed. Not continuing.") + + # Linux install and cleanup not supported from rmake yet + + +if __name__ == '__main__': + main() \ No newline at end of file diff --git a/tensilelite/Tensile/Source/lib/CMakeLists.txt b/tensilelite/Tensile/Source/lib/CMakeLists.txt index 3d752f96f5..3eb661c477 100644 --- a/tensilelite/Tensile/Source/lib/CMakeLists.txt +++ b/tensilelite/Tensile/Source/lib/CMakeLists.txt @@ -115,11 +115,15 @@ if(TENSILE_USE_LLVM OR TENSILE_USE_MSGPACK) endif() if(TENSILE_USE_MSGPACK) - find_package(msgpack REQUIRED) + if (WIN32) + find_package(msgpack-cxx REQUIRED) + else() + find_package(msgpack REQUIRED) + endif() target_compile_definitions(TensileHost PUBLIC -DTENSILE_MSGPACK=1) - if(TARGET msgpackc-cxx) - get_target_property(msgpack_inc msgpackc-cxx INTERFACE_INCLUDE_DIRECTORIES) + if(TARGET msgpack-cxx) + get_target_property(msgpack_inc msgpack-cxx INTERFACE_INCLUDE_DIRECTORIES) elseif(TARGET msgpackc) get_target_property(msgpack_inc msgpackc INTERFACE_INCLUDE_DIRECTORIES) endif() diff --git a/tensilelite/Tensile/Toolchain/Component.py b/tensilelite/Tensile/Toolchain/Component.py index 1a1f1528ec..5679fbee74 100644 --- a/tensilelite/Tensile/Toolchain/Component.py +++ b/tensilelite/Tensile/Toolchain/Component.py @@ -24,7 +24,6 @@ ################################################################################ from os import name as os_name from os import environ -from os import sysconf from pathlib import Path from re import search, IGNORECASE from shlex import split @@ -34,6 +33,9 @@ from Tensile.Common import SemanticVersion, print1 from .Validators import ToolchainDefaults, validateToolchain +if os_name != "nt": + from os import sysconf + def _invoke(args: List[str], desc: str=""): """Invokes a command with the provided arguments in a subprocess. Args: @@ -368,9 +370,10 @@ def _use_response_file(self, args: List[str]) -> bool: On Unix: check against system argument length limit """ if os_name == "nt": - return True - line_length = sum(len(arg) for arg in args) + len(args) - 1 - return line_length >= sysconf("SC_ARG_MAX") + return True + else: + line_length = sum(len(arg) for arg in args) + len(args) - 1 + return line_length >= sysconf("SC_ARG_MAX") def __call__(self, srcPaths: List[str], destPath: str): """ diff --git a/toolchain-windows.cmake b/toolchain-windows.cmake new file mode 100644 index 0000000000..b36fd1ca4a --- /dev/null +++ b/toolchain-windows.cmake @@ -0,0 +1,73 @@ +# ######################################################################## +# Copyright (C) 2022-2023 Advanced Micro Devices, Inc. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in +# all copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +# THE SOFTWARE. +# +# ######################################################################## + +if (DEFINED ENV{HIP_PATH}) + file(TO_CMAKE_PATH "$ENV{HIP_PATH}" HIP_DIR) + set(rocm_bin "${HIP_DIR}/bin") +elseif (DEFINED ENV{HIP_DIR}) + file(TO_CMAKE_PATH "$ENV{HIP_DIR}" HIP_DIR) + set(rocm_bin "${HIP_DIR}/bin") +else() + set(HIP_DIR "C:/hip") + set(rocm_bin "C:/hip/bin") +endif() + +set(CMAKE_CXX_COMPILER "${rocm_bin}/clang++.exe") +set(CMAKE_C_COMPILER "${rocm_bin}/clang.exe") + +if (NOT python) + set(python "python") # take default for windows +endif() + +# our usage flags +set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DWIN32 -DWIN32_LEAN_AND_MEAN -DNOMINMAX -D_CRT_SECURE_NO_WARNINGS -D_SILENCE_EXPERIMENTAL_FILESYSTEM_DEPRECATION_WARNING") + +# flags for clang direct use + +# -Wno-ignored-attributes to avoid warning: __declspec attribute 'dllexport' is not supported [-Wignored-attributes] which is used by msvc compiler +set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -Wno-ignored-attributes") + +set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DHIP_CLANG_HCC_COMPAT_MODE=1") + +# args also in hipcc.bat +set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -fms-extensions -fms-compatibility -D__HIP_ROCclr__=1 -D__HIP_PLATFORM_AMD__=1 ") + +if (DEFINED ENV{OPENBLAS_DIR}) + file(TO_CMAKE_PATH "$ENV{OPENBLAS_DIR}" OPENBLAS_DIR) +else() + set(OPENBLAS_DIR "C:/OpenBLAS/OpenBLAS-0.3.18-x64") +endif() + +if (DEFINED ENV{VCPKG_PATH}) + file(TO_CMAKE_PATH "$ENV{VCPKG_PATH}" VCPKG_PATH) +else() + set(VCPKG_PATH "C:/github/vcpkg") +endif() +include("${VCPKG_PATH}/scripts/buildsystems/vcpkg.cmake") + +set(CMAKE_STATIC_LIBRARY_SUFFIX ".a") +set(CMAKE_STATIC_LIBRARY_PREFIX "static_") +set(CMAKE_SHARED_LIBRARY_SUFFIX ".dll") +set(CMAKE_SHARED_LIBRARY_PREFIX "") + +set(BUILD_FORTRAN_CLIENTS OFF) \ No newline at end of file From 88ee306db609d9acb64caf7e90add730ae514743 Mon Sep 17 00:00:00 2001 From: Milica Trifunovic Date: Mon, 16 Jun 2025 08:20:20 -0400 Subject: [PATCH 2/4] Add flush after write in clang_args file --- tensilelite/Tensile/Toolchain/Component.py | 1 + 1 file changed, 1 insertion(+) diff --git a/tensilelite/Tensile/Toolchain/Component.py b/tensilelite/Tensile/Toolchain/Component.py index 5679fbee74..9c1d0a1c0b 100644 --- a/tensilelite/Tensile/Toolchain/Component.py +++ b/tensilelite/Tensile/Toolchain/Component.py @@ -360,6 +360,7 @@ def _response_file_args(self, srcPaths: List[str], destPath: str) -> List[str]: """ with open(Path.cwd() / "clang_args.txt", "wt") as file: file.write(" ".join(srcPaths).replace('\\', '\\\\') if os_name == "nt" else " ".join(srcPaths)) + file.flush() return [*(self.default_args), "-o", destPath, "@clang_args.txt"] def _use_response_file(self, args: List[str]) -> bool: From ceec597997d1be8cb263a36ea161a84974da5f9d Mon Sep 17 00:00:00 2001 From: Milica Trifunovic Date: Thu, 19 Jun 2025 05:33:01 -0400 Subject: [PATCH 3/4] Fix copying *.co files --- .../rocblaslt/src/extops/CMakeLists.txt | 14 +++--------- tensilelite/copy_files.py | 22 +++++++++++++++++++ 2 files changed, 25 insertions(+), 11 deletions(-) create mode 100644 tensilelite/copy_files.py diff --git a/library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt b/library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt index 721d49a395..c76c7e451c 100644 --- a/library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt +++ b/library/src/amd_detail/rocblaslt/src/extops/CMakeLists.txt @@ -157,16 +157,8 @@ foreach(arch IN LISTS archs) endforeach() add_custom_target(ExtOpCp ALL - DEPENDS ${dat_depends} TENSILE_LIBRARY_TARGET + DEPENDS ${dat_depends} TENSILE_LIBRARY_TARGET + COMMAND ${CMAKE_COMMAND} -E env ${python_launch_prefix} "${ops_path}/copy_files.py" ${CMAKE_CURRENT_BINARY_DIR} ${PROJECT_BINARY_DIR}/Tensile/library/ COMMENT "Copying .co and .dat files" -) +) -file(GLOB FILES_TO_COPY "${CMAKE_CURRENT_BINARY_DIR}/*.co" "${CMAKE_CURRENT_BINARY_DIR}/*.dat") - -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() diff --git a/tensilelite/copy_files.py b/tensilelite/copy_files.py new file mode 100644 index 0000000000..8b84abbc41 --- /dev/null +++ b/tensilelite/copy_files.py @@ -0,0 +1,22 @@ +import shutil +import sys +from pathlib import Path + +def copy_files(source_dir, dest_dir, extensions): + source = Path(source_dir) + destination = Path(dest_dir) + + for ext in extensions: + for file_path in source.rglob(f"*{ext}"): + try: + shutil.copy2(file_path, destination) + print(f"Copied {file_path} to {destination}") + except Exception as e: + print(f"Failed to copy {file_path}: {e}", file=sys.stderr) + +if __name__ == "__main__": + source_directory = sys.argv[1] + destination_directory = sys.argv[2] + file_extensions = ['.co', '.dat'] + + copy_files(source_directory, destination_directory, file_extensions) \ No newline at end of file From 02ba28bce11e1e08f3560cbf9e79f70284685164 Mon Sep 17 00:00:00 2001 From: Milica Trifunovic Date: Thu, 19 Jun 2025 10:29:02 -0400 Subject: [PATCH 4/4] Add print messages --- tensilelite/Tensile/Toolchain/Assembly.py | 7 +++++++ 1 file changed, 7 insertions(+) diff --git a/tensilelite/Tensile/Toolchain/Assembly.py b/tensilelite/Tensile/Toolchain/Assembly.py index 4f3db710b7..81c127f15f 100644 --- a/tensilelite/Tensile/Toolchain/Assembly.py +++ b/tensilelite/Tensile/Toolchain/Assembly.py @@ -26,6 +26,7 @@ import math import shutil import subprocess +import os from pathlib import Path from typing import List, Union, NamedTuple @@ -122,6 +123,12 @@ def buildAssemblyCodeObjectFiles( linker(objFiles, str(coFileRaw)) coFile = destDir / coFileRaw.name.replace(extCoRaw, extCo) if compress: + if os.path.exists(str(coFileRaw)): + print(f"exist") + print(str(coFileRaw)) + else: + print(f"non exist") + print(str(coFileRaw)) bundler.compress(str(coFileRaw), str(coFile), gfx) else: shutil.move(coFileRaw, coFile)