diff --git a/projects/hipblas/CMakeLists.txt b/projects/hipblas/CMakeLists.txt index 4ed583e64d6a..87bb9b4ceda5 100644 --- a/projects/hipblas/CMakeLists.txt +++ b/projects/hipblas/CMakeLists.txt @@ -77,37 +77,25 @@ endif( ) # BUILD_SHARED_LIBS is a cmake built-in; we make it an explicit option such that it shows in cmake-gui option( BUILD_SHARED_LIBS "Build hipBLAS as a shared library" ON ) -# Deprecated USE_CUDA option -if(DEFINED USE_CUDA) - if(USE_CUDA) - set(ENV{HIP_PLATFORM} nvidia) - message(DEPRECATED "USE_CUDA is deprecated (use environment variable HIP_PLATFORM=nvidia)") - else() - set(ENV{HIP_PLATFORM} amd) - message(DEPRECATION "USE_CUDA is deprecated (use environment variable HIP_PLATFORM=amd)") - endif() -endif() - -# Hip headers required of all clients; clients use hip to allocate device memory -find_package( hip CONFIG PATHS ${HIP_DIR} ${ROCM_PATH} /opt/rocm ) - -# support for cuda backend with hip < 6.0 -if(NOT hip_FOUND) - find_package( HIP MODULE REQUIRED ) - list( APPEND HIP_INCLUDE_DIRS "${HIP_ROOT_DIR}/include" ) - set(ENV{HIP_PLATFORM} nvidia) - if (NOT HIP_PLATFORM) - set(HIP_PLATFORM nvidia) - endif() +# USE_CUDA: build hipBLAS for the CUDA backend using nvcc directly. +# Pass -DUSE_CUDA=ON to cmake (or --cuda to rmake.py/install.sh) to enable. +option(USE_CUDA "Build hipBLAS for the CUDA backend using nvcc directly" OFF) + +# Define the platform before loading HIP's CMake package so it does not need to +# invoke hipconfig to select a backend. The HIP headers and hip::host target are +# required by the public hipBLAS headers on both backends. +if(USE_CUDA) + set(HIP_PLATFORM nvidia) else() - if (NOT HIP_PLATFORM) - set(HIP_PLATFORM amd) - endif() + set(HIP_PLATFORM amd) endif() +find_package(hip CONFIG REQUIRED PATHS ${HIP_DIR} ${ROCM_PATH} /opt/rocm) -if(HIP_PLATFORM STREQUAL nvidia) - find_package( CUDA REQUIRED ) - set( HIPBLAS_HIP_PLATFORM_COMPILER_DEFINES __HIP_PLATFORM_NVCC__ __HIP_PLATFORM_NVIDIA__ ) +if(USE_CUDA) + # CUDA backend: use the CUDAToolkit module (requires nvcc on PATH or + # CMAKE_CUDA_COMPILER set by the caller). No hipcc involvement. + find_package(CUDAToolkit REQUIRED) + set(HIPBLAS_CUDA_COMPILER_DEFINES __HIP_PLATFORM_NVCC__ __HIP_PLATFORM_NVIDIA__) endif() option(BUILD_CODE_COVERAGE "Build with code coverage enabled" OFF) @@ -198,7 +186,7 @@ if( WIN32 ) endif( ) # Package specific CPACK vars -if(HIP_PLATFORM STREQUAL amd) +if(NOT USE_CUDA) set(rocblas_minimum 5.7.0) set(rocsolver_minimum 3.33.0) rocm_package_add_dependencies(SHARED_DEPENDS "rocblas >= ${rocblas_minimum}" "rocsolver >= ${rocsolver_minimum}") @@ -237,7 +225,7 @@ endif( ) set( CPACK_RPM_EXCLUDE_FROM_AUTO_FILELIST_ADDITION "\${CPACK_PACKAGING_INSTALL_PREFIX}" ) # Give hipblas compiled for CUDA backend a different name -if(HIP_PLATFORM STREQUAL amd) +if(NOT USE_CUDA) set( package_name hipblas ) else( ) set( package_name hipblas-alt ) diff --git a/projects/hipblas/clients/CMakeLists.txt b/projects/hipblas/clients/CMakeLists.txt index ffb6ef325553..acce051b393a 100644 --- a/projects/hipblas/clients/CMakeLists.txt +++ b/projects/hipblas/clients/CMakeLists.txt @@ -233,7 +233,7 @@ if( BUILD_CLIENTS_BENCHMARKS OR BUILD_CLIENTS_TESTS) list( APPEND COMMON_LINK_LIBS "Threads::Threads") endif() - set( COMMON_DEFINES HIPBLAS_USE_HIP_HALF HIPBLAS_BFLOAT16_CLASS ROCM_USE_FLOAT16 HIPBLAS_NO_DEPRECATED_WARNINGS ${HIPBLAS_HIP_PLATFORM_COMPILER_DEFINES} ) + set( COMMON_DEFINES HIPBLAS_USE_HIP_HALF HIPBLAS_BFLOAT16_CLASS ROCM_USE_FLOAT16 HIPBLAS_NO_DEPRECATED_WARNINGS ${HIPBLAS_CUDA_COMPILER_DEFINES} ) if (BUILD_FORTRAN_CLIENTS) list( APPEND COMMON_DEFINES "HIPBLAS_FORTRAN_CLIENTS" ) endif() diff --git a/projects/hipblas/clients/benchmarks/CMakeLists.txt b/projects/hipblas/clients/benchmarks/CMakeLists.txt index a9adb3e1669a..d8e00cc081b6 100644 --- a/projects/hipblas/clients/benchmarks/CMakeLists.txt +++ b/projects/hipblas/clients/benchmarks/CMakeLists.txt @@ -81,7 +81,7 @@ if (NOT WIN32) target_link_libraries( hipblas-bench PRIVATE stdc++fs ) endif() -if(HIP_PLATFORM STREQUAL amd) +if(NOT USE_CUDA) target_link_libraries( hipblas-bench PRIVATE hip::host ) if( CUSTOM_TARGET ) @@ -89,12 +89,7 @@ if(HIP_PLATFORM STREQUAL amd) endif() else( ) - target_include_directories( hipblas-bench - PRIVATE - $ - ) - - target_link_libraries( hipblas-bench PRIVATE ${CUDA_LIBRARIES} ) + target_link_libraries( hipblas-bench PRIVATE hip::host CUDA::cudart ) endif( ) set_target_properties( hipblas-bench PROPERTIES diff --git a/projects/hipblas/clients/cmake/build-options.cmake b/projects/hipblas/clients/cmake/build-options.cmake index 20f6916ff112..4bab5dc60261 100644 --- a/projects/hipblas/clients/cmake/build-options.cmake +++ b/projects/hipblas/clients/cmake/build-options.cmake @@ -32,7 +32,7 @@ if( NOT CREATE_TEST_APP_LOCAL_DEPLOY ) option( CREATE_TEST_APP_LOCAL_DEPLOY "Create test sandbox for app local deployment" OFF ) endif( ) -if( HIP_PLATFORM STREQUAL nvidia ) +if( USE_CUDA ) option( LINK_BLIS "Link AOCL Blis reference library" OFF ) else() option( LINK_BLIS "Link AOCL Blis reference library" ON ) diff --git a/projects/hipblas/clients/gtest/CMakeLists.txt b/projects/hipblas/clients/gtest/CMakeLists.txt index 9108333e161d..3779a3294e09 100644 --- a/projects/hipblas/clients/gtest/CMakeLists.txt +++ b/projects/hipblas/clients/gtest/CMakeLists.txt @@ -166,7 +166,7 @@ if (NOT WIN32) target_link_libraries( hipblas-test PRIVATE stdc++fs ) endif() -if(HIP_PLATFORM STREQUAL amd) +if(NOT USE_CUDA) target_link_libraries( hipblas-test PRIVATE hip::host ) if( CUSTOM_TARGET ) @@ -174,12 +174,7 @@ if(HIP_PLATFORM STREQUAL amd) endif( ) else( ) - target_include_directories( hipblas-test - PRIVATE - $ - ) - - target_link_libraries( hipblas-test PRIVATE ${CUDA_LIBRARIES} ) + target_link_libraries( hipblas-test PRIVATE hip::host CUDA::cudart ) endif( ) if (CREATE_TEST_APP_LOCAL_DEPLOY AND WIN32) diff --git a/projects/hipblas/clients/samples/CMakeLists.txt b/projects/hipblas/clients/samples/CMakeLists.txt index 40a7f3204b18..c36a576101f6 100644 --- a/projects/hipblas/clients/samples/CMakeLists.txt +++ b/projects/hipblas/clients/samples/CMakeLists.txt @@ -42,7 +42,7 @@ if(NOT WIN32) set( sample_list_fortran hipblas-example-sscal-fortran hipblas-example-gemmEx-fortran ) endif() -if(HIP_PLATFORM STREQUAL amd) +if(NOT USE_CUDA) add_executable(hipblas-example-bfdot-hip-bfloat16 example_bfdot_hip_bfloat16.cpp ${hipblas_samples_common}) endif() @@ -66,7 +66,7 @@ list (APPEND hipblas-example-executables hipblas-example-sscal hipblas-example-s if( CMAKE_CXX_COMPILER MATCHES ".*/hipcc$" OR CMAKE_CXX_COMPILER MATCHES ".*/amdclang\\+\\+$") list (APPEND hipblas-example-executables hipblas-example-hgemm) endif( ) -if(HIP_PLATFORM STREQUAL amd) +if(NOT USE_CUDA) list (APPEND hipblas-example-executables hipblas-example-bfdot-hip-bfloat16) target_compile_options(hipblas-example-bfdot-hip-bfloat16 PRIVATE -DHIPBLAS_USE_HIP_BFLOAT16) endif() @@ -99,7 +99,7 @@ foreach( exe IN LISTS hipblas-example-executables) target_link_libraries( ${exe} PRIVATE roc::hipblas ) - if(HIP_PLATFORM STREQUAL amd) + if(NOT USE_CUDA) target_link_libraries( ${exe} PRIVATE hip::host ) if( CUSTOM_TARGET ) @@ -107,14 +107,8 @@ foreach( exe IN LISTS hipblas-example-executables) endif( ) else( ) - target_compile_definitions( ${exe} PRIVATE ${HIPBLAS_HIP_PLATFORM_COMPILER_DEFINES} ) - - target_include_directories( ${exe} - PRIVATE - $ - ) - - target_link_libraries( ${exe} PRIVATE ${CUDA_LIBRARIES} ) + target_compile_definitions( ${exe} PRIVATE ${HIPBLAS_CUDA_COMPILER_DEFINES} ) + target_link_libraries( ${exe} PRIVATE hip::host CUDA::cudart ) endif( ) set_target_properties( ${exe} PROPERTIES diff --git a/projects/hipblas/docs/install/Linux_Install_Guide.rst b/projects/hipblas/docs/install/Linux_Install_Guide.rst index 941c36a85c57..5f9213de0c94 100644 --- a/projects/hipblas/docs/install/Linux_Install_Guide.rst +++ b/projects/hipblas/docs/install/Linux_Install_Guide.rst @@ -92,8 +92,10 @@ A few commands in the script require ``sudo`` access, which might prompt you for You can run the ``rmake.py`` script from the ``projects/hipblas`` directory. -The install script determines the build platform by querying ``hipconfig --platform``. This value can be explicitly defined -by setting the environment variable ``HIP_PLATFORM`` to ``HIP_PLATFORM=amd`` or ``HIP_PLATFORM=nvidia``. +To build for the CUDA backend, pass ``--cuda`` to ``rmake.py`` or ``install.sh``. +This sets ``-DUSE_CUDA=ON`` in CMake and invokes ``nvcc`` directly without +relying on ``hipcc``. ``nvcc`` must be on ``PATH`` or the CUDA install location +specified via ``--cudapath``. Common examples showing how to use ``rmake.py`` to build the library dependencies and library are listed in this table. diff --git a/projects/hipblas/install.sh b/projects/hipblas/install.sh index 1b214c77b92e..ae4ac47c0e83 100755 --- a/projects/hipblas/install.sh +++ b/projects/hipblas/install.sh @@ -168,7 +168,7 @@ install_packages( ) local library_dependencies_fedora=( "make" "gcc-c++" "libcxx-devel" "rpm-build" ) local library_dependencies_sles=( "make" "gcc-c++" "rpm-build" ) - if [[ $HIP_PLATFORM == "nvidia" ]]; then + if [[ "${build_cuda}" == true ]]; then # Ideally, this could be cuda-cublas-dev, but the package name has a version number in it library_dependencies_ubuntu+=( "" ) # removed, use --installcuda option to install cuda else @@ -360,9 +360,10 @@ cat < - ) + # hipBLAS' public headers expose HIP and cuBLAS declarations on this backend, + # so propagate their usage requirements to clients of roc::hipblas. + target_link_libraries( hipblas PUBLIC CUDA::cublas CUDA::cudart ) endif( ) # External header includes included as system files @@ -179,7 +174,7 @@ rocm_install_targets( ) # PERMISSIONS OWNER_EXECUTE OWNER_WRITE OWNER_READ GROUP_EXECUTE GROUP_READ WORLD_EXECUTE WORLD_READ -if(HIP_PLATFORM STREQUAL amd ) +if(NOT USE_CUDA ) rocm_export_targets( TARGETS roc::hipblas DEPENDS PACKAGE hip @@ -191,6 +186,7 @@ else( ) rocm_export_targets( TARGETS roc::hipblas DEPENDS PACKAGE HIP + DEPENDS PACKAGE CUDAToolkit DEPENDS PACKAGE hipblas-common NAMESPACE roc:: ) diff --git a/projects/hipblas/rmake.py b/projects/hipblas/rmake.py index 0ae6238f12d7..36674b9a354e 100755 --- a/projects/hipblas/rmake.py +++ b/projects/hipblas/rmake.py @@ -73,7 +73,8 @@ def parse_args(): help='Specify path to host compiler. Default is amdclang++.') parser.add_argument('--cuda', '--use-cuda', dest='use_cuda', required=False, default=False, action='store_true', - help='[DEPRECATED] Build library for CUDA backend. Deprecated, use HIP_PLATFORM environment variable to override default which is determined by `hipconfig --platform`') + help='Build library for CUDA backend using nvcc directly (sets -DUSE_CUDA=ON). ' + 'nvcc must be on PATH or specified via --cudapath.') parser.add_argument( '--cudapath', type=str, required=False, default='/usr/local/cuda', dest='cuda_path', help='Specify path of CUDA install.') @@ -280,9 +281,11 @@ def config_cmd(): if args.compiler is not None: cmake_options.append(f"-DCMAKE_CXX_COMPILER={args.compiler}") - # amdclang++ default in linux toolchain, clang++ in windows toolchain + # amdclang++ default in linux toolchain, clang++ in windows toolchain. + # Skip when targeting CUDA: nvcc drives device compilation and g++ is not + # the right host compiler for CUDA workflows. if os.name != "nt" and args.compiler is None: - if not args.use_amdclang_compiler: + if not args.use_amdclang_compiler and not args.use_cuda: cmake_options.append(f"-DCMAKE_CXX_COMPILER=g++") if args.static_lib: @@ -337,8 +340,19 @@ def config_cmd(): rocsolver_path_cmake = f'"{raw_rocsolver_path}"' cmake_options.append( f"-DCUSTOM_ROCSOLVER={rocsolver_path_cmake}") - if args.cuda_path: - os.environ['CUDA_BIN_PATH'] = args.cuda_path + # CUDA backend: pass USE_CUDA=ON and locate nvcc so CMake's find_package(CUDAToolkit) + # resolves without any hipcc involvement. + if args.use_cuda: + cmake_options.append("-DUSE_CUDA=ON") + cuda_path = args.cuda_path # defaults to /usr/local/cuda + nvcc_candidate = os.path.join(cuda_path, 'bin', 'nvcc') + if not os.path.exists(nvcc_candidate): + nvcc_candidate = which('nvcc') or '' + if not nvcc_candidate: + fatal("nvcc not found. Install the CUDA toolkit or pass " + "--cudapath=/path/to/cuda to specify its location.") + cmake_options.append(f"-DCUDA_TOOLKIT_ROOT_DIR={cuda_path}") + cmake_options.append(f"-DCMAKE_CUDA_COMPILER={nvcc_candidate}") if args.cmake_dargs: for i in args.cmake_dargs: @@ -385,22 +399,7 @@ def main(): os_detect() args = parse_args() - rocm_path = get_rocm_path() - - # support --cuda flag for now by overwriting HIP_PLATFORM envvar - if args.use_cuda: - os.environ['HIP_PLATFORM'] = 'nvidia' - print("--cuda option is deprecated (use environment variable HIP_PLATFORM=nvidia)") - - # otherwise query hipconfig to get platform - if which('hipconfig') is not None: - os.environ['HIP_PLATFORM'] = subprocess.getoutput('hipconfig --platform') - elif which(f'{rocm_path}/bin/hipconfig') is not None: - os.environ['HIP_PLATFORM'] = subprocess.getoutput(f'{rocm_path}/bin/hipconfig --platform') - else: - os.environ['HIP_PLATFORM'] = 'amd' # if can't find hipconfig, default amd - - if os.environ['HIP_PLATFORM'] == 'nvidia' and args.static_lib: + if args.use_cuda and args.static_lib: fatal("Static library not supported for CUDA backend. Not continuing.") if args.install_invoked: diff --git a/projects/hipblaslt/cmake/hipblaslt_python.cmake b/projects/hipblaslt/cmake/hipblaslt_python.cmake index 658f50fe23cb..0eb4c71e01fe 100644 --- a/projects/hipblaslt/cmake/hipblaslt_python.cmake +++ b/projects/hipblaslt/cmake/hipblaslt_python.cmake @@ -38,6 +38,7 @@ function(hipblaslt_configure_bundled_python_command python_binary_dir asan_optio "${CMAKE_COMMAND}" -E env "PYTHONPATH=${_python_path}" "PATH=${_path}" + "ROCM_VERSION=${hip_VERSION}" "${asan_options}" -- "${Python3_EXECUTABLE}" diff --git a/projects/hipblaslt/tensilelite/Tensile/TensileCreateLibrary/Run.py b/projects/hipblaslt/tensilelite/Tensile/TensileCreateLibrary/Run.py index b021f212ac6c..4023149a5728 100644 --- a/projects/hipblaslt/tensilelite/Tensile/TensileCreateLibrary/Run.py +++ b/projects/hipblaslt/tensilelite/Tensile/TensileCreateLibrary/Run.py @@ -1027,12 +1027,11 @@ def run(): arguments = parseArguments() setVerbosity(arguments["PrintLevel"]) outputPath = Path(ensurePath(os.path.abspath(arguments["OutputPath"]))) - cxxCompiler, _, offloadBundler, _, _ = validateToolchain( + cxxCompiler, _, offloadBundler, _ = validateToolchain( arguments["CxxCompiler"], arguments["CCompiler"], arguments["OffloadBundler"], arguments["Assembler"], - ToolchainDefaults.HIP_CONFIG, ) if ";" in arguments["Architecture"]: diff --git a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/PublicInputSurfaceDeep/test_pchaos_Validators_L226_char.py b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/PublicInputSurfaceDeep/test_pchaos_Validators_L226_char.py index ba87753c5dad..58303b79beba 100644 --- a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/PublicInputSurfaceDeep/test_pchaos_Validators_L226_char.py +++ b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/PublicInputSurfaceDeep/test_pchaos_Validators_L226_char.py @@ -35,7 +35,6 @@ supportedCCompiler, supportedCxxCompiler, supportedDeviceEnumerator, - supportedHip, supportedOffloadBundler, ) @@ -123,7 +122,6 @@ def test_real_guard_true_hipinfo_posix(): supportedCxxCompiler(file), supportedCCompiler(file), supportedOffloadBundler(file), - supportedHip(file), supportedDeviceEnumerator(file), )) assert result is True, f"Expected guard=True for {file!r} on POSIX, got False" @@ -136,7 +134,6 @@ def test_real_guard_true_gcc_posix(): supportedCxxCompiler(file), supportedCCompiler(file), supportedOffloadBundler(file), - supportedHip(file), supportedDeviceEnumerator(file), )) assert result is True, f"Expected guard=True for {file!r} on POSIX, got False" @@ -149,7 +146,6 @@ def test_real_guard_false_amdclang_plus_plus_posix(): supportedCxxCompiler(file), supportedCCompiler(file), supportedOffloadBundler(file), - supportedHip(file), supportedDeviceEnumerator(file), )) assert result is False, f"Expected guard=False for {file!r} on POSIX, got True" @@ -162,7 +158,6 @@ def test_real_guard_false_amdgpu_arch_posix(): supportedCxxCompiler(file), supportedCCompiler(file), supportedOffloadBundler(file), - supportedHip(file), supportedDeviceEnumerator(file), )) assert result is False, f"Expected guard=False for {file!r} on POSIX, got True" @@ -180,7 +175,6 @@ def test_helper_agrees_with_real_guard_all_witnesses(): supportedCxxCompiler(file), supportedCCompiler(file), supportedOffloadBundler(file), - supportedHip(file), supportedDeviceEnumerator(file), )) helper = toolchain_component_rejected(file) diff --git a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/TensileCreateLibraryRun/test_r7_createlib_deep_char.py b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/TensileCreateLibraryRun/test_r7_createlib_deep_char.py index edd00cfabf77..058c486ae5dd 100644 --- a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/TensileCreateLibraryRun/test_r7_createlib_deep_char.py +++ b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/TensileCreateLibraryRun/test_r7_createlib_deep_char.py @@ -654,7 +654,7 @@ def _run_with_stubs(self, logic_dir, output_dir, extra_args=None): with patch.object(M, "parseArguments", return_value=base_args), \ patch.object(M, "setVerbosity"), \ patch.object(M, "validateToolchain", return_value=( - "/fake/hipcc", None, "/fake/bundler", None, None)), \ + "/fake/hipcc", None, "/fake/bundler", None)), \ patch.object(M, "makeIsaInfoMap", return_value=info_map), \ patch.object(M, "assignGlobalParameters"), \ patch.object(M, "makeAssemblyToolchain", return_value=MagicMock()), \ @@ -694,7 +694,7 @@ def _spy_write(name, data, *a, **kw): with patch.object(M, "parseArguments", return_value=base_args), \ patch.object(M, "setVerbosity"), \ patch.object(M, "validateToolchain", return_value=( - "/fake/hipcc", None, "/fake/bundler", None, None)), \ + "/fake/hipcc", None, "/fake/bundler", None)), \ patch.object(M, "makeIsaInfoMap", return_value=info_map), \ patch.object(M, "assignGlobalParameters"), \ patch.object(M, "makeAssemblyToolchain", return_value=MagicMock()), \ @@ -733,7 +733,7 @@ def _spy_write(name, data, *a, **kw): with patch.object(M, "parseArguments", return_value=base_args), \ patch.object(M, "setVerbosity"), \ patch.object(M, "validateToolchain", return_value=( - "/fake/hipcc", None, "/fake/bundler", None, None)), \ + "/fake/hipcc", None, "/fake/bundler", None)), \ patch.object(M, "makeIsaInfoMap", return_value=info_map), \ patch.object(M, "assignGlobalParameters"), \ patch.object(M, "makeAssemblyToolchain", return_value=MagicMock()), \ @@ -773,7 +773,7 @@ def _spy_write(name, data, fmt=None, *a, **kw): with patch.object(M, "parseArguments", return_value=base_args), \ patch.object(M, "setVerbosity"), \ patch.object(M, "validateToolchain", return_value=( - "/fake/hipcc", None, "/fake/bundler", None, None)), \ + "/fake/hipcc", None, "/fake/bundler", None)), \ patch.object(M, "makeIsaInfoMap", return_value=info_map), \ patch.object(M, "assignGlobalParameters"), \ patch.object(M, "makeAssemblyToolchain", return_value=MagicMock()), \ @@ -802,7 +802,7 @@ def test_run_logic_path_not_exist_exits(self, tmp_path): with patch.object(M, "parseArguments", return_value=base_args), \ patch.object(M, "setVerbosity"), \ patch.object(M, "validateToolchain", return_value=( - "/fake/hipcc", None, "/fake/bundler", None, None)), \ + "/fake/hipcc", None, "/fake/bundler", None)), \ patch.object(M, "makeIsaInfoMap", return_value=info_map), \ patch.object(M, "assignGlobalParameters"), \ patch.object(M, "makeAssemblyToolchain", return_value=MagicMock()), \ @@ -819,7 +819,7 @@ def test_run_semicolon_arch_is_split(self, logic_dir, output_dir): with patch.object(M, "parseArguments", return_value=base_args), \ patch.object(M, "setVerbosity"), \ patch.object(M, "validateToolchain", return_value=( - "/fake/hipcc", None, "/fake/bundler", None, None)), \ + "/fake/hipcc", None, "/fake/bundler", None)), \ patch.object(M, "makeIsaInfoMap", return_value=info_map), \ patch.object(M, "assignGlobalParameters"), \ patch.object(M, "makeAssemblyToolchain", return_value=MagicMock()), \ @@ -851,7 +851,7 @@ def test_run_keep_build_tmp_suppresses_rmtree(self, logic_dir, output_dir): with patch.object(M, "parseArguments", return_value=base_args), \ patch.object(M, "setVerbosity"), \ patch.object(M, "validateToolchain", return_value=( - "/fake/hipcc", None, "/fake/bundler", None, None)), \ + "/fake/hipcc", None, "/fake/bundler", None)), \ patch.object(M, "makeIsaInfoMap", return_value=info_map), \ patch.object(M, "assignGlobalParameters"), \ patch.object(M, "makeAssemblyToolchain", return_value=MagicMock()), \ @@ -909,7 +909,7 @@ def _spy_write(name, data, fmt=None, *a, **kw): with patch.object(M, "parseArguments", return_value=base_args), \ patch.object(M, "setVerbosity"), \ patch.object(M, "validateToolchain", return_value=( - "/fake/hipcc", None, "/fake/bundler", None, None)), \ + "/fake/hipcc", None, "/fake/bundler", None)), \ patch.object(M, "makeIsaInfoMap", return_value=info_map), \ patch.object(M, "assignGlobalParameters"), \ patch.object(M, "makeAssemblyToolchain", return_value=MagicMock()), \ diff --git a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/ToolchainComponent/test_toolchain_component_char.py b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/ToolchainComponent/test_toolchain_component_char.py index 560a3347e7de..e06a8acc607d 100644 --- a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/ToolchainComponent/test_toolchain_component_char.py +++ b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/ToolchainComponent/test_toolchain_component_char.py @@ -79,45 +79,92 @@ class _R: C._getVersion("amdclang++", "--version", r"version\s+([\d.]+)") -def test_get_rocm_version_uses_hipconfig(monkeypatch): - seen = {} - - def _fake(exe, flag, regex): - seen["exe"], seen["flag"] = exe, flag - return SemanticVersion(6, 4, 0) - - monkeypatch.setattr(C, "_getVersion", _fake) +def test_get_rocm_version_uses_rocm_version_env(monkeypatch, tmp_path): + """ROCM_VERSION env var takes priority over all other sources.""" + rocm_info = tmp_path / "rocm" / ".info" + hip_info = tmp_path / "hip" / ".info" + rocm_info.mkdir(parents=True) + hip_info.mkdir(parents=True) + (rocm_info / "version").write_text("7.1.0") + (hip_info / "version").write_text("7.2.0") + monkeypatch.setenv("ROCM_VERSION", "6.4.0") + monkeypatch.setenv("ROCM_PATH", str(rocm_info.parent)) + monkeypatch.setenv("HIP_PATH", str(hip_info.parent)) assert C.get_rocm_version() == SemanticVersion(6, 4, 0) - assert seen["flag"] == "--version" - assert seen["exe"] == C.ToolchainDefaults.HIP_CONFIG @pytest.mark.parametrize( - "hipconfig_output, expected_version", + "version_str, expected_version", [ pytest.param( - b"7.1.25424-4179531dcd", + "7.1.0", + SemanticVersion(7, 1, 0), + id="simple_version", + ), + pytest.param( + "7.1.25424-4179531dcd", SemanticVersion(7, 1, 25424), - id="rocm_7_1_build_suffix", + id="build_suffix", + ), + pytest.param( + "10.1.0a20260813", + SemanticVersion(10, 1, 0), + id="alpha_suffix", ), pytest.param( - b"7.2.26015-fc0010cf6a", - SemanticVersion(7, 2, 26015), - id="rocm_7_2_build_suffix", + "7.2.1rc2", + SemanticVersion(7, 2, 1), + id="release_candidate_suffix", ), ], ) -def test_get_rocm_version_parses_hipconfig_build_suffix( - monkeypatch, hipconfig_output, expected_version +def test_get_rocm_version_reads_info_version_file( + monkeypatch, tmp_path, version_str, expected_version ): - monkeypatch.setattr(C, "validateToolchain", lambda x: x) - - class _R: - stdout = hipconfig_output + """get_rocm_version reads .info/version when ROCM_VERSION is not set.""" + info_dir = tmp_path / ".info" + info_dir.mkdir() + (info_dir / "version").write_text(version_str) + monkeypatch.delenv("ROCM_VERSION", raising=False) + monkeypatch.setenv("ROCM_PATH", str(tmp_path)) + assert C.get_rocm_version() == expected_version - monkeypatch.setattr(C, "run", lambda *a, **k: _R()) - assert C.get_rocm_version() == expected_version +def test_get_rocm_version_prefers_rocm_path_over_hip_path(monkeypatch, tmp_path): + """ROCM_PATH takes priority over HIP_PATH.""" + rocm_info = tmp_path / "rocm" / ".info" + hip_info = tmp_path / "hip" / ".info" + rocm_info.mkdir(parents=True) + hip_info.mkdir(parents=True) + (rocm_info / "version").write_text("7.1.0") + (hip_info / "version").write_text("7.2.0") + monkeypatch.delenv("ROCM_VERSION", raising=False) + monkeypatch.setenv("ROCM_PATH", str(rocm_info.parent)) + monkeypatch.setenv("HIP_PATH", str(hip_info.parent)) + assert C.get_rocm_version() == SemanticVersion(7, 1, 0) + + +def test_get_rocm_version_falls_back_from_rocm_path_to_hip_path(monkeypatch, tmp_path): + """An unreadable ROCM_PATH falls through to HIP_PATH.""" + hip_root = tmp_path / "hip" + info_dir = hip_root / ".info" + info_dir.mkdir(parents=True) + (info_dir / "version").write_text("7.2.3") + monkeypatch.delenv("ROCM_VERSION", raising=False) + monkeypatch.setenv("ROCM_PATH", str(tmp_path / "missing")) + monkeypatch.setenv("HIP_PATH", str(hip_root)) + assert C.get_rocm_version() == SemanticVersion(7, 2, 3) + + +def test_get_rocm_version_raises_when_no_source_exists(monkeypatch, tmp_path): + """Missing environment and installation files report the documented error.""" + real_path = C.Path + monkeypatch.delenv("ROCM_VERSION", raising=False) + monkeypatch.delenv("ROCM_PATH", raising=False) + monkeypatch.delenv("HIP_PATH", raising=False) + monkeypatch.setattr(C, "Path", lambda _root: real_path(tmp_path / "missing")) + with pytest.raises(RuntimeError, match="Failed to get ROCm version"): + C.get_rocm_version() # --------------------------------------------------------------------------- diff --git a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/ToolchainValidators/test_toolchain_validators_char.py b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/ToolchainValidators/test_toolchain_validators_char.py index ac293fb57bff..3931abe2a110 100644 --- a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/ToolchainValidators/test_toolchain_validators_char.py +++ b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/characterization/ToolchainValidators/test_toolchain_validators_char.py @@ -35,9 +35,6 @@ ("supportedCCompiler", "amdclang++", False), ("supportedOffloadBundler", "clang-offload-bundler", True), ("supportedOffloadBundler", "clang", False), - ("supportedHip", "hipcc", True), - ("supportedHip", "hipconfig", True), - ("supportedHip", "hipcc.exe", False), ], ) def test_supported_predicates(fn, name, expected): @@ -180,7 +177,6 @@ def test_toolchain_defaults_posix(): assert d.C_COMPILER == "amdclang" assert d.OFFLOAD_BUNDLER == "clang-offload-bundler" assert d.ASSEMBLER == "amdclang++" - assert d.HIP_CONFIG == "hipconfig" assert d.DEVICE_ENUMERATOR in ("rocm_agent_enumerator", "amdgpu-arch") diff --git a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/test_gfx1250_asic_revision.py b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/test_gfx1250_asic_revision.py index e1cf92b7065b..d362adb9ba53 100644 --- a/projects/hipblaslt/tensilelite/Tensile/Tests/unit/test_gfx1250_asic_revision.py +++ b/projects/hipblaslt/tensilelite/Tensile/Tests/unit/test_gfx1250_asic_revision.py @@ -1204,7 +1204,7 @@ def _capture_archs(*args, **kwargs): monkeypatch.setattr( RunModule, "validateToolchain", - lambda *a: ("/fake/hipcc", None, "/fake/bundler", None, None), + lambda *a: ("/fake/hipcc", None, "/fake/bundler", None), ) monkeypatch.setattr(RunModule, "makeIsaInfoMap", lambda _isas, _cxx: _stub_iim()) monkeypatch.setattr(RunModule, "assignGlobalParameters", _capture_gp) @@ -1826,7 +1826,7 @@ def _capture_write(filename, *a, **kw): monkeypatch.setattr( RunModule, "validateToolchain", - lambda *a: ("/fake/hipcc", None, "/fake/bundler", None, None), + lambda *a: ("/fake/hipcc", None, "/fake/bundler", None), ) monkeypatch.setattr(RunModule, "makeIsaInfoMap", _iim) monkeypatch.setattr(RunModule, "assignGlobalParameters", lambda *a, **kw: None) diff --git a/projects/hipblaslt/tensilelite/Tensile/Toolchain/Component.py b/projects/hipblaslt/tensilelite/Tensile/Toolchain/Component.py index fd5a986f444a..60942d46219e 100644 --- a/projects/hipblaslt/tensilelite/Tensile/Toolchain/Component.py +++ b/projects/hipblaslt/tensilelite/Tensile/Toolchain/Component.py @@ -25,7 +25,7 @@ from os import name as os_name from os import environ from pathlib import Path -from re import search, IGNORECASE +from re import match, search, IGNORECASE from shlex import split from subprocess import check_output, STDOUT, CalledProcessError, PIPE, run from typing import List @@ -79,15 +79,38 @@ def _getVersion(executable: str, versionFlag: str, regex: str) -> str: raise RuntimeError(f"Failed to get version when calling {args}: {e}") -def get_rocm_version() -> str: - """Compute the ROCm version string using hipconfig. +def get_rocm_version() -> SemanticVersion: + """Compute the ROCm version string. + + Reads ROCM_VERSION env var (set by CMake from hip_VERSION) if available, + otherwise falls back to reading .info/version from ROCM_PATH, HIP_PATH, + or /opt/rocm. + + Note: Python ROCm SDK (pip) version detection is intentionally omitted here; + it will be handled by PR #11023 with proper feature-gating once the + infrastructure is in place. Raises: - RuntimeError: If hipconfig fails to execute. + RuntimeError: If the version cannot be determined. Return: - ROCm version string + ROCm SemanticVersion """ - return _getVersion(ToolchainDefaults.HIP_CONFIG, "--version", r'(.+)') + version_str = environ.get("ROCM_VERSION") + if not version_str: + for root in [environ.get("ROCM_PATH"), environ.get("HIP_PATH"), "/opt/rocm"]: + if root: + try: + version_str = (Path(root) / ".info" / "version").read_text().strip() + break + except OSError: + continue + if not version_str: + raise RuntimeError("Failed to get ROCm version: ROCM_VERSION not set and " + ".info/version not found in ROCM_PATH, HIP_PATH, or /opt/rocm") + # Strip pre-release suffixes before parsing (e.g. "0a20260813" -> "0", "1rc2" -> "1") + # so that nightly/alpha version strings like "10.1.0a20260813" parse correctly. + return SemanticVersion(*[int(match(r'\d+', c.split("-")[0]).group()) + for c in version_str.split(".")[:3]]) class Component: diff --git a/projects/hipblaslt/tensilelite/Tensile/Toolchain/Validators.py b/projects/hipblaslt/tensilelite/Tensile/Toolchain/Validators.py index 75f786868f14..78caa49e98d4 100644 --- a/projects/hipblaslt/tensilelite/Tensile/Toolchain/Validators.py +++ b/projects/hipblaslt/tensilelite/Tensile/Toolchain/Validators.py @@ -120,7 +120,6 @@ class ToolchainDefaults(NamedTuple): OFFLOAD_BUNDLER = osSelect(linux="clang-offload-bundler", windows="clang-offload-bundler.exe") DEVICE_ENUMERATOR = osSelect(linux="rocm_agent_enumerator" if isRhel8() or inFFMEnv else "amdgpu-arch", windows="hipinfo") ASSEMBLER = osSelect(linux="amdclang++", windows="clang++.exe") - HIP_CONFIG = osSelect(linux="hipconfig", windows="hipconfig.exe") def _supportedComponent(component: str, targets: List[str]) -> bool: @@ -169,18 +168,6 @@ def supportedOffloadBundler(bundler: str) -> bool: return _supportedComponent(bundler, ["clang-offload-bundler"]) -def supportedHip(hip: str) -> bool: - """ - Determine if a hip callable binary is supported by Tensile. - - Args: - hip: The name of an offload bundler to test for support. - - Return: - If supported True; otherwise, False. - """ - return _supportedComponent(hip, ["hipcc", "hipconfig"]) - def supportedDeviceEnumerator(enumerator: str) -> bool: """ @@ -227,7 +214,6 @@ def _validateExecutable(file: str, searchPaths: List[Path]) -> str: supportedCxxCompiler(file), supportedCCompiler(file), supportedOffloadBundler(file), - supportedHip(file), supportedDeviceEnumerator(file) )): raise ValueError(f"`{file}` is not a supported toolchain component on {'Windows' if os.name == 'nt' else 'Linux'}") diff --git a/projects/hipsolver/install.sh b/projects/hipsolver/install.sh index 6d855b4e0d16..a83e386eb2cc 100755 --- a/projects/hipsolver/install.sh +++ b/projects/hipsolver/install.sh @@ -629,7 +629,7 @@ fi if [[ "${build_cuda}" == true ]]; then export HIP_PLATFORM="nvidia" else - export HIP_PLATFORM="$(${rocm_path}/bin/hipconfig --platform)" + export HIP_PLATFORM="amd" fi # clients diff --git a/shared/tensile/Tensile/AsmCaps.py b/shared/tensile/Tensile/AsmCaps.py index b3503840b40d..1fe284b9e583 100644 --- a/shared/tensile/Tensile/AsmCaps.py +++ b/shared/tensile/Tensile/AsmCaps.py @@ -27,7 +27,7 @@ def getCapabilitiesCache(rocmVersion: NamedTuple) -> dict: """Returns a compiler version dependent capabilities cache. Args: - rocmVersion: The rocm major/minor version (often obtained from hipconfig maj/min version). + rocmVersion: The rocm major/minor version. Returns: A dictionary representing the capabilities cache. diff --git a/shared/tensile/Tensile/Common.py b/shared/tensile/Tensile/Common.py index cc721bd753c9..850e15fe6660 100644 --- a/shared/tensile/Tensile/Common.py +++ b/shared/tensile/Tensile/Common.py @@ -25,10 +25,10 @@ from . import __version__ from . import Parallel from .Utilities.ConditionalImports import print, TENSILE_TERM_COLORS -from .Utilities.Toolchain import getVersion from collections import OrderedDict from copy import deepcopy +from pathlib import Path from .AsmCaps import getCapabilitiesCache from typing import Any, NamedTuple, Optional, Tuple, Dict @@ -2410,12 +2410,26 @@ def assignGlobalParameters( config, capabilitiesCache: Optional[dict] = None, *, printExit("Config file requires version=%s is not compatible with current Tensile version=%s" \ % (config["MinimumRequiredVersion"], __version__) ) - if "HipConfig" in config: - output = getVersion(config["HipConfig"], regex=r'(.+)') - globalParameters["HipClangVersion"] = output.strip() + version_str = os.environ.get("ROCM_VERSION") + if not version_str: + # Note: Python ROCm SDK (pip) version detection is intentionally omitted + # here; it will be handled by PR #11023 with proper feature-gating. + for root in [os.environ.get("ROCM_PATH"), os.environ.get("HIP_PATH"), "/opt/rocm"]: + if root: + try: + version_str = (Path(root) / ".info" / "version").read_text().strip() + break + except OSError: + continue + if version_str: + # Strip pre-release suffixes before storing (e.g. "0a20260813" -> "0") + # so that nightly version strings like "10.1.0a20260813" are handled correctly. + parts = version_str.split(".")[:3] + version_str = ".".join(re.match(r'\d+', p.split("-")[0]).group() for p in parts) + globalParameters["HipClangVersion"] = version_str tPrint(1, f"# Found HIP version: {globalParameters['HipClangVersion']}") else: - raise ValueError("HipConfig not specified in config, could not set HipClangVersion") + raise ValueError("ROCM_VERSION not set, could not set HipClangVersion") # User-specified global parameters tPrint(3, "GlobalParameters:") @@ -2529,7 +2543,6 @@ def assignGlobalParameters( config, capabilitiesCache: Optional[dict] = None, *, "Version", "ClientConfig", "WriteMasterSolutionIndex", - "HipConfig", "OffloadBundler", "Assembler", } diff --git a/shared/tensile/Tensile/Tensile.py b/shared/tensile/Tensile/Tensile.py index 0f458514d219..8689e73873fc 100644 --- a/shared/tensile/Tensile/Tensile.py +++ b/shared/tensile/Tensile/Tensile.py @@ -319,14 +319,12 @@ def Tensile(userArgs): cCompiler, assembler, offloadBundler, - hipconfig, deviceEnumerator ) = validateToolchain( args.CxxCompiler, args.CCompiler, args.Assembler, args.OffloadBundler, - ToolchainDefaults.HIP_CONFIG, ToolchainDefaults.DEVICE_ENUMERATOR ) params = config.get("GlobalParameters", {}) @@ -334,7 +332,6 @@ def Tensile(userArgs): params["CCompiler"] = cCompiler params["Assembler"] = assembler params["OffloadBundler"] = offloadBundler - params["HipConfig"] = hipconfig params["ROCmAgentEnumeratorPath"] = deviceEnumerator assignGlobalParameters(params, capabilitiesCache) diff --git a/shared/tensile/Tensile/TensileCreateLib/ParseArguments.py b/shared/tensile/Tensile/TensileCreateLib/ParseArguments.py index ee6225d59ac0..fc961bd9c630 100644 --- a/shared/tensile/Tensile/TensileCreateLib/ParseArguments.py +++ b/shared/tensile/Tensile/TensileCreateLib/ParseArguments.py @@ -339,13 +339,11 @@ def parseArguments(input: Optional[List[str]] = None) -> Dict[str, Any]: arguments["CCompiler"], arguments["Assembler"], arguments["OffloadBundler"], - arguments["HipConfig"], ) = validateToolchain( arguments["CxxCompiler"], arguments["CCompiler"], arguments["Assembler"], arguments["OffloadBundler"], - ToolchainDefaults.HIP_CONFIG, ) if args.NoEnumerate: arguments["ROCmAgentEnumeratorPath"] = False diff --git a/shared/tensile/Tensile/Tests/conftest.py b/shared/tensile/Tensile/Tests/conftest.py index dc3791df0b8e..9680cdb45188 100644 --- a/shared/tensile/Tensile/Tests/conftest.py +++ b/shared/tensile/Tensile/Tests/conftest.py @@ -144,14 +144,12 @@ def __enter__(self): cCompiler, assembler, offloadBundler, - hipconfig, deviceEnumerator ) = validateToolchain( args.CxxCompiler, args.CCompiler, args.Assembler, args.OffloadBundler, - ToolchainDefaults.HIP_CONFIG, ToolchainDefaults.DEVICE_ENUMERATOR ) params = { @@ -159,7 +157,6 @@ def __enter__(self): "CCompiler": cCompiler, "Assembler": assembler, "OffloadBundler": offloadBundler, - "HipConfig": hipconfig, "ROCmAgentEnumeratorPath": deviceEnumerator, } Common.assignGlobalParameters(params) diff --git a/shared/tensile/Tensile/Tests/unit/test_CustomKernels.py b/shared/tensile/Tensile/Tests/unit/test_CustomKernels.py index 85fc967af76e..f3dc013bf929 100644 --- a/shared/tensile/Tensile/Tests/unit/test_CustomKernels.py +++ b/shared/tensile/Tensile/Tests/unit/test_CustomKernels.py @@ -80,14 +80,12 @@ def test_CreateSolutionFromCustomKernel(objs): cCompiler, assembler, offloadBundler, - hipconfig, # deviceEnumerator ) = validateToolchain( ToolchainDefaults.CXX_COMPILER, ToolchainDefaults.C_COMPILER, ToolchainDefaults.ASSEMBLER, ToolchainDefaults.OFFLOAD_BUNDLER, - ToolchainDefaults.HIP_CONFIG, # ToolchainDefaults.DEVICE_ENUMERATOR ) params = { @@ -95,7 +93,6 @@ def test_CreateSolutionFromCustomKernel(objs): "CCompiler": cCompiler, "Assembler": assembler, "OffloadBundler": offloadBundler, - "HipConfig": hipconfig, # "ROCmAgentEnumeratorPath": deviceEnumerator, } params["ROCmAgentEnumeratorPath"] = None diff --git a/shared/tensile/Tensile/Tests/unit/test_TensileCreateLibrary.py b/shared/tensile/Tensile/Tests/unit/test_TensileCreateLibrary.py index 08a0633a83a1..e0bd6fb247f7 100644 --- a/shared/tensile/Tensile/Tests/unit/test_TensileCreateLibrary.py +++ b/shared/tensile/Tensile/Tests/unit/test_TensileCreateLibrary.py @@ -539,22 +539,18 @@ def setupSolutionsAndKernels( ) -> Tuple[List[Solution], List[Solution], KernelWriterAssembly, KernelWriterSource]: """Reusable logic for setting up testable solutions and kernels""" - (cxxCompiler, cCompiler, assembler, offloadBundler, hipconfig, deviceEnumerator) = ( - validateToolchain( - ToolchainDefaults.CXX_COMPILER, - ToolchainDefaults.C_COMPILER, - ToolchainDefaults.ASSEMBLER, - ToolchainDefaults.OFFLOAD_BUNDLER, - ToolchainDefaults.HIP_CONFIG, - ToolchainDefaults.DEVICE_ENUMERATOR, - ) + (cxxCompiler, cCompiler, assembler, offloadBundler, deviceEnumerator) = validateToolchain( + ToolchainDefaults.CXX_COMPILER, + ToolchainDefaults.C_COMPILER, + ToolchainDefaults.ASSEMBLER, + ToolchainDefaults.OFFLOAD_BUNDLER, + ToolchainDefaults.DEVICE_ENUMERATOR, ) params = { "CxxCompiler": cxxCompiler, "CCompiler": cCompiler, "Assembler": assembler, "OffloadBundler": offloadBundler, - "HipConfig": hipconfig, "ROCmAgentEnumeratorPath": deviceEnumerator, } Common.assignGlobalParameters(params) diff --git a/shared/tensile/Tensile/Utilities/Toolchain.py b/shared/tensile/Tensile/Utilities/Toolchain.py index e9576faba5fb..92527c41e42f 100644 --- a/shared/tensile/Tensile/Utilities/Toolchain.py +++ b/shared/tensile/Tensile/Utilities/Toolchain.py @@ -114,7 +114,6 @@ class ToolchainDefaults(NamedTuple): C_COMPILER = osSelect(linux="amdclang", windows="clang.exe") OFFLOAD_BUNDLER = osSelect(linux="clang-offload-bundler", windows="clang-offload-bundler.exe") ASSEMBLER = osSelect(linux="amdclang++", windows="clang++.exe") - HIP_CONFIG = osSelect(linux="hipconfig", windows="hipconfig") DEVICE_ENUMERATOR = osSelect(linux="rocm_agent_enumerator", windows="hipinfo.exe") @@ -163,18 +162,6 @@ def supportedOffloadBundler(bundler: str) -> bool: return _supportedComponent(bundler, ["clang-offload-bundler"]) -def supportedHip(exe: str) -> bool: - """Determine if a HIP config executable is supported by Tensile. - - Args: - bundler: The name of an offload bundler to test for support. - - Return: - If supported True; otherwise, False. - """ - return _supportedComponent(exe, ["hipconfig", "hipcc"]) - - def supportedDeviceEnumerator(enumerator: str) -> bool: """Determine if a device enumerator is supported by Tensile. @@ -216,7 +203,6 @@ def _validateExecutable(file: str, searchPaths: List[Path]) -> str: supportedCxxCompiler(file), supportedCCompiler(file), supportedOffloadBundler(file), - supportedHip(file), supportedDeviceEnumerator(file), ) ): diff --git a/shared/tensile/Tensile/cmake/TensileConfig.cmake b/shared/tensile/Tensile/cmake/TensileConfig.cmake index 0a078ab798f3..619c04d6c78d 100644 --- a/shared/tensile/Tensile/cmake/TensileConfig.cmake +++ b/shared/tensile/Tensile/cmake/TensileConfig.cmake @@ -232,7 +232,7 @@ function(TensileCreateLibraryFiles endif() set(ENV_PATH_ARG "PATH=${ESC_PATH}") set(CommandLine - "${CMAKE_COMMAND}" -E env "PATH=${ESC_PATH}" -- + "${CMAKE_COMMAND}" -E env "PATH=${ESC_PATH}" "ROCM_VERSION=${HIP_VERSION}" -- ${CommandLine}) message(STATUS "Tensile_CREATE_COMMAND: ${CommandLine}")