From 147a175e37a444f4714c004c02534acecb340aee Mon Sep 17 00:00:00 2001
From: Deven Desai <deven.desai.amd@gmail.com>
Date: Fri, 4 Dec 2020 02:19:59 +0000
Subject: [PATCH] Porting ROCm toolchain specific changes from the master to
r2.4
---
third_party/gpus/compress_find_rocm_config.py | 36 +++
.../bin/crosstool_wrapper_driver_rocm.tpl | 42 +--
third_party/gpus/find_rocm_config.py | 286 ++++++++++++++++++
.../gpus/find_rocm_config.py.gz.base64 | 1 +
third_party/gpus/rocm/BUILD.tpl | 1 -
third_party/gpus/rocm/rocm_config.h.tpl | 4 +
third_party/gpus/rocm_configure.bzl | 152 ++++------
7 files changed, 391 insertions(+), 131 deletions(-)
create mode 100644 third_party/gpus/compress_find_rocm_config.py
create mode 100644 third_party/gpus/find_rocm_config.py
create mode 100644 third_party/gpus/find_rocm_config.py.gz.base64
diff --git a/third_party/gpus/compress_find_rocm_config.py b/third_party/gpus/compress_find_rocm_config.py
new file mode 100644
index 00000000000..90615d4b1ea
--- /dev/null
+++ b/third_party/gpus/compress_find_rocm_config.py
@@ -0,0 +1,36 @@
+# Copyright 2020 The TensorFlow Authors. All Rights Reserved.
+#
+# Licensed under the Apache License, Version 2.0 (the "License");
+# you may not use this file except in compliance with the License.
+# You may obtain a copy of the License at
+#
+# http://www.apache.org/licenses/LICENSE-2.0
+#
+# Unless required by applicable law or agreed to in writing, software
+# distributed under the License is distributed on an "AS IS" BASIS,
+# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+# See the License for the specific language governing permissions and
+# limitations under the License.
+# ==============================================================================
+"""Compresses the contents of 'find_rocm_config.py'.
+
+The compressed file is what is actually being used. It works around remote
+config not being able to upload files yet.
+"""
+import base64
+import zlib
+
+
+def main():
+ with open('find_rocm_config.py', 'rb') as f:
+ data = f.read()
+
+ compressed = zlib.compress(data)
+ b64encoded = base64.b64encode(compressed)
+
+ with open('find_rocm_config.py.gz.base64', 'wb') as f:
+ f.write(b64encoded)
+
+
+if __name__ == '__main__':
+ main()
diff --git a/third_party/gpus/crosstool/clang/bin/crosstool_wrapper_driver_rocm.tpl b/third_party/gpus/crosstool/clang/bin/crosstool_wrapper_driver_rocm.tpl
index d5bfe78c644..161bc7c8df4 100755
--- a/third_party/gpus/crosstool/clang/bin/crosstool_wrapper_driver_rocm.tpl
+++ b/third_party/gpus/crosstool/clang/bin/crosstool_wrapper_driver_rocm.tpl
@@ -26,12 +26,9 @@ import pipes
# Template values set by rocm_configure.bzl.
CPU_COMPILER = ('%{cpu_compiler}')
-GCC_HOST_COMPILER_PATH = ('%{gcc_host_compiler_path}')
HIPCC_PATH = '%{hipcc_path}'
-PREFIX_DIR = os.path.dirname(GCC_HOST_COMPILER_PATH)
HIPCC_ENV = '%{hipcc_env}'
-HIPCC_IS_HIPCLANG = '%{hipcc_is_hipclang}'=="True"
HIP_RUNTIME_PATH = '%{hip_runtime_path}'
HIP_RUNTIME_LIBRARY = '%{hip_runtime_library}'
ROCR_RUNTIME_PATH = '%{rocr_runtime_path}'
@@ -98,27 +95,6 @@ def GetHostCompilerOptions(argv):
return opts
-def GetHipccOptions(argv):
- """Collect the -hipcc_options values from argv.
-
- Args:
- argv: A list of strings, possibly the argv passed to main().
-
- Returns:
- The string that can be passed directly to hipcc.
- """
-
- parser = ArgumentParser()
- parser.add_argument('-hipcc_options', nargs='*', action='append')
-
- args, _ = parser.parse_known_args(argv)
-
- if args.hipcc_options:
- options = _update_options(sum(args.hipcc_options, []))
- return ' '.join(['--'+a for a in options])
- return ''
-
-
def system(cmd):
"""Invokes cmd with os.system().
@@ -148,7 +124,6 @@ def InvokeHipcc(argv, log=False):
"""
host_compiler_options = GetHostCompilerOptions(argv)
- hipcc_compiler_options = GetHipccOptions(argv)
opt_option = GetOptionValue(argv, 'O')
m_options = GetOptionValue(argv, 'm')
m_options = ''.join([' -m' + m for m in m_options if m in ['32', '64']])
@@ -193,14 +168,13 @@ def InvokeHipcc(argv, log=False):
# Otherwise, we get build error.
# Also we need to retain warning about uninitialised shared variable as
# warning only, even when -Werror option is specified.
- if HIPCC_IS_HIPCLANG:
- hipccopts += ' --include=hip/hip_runtime.h '
- hipccopts += ' ' + hipcc_compiler_options
+ hipccopts += ' --include=hip/hip_runtime.h '
# Use -fno-gpu-rdc by default for early GPU kernel finalization
# This flag would trigger GPU kernels be generated at compile time, instead
# of link time. This allows the default host compiler (gcc) be used as the
# linker for TensorFlow on ROCm platform.
hipccopts += ' -fno-gpu-rdc '
+ hipccopts += ' -fcuda-flush-denormals-to-zero '
hipccopts += undefines
hipccopts += defines
hipccopts += std_options
@@ -211,22 +185,19 @@ def InvokeHipcc(argv, log=False):
depfile = depfiles[0]
cmd = (HIPCC_PATH + ' ' + hipccopts +
host_compiler_options +
- ' ' + GCC_HOST_COMPILER_PATH +
' -I .' + includes + ' ' + srcs + ' -M -o ' + depfile)
+ cmd = HIPCC_ENV.replace(';', ' ') + ' ' + cmd
if log: Log(cmd)
+ if VERBOSE: print(cmd)
exit_status = os.system(cmd)
if exit_status != 0:
return exit_status
cmd = (HIPCC_PATH + ' ' + hipccopts +
host_compiler_options + ' -fPIC' +
- ' ' + GCC_HOST_COMPILER_PATH +
' -I .' + opt + includes + ' -c ' + srcs + out)
- # TODO(zhengxq): for some reason, 'gcc' needs this help to find 'as'.
- # Need to investigate and fix.
- cmd = 'PATH=' + PREFIX_DIR + ':$PATH '\
- + HIPCC_ENV.replace(';', ' ') + ' '\
+ cmd = HIPCC_ENV.replace(';', ' ') + ' '\
+ cmd
if log: Log(cmd)
if VERBOSE: print(cmd)
@@ -268,8 +239,7 @@ def main():
gpu_linker_flags.append('-L' + HIP_RUNTIME_PATH)
gpu_linker_flags.append('-Wl,-rpath=' + HIP_RUNTIME_PATH)
gpu_linker_flags.append('-l' + HIP_RUNTIME_LIBRARY)
- if HIPCC_IS_HIPCLANG:
- gpu_linker_flags.append("-lrt")
+ gpu_linker_flags.append("-lrt")
if VERBOSE: print(' '.join([CPU_COMPILER] + gpu_linker_flags))
return subprocess.call([CPU_COMPILER] + gpu_linker_flags)
diff --git a/third_party/gpus/find_rocm_config.py b/third_party/gpus/find_rocm_config.py
new file mode 100644
index 00000000000..c1eb119612b
--- /dev/null
+++ b/third_party/gpus/find_rocm_config.py
@@ -0,0 +1,286 @@
+# Copyright 2020 The TensorFlow Authors. All Rights Reserved.
+#
+# Licensed under the Apache License, Version 2.0 (the "License");
+# you may not use this file except in compliance with the License.
+# You may obtain a copy of the License at
+#
+# http://www.apache.org/licenses/LICENSE-2.0
+#
+# Unless required by applicable law or agreed to in writing, software
+# distributed under the License is distributed on an "AS IS" BASIS,
+# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+# See the License for the specific language governing permissions and
+# limitations under the License.
+# ==============================================================================
+"""Prints ROCm library and header directories and versions found on the system.
+
+The script searches for ROCm library and header files on the system, inspects
+them to determine their version and prints the configuration to stdout.
+The path to inspect is specified through an environment variable (ROCM_PATH).
+If no valid configuration is found, the script prints to stderr and
+returns an error code.
+
+The script takes the directory specified by the ROCM_PATH environment variable.
+The script looks for headers and library files in a hard-coded set of
+subdirectories from base path of the specified directory. If ROCM_PATH is not
+specified, then "/opt/rocm" is used as it default value
+
+"""
+
+import io
+import os
+import re
+import sys
+
+
+class ConfigError(Exception):
+ pass
+
+
+def _get_default_rocm_path():
+ return "/opt/rocm"
+
+
+def _get_rocm_install_path():
+ """Determines and returns the ROCm installation path."""
+ rocm_install_path = _get_default_rocm_path()
+ if "ROCM_PATH" in os.environ:
+ rocm_install_path = os.environ["ROCM_PATH"]
+ # rocm_install_path = os.path.realpath(rocm_install_path)
+ return rocm_install_path
+
+
+def _get_composite_version_number(major, minor, patch):
+ return 10000 * major + 100 * minor + patch
+
+
+def _get_header_version(path, name):
+ """Returns preprocessor defines in C header file."""
+ for line in io.open(path, "r", encoding="utf-8"):
+ match = re.match(r"#define %s +(\d+)" % name, line)
+ if match:
+ value = match.group(1)
+ return int(value)
+
+ raise ConfigError('#define "{}" is either\n'.format(name) +
+ " not present in file {} OR\n".format(path) +
+ " its value is not an integer literal")
+
+
+def _find_rocm_config(rocm_install_path):
+
+ def rocm_version_numbers(path):
+ version_file = os.path.join(path, ".info/version-dev")
+ if not os.path.exists(version_file):
+ raise ConfigError('ROCm version file "{}" not found'.format(version_file))
+ version_numbers = []
+ with open(version_file) as f:
+ version_string = f.read().strip()
+ version_numbers = version_string.split(".")
+ major = int(version_numbers[0])
+ minor = int(version_numbers[1])
+ patch = int(version_numbers[2].split("-")[0])
+ return major, minor, patch
+
+ major, minor, patch = rocm_version_numbers(rocm_install_path)
+
+ rocm_config = {
+ "rocm_version_number": _get_composite_version_number(major, minor, patch)
+ }
+
+ return rocm_config
+
+
+def _find_hipruntime_config(rocm_install_path):
+
+ def hipruntime_version_number(path):
+ version_file = os.path.join(path, "hip/include/hip/hip_version.h")
+ if not os.path.exists(version_file):
+ raise ConfigError(
+ 'HIP Runtime version file "{}" not found'.format(version_file))
+ # This header file has an explicit #define for HIP_VERSION, whose value
+ # is (HIP_VERSION_MAJOR * 100 + HIP_VERSION_MINOR)
+ # Retreive the major + minor and re-calculate here, since we do not
+ # want get into the business of parsing arith exprs
+ major = _get_header_version(version_file, "HIP_VERSION_MAJOR")
+ minor = _get_header_version(version_file, "HIP_VERSION_MINOR")
+ return 100 * major + minor
+
+ hipruntime_config = {
+ "hipruntime_version_number": hipruntime_version_number(rocm_install_path)
+ }
+
+ return hipruntime_config
+
+
+def _find_miopen_config(rocm_install_path):
+
+ def miopen_version_numbers(path):
+ version_file = os.path.join(path, "miopen/include/miopen/version.h")
+ if not os.path.exists(version_file):
+ raise ConfigError(
+ 'MIOpen version file "{}" not found'.format(version_file))
+ major = _get_header_version(version_file, "MIOPEN_VERSION_MAJOR")
+ minor = _get_header_version(version_file, "MIOPEN_VERSION_MINOR")
+ patch = _get_header_version(version_file, "MIOPEN_VERSION_PATCH")
+ return major, minor, patch
+
+ major, minor, patch = miopen_version_numbers(rocm_install_path)
+
+ miopen_config = {
+ "miopen_version_number":
+ _get_composite_version_number(major, minor, patch)
+ }
+
+ return miopen_config
+
+
+def _find_rocblas_config(rocm_install_path):
+
+ def rocblas_version_numbers(path):
+ possible_version_files = [
+ "rocblas/include/rocblas-version.h", # ROCm 3.7 and prior
+ "rocblas/include/internal/rocblas-version.h", # ROCm 3.8
+ ]
+ version_file = None
+ for f in possible_version_files:
+ version_file_path = os.path.join(path, f)
+ if os.path.exists(version_file_path):
+ version_file = version_file_path
+ break
+ if not version_file:
+ raise ConfigError(
+ "rocblas version file not found in {}".format(
+ possible_version_files))
+ major = _get_header_version(version_file, "ROCBLAS_VERSION_MAJOR")
+ minor = _get_header_version(version_file, "ROCBLAS_VERSION_MINOR")
+ patch = _get_header_version(version_file, "ROCBLAS_VERSION_PATCH")
+ return major, minor, patch
+
+ major, minor, patch = rocblas_version_numbers(rocm_install_path)
+
+ rocblas_config = {
+ "rocblas_version_number":
+ _get_composite_version_number(major, minor, patch)
+ }
+
+ return rocblas_config
+
+
+def _find_rocrand_config(rocm_install_path):
+
+ def rocrand_version_number(path):
+ version_file = os.path.join(path, "rocrand/include/rocrand_version.h")
+ if not os.path.exists(version_file):
+ raise ConfigError(
+ 'rocblas version file "{}" not found'.format(version_file))
+ version_number = _get_header_version(version_file, "ROCRAND_VERSION")
+ return version_number
+
+ rocrand_config = {
+ "rocrand_version_number": rocrand_version_number(rocm_install_path)
+ }
+
+ return rocrand_config
+
+
+def _find_rocfft_config(rocm_install_path):
+
+ def rocfft_version_numbers(path):
+ version_file = os.path.join(path, "rocfft/include/rocfft-version.h")
+ if not os.path.exists(version_file):
+ raise ConfigError(
+ 'rocfft version file "{}" not found'.format(version_file))
+ major = _get_header_version(version_file, "rocfft_version_major")
+ minor = _get_header_version(version_file, "rocfft_version_minor")
+ patch = _get_header_version(version_file, "rocfft_version_patch")
+ return major, minor, patch
+
+ major, minor, patch = rocfft_version_numbers(rocm_install_path)
+
+ rocfft_config = {
+ "rocfft_version_number":
+ _get_composite_version_number(major, minor, patch)
+ }
+
+ return rocfft_config
+
+
+def _find_roctracer_config(rocm_install_path):
+
+ def roctracer_version_numbers(path):
+ version_file = os.path.join(path, "roctracer/include/roctracer.h")
+ if not os.path.exists(version_file):
+ raise ConfigError(
+ 'roctracer version file "{}" not found'.format(version_file))
+ major = _get_header_version(version_file, "ROCTRACER_VERSION_MAJOR")
+ minor = _get_header_version(version_file, "ROCTRACER_VERSION_MINOR")
+ # roctracer header does not have a patch version number
+ patch = 0
+ return major, minor, patch
+
+ major, minor, patch = roctracer_version_numbers(rocm_install_path)
+
+ roctracer_config = {
+ "roctracer_version_number":
+ _get_composite_version_number(major, minor, patch)
+ }
+
+ return roctracer_config
+
+
+def _find_hipsparse_config(rocm_install_path):
+
+ def hipsparse_version_numbers(path):
+ version_file = os.path.join(path, "hipsparse/include/hipsparse-version.h")
+ if not os.path.exists(version_file):
+ raise ConfigError(
+ 'hipsparse version file "{}" not found'.format(version_file))
+ major = _get_header_version(version_file, "hipsparseVersionMajor")
+ minor = _get_header_version(version_file, "hipsparseVersionMinor")
+ patch = _get_header_version(version_file, "hipsparseVersionPatch")
+ return major, minor, patch
+
+ major, minor, patch = hipsparse_version_numbers(rocm_install_path)
+
+ hipsparse_config = {
+ "hipsparse_version_number":
+ _get_composite_version_number(major, minor, patch)
+ }
+
+ return hipsparse_config
+
+
+def find_rocm_config():
+ """Returns a dictionary of ROCm components config info."""
+ rocm_install_path = _get_rocm_install_path()
+ if not os.path.exists(rocm_install_path):
+ raise ConfigError(
+ 'Specified ROCM_PATH "{}" does not exist'.format(rocm_install_path))
+
+ result = {}
+
+ result["rocm_toolkit_path"] = rocm_install_path
+ result.update(_find_rocm_config(rocm_install_path))
+ result.update(_find_hipruntime_config(rocm_install_path))
+ result.update(_find_miopen_config(rocm_install_path))
+ result.update(_find_rocblas_config(rocm_install_path))
+ result.update(_find_rocrand_config(rocm_install_path))
+ result.update(_find_rocfft_config(rocm_install_path))
+ result.update(_find_roctracer_config(rocm_install_path))
+ result.update(_find_hipsparse_config(rocm_install_path))
+
+ return result
+
+
+def main():
+ try:
+ for key, value in sorted(find_rocm_config().items()):
+ print("%s: %s" % (key, value))
+ except ConfigError as e:
+ sys.stderr.write("\nERROR: {}\n\n".format(str(e)))
+ sys.exit(1)
+
+
+if __name__ == "__main__":
+ main()
diff --git a/third_party/gpus/find_rocm_config.py.gz.base64 b/third_party/gpus/find_rocm_config.py.gz.base64
new file mode 100644
index 00000000000..f38b64fda87
--- /dev/null
+++ b/third_party/gpus/find_rocm_config.py.gz.base64
@@ -0,0 +1 @@
+eJy9WtFu27gSfddXEAqKyhtHSXsfdpFFHrxpFvXe1gns7C4WTWDQMm1zK4u6JJXUKPrvd4akZEqWEydOa6CoZQ0PhzNnDkdiDsi5yFeSzxeavD15e0KuF4xcs0wJ+Xsq7kmv0AshVUx6aUqGaKbIkCkm79g0Dg6CA/KBJ2DOpqTIpkwSDeN7OU3gP3enS/5iUnGRkbfxCYnQIHS3ws6vgLASBVnSFcmEJoViAMEVmfGUEfYlYbkmPCOJWOYpp1nCyD3XCzONAwE3yD8OQkw0BWsK9jlczXw7QrVxGD8LrfPT4+P7+/uYGmdjIefHqTVUxx/65xeD0cUROGyG/JmlTCki2f8KLmGpkxWhOfiT0Al4mdJ7IiShc8ngnhbo773kmmfzLlFipu+pZIAy5UpLPil0LVild7Bm3wDCRTMS9kakPwrJb71Rf9QFjL/71+8v/7wmf/eGw97gun8xIpdDcn45eNe/7l8O4Op30hv8Q/7bH7zrEgahgmnYl1yi/+AkxzCa1JERYzUHZsI6pHKW8BlPYF3ZvKBzRubijskMlkNyJpdcYTIVuDcFlJQvuaba/LKxKJzm7EU/QRiGV5JnSMPL8yVMP5FUrtAZsmAU559CihItJGfGR3Jn2QeUEuAgBtascqU0W8ZBgIRXieTAM8WoBC4oE4pt8EhMVUfpQsYxaloF8OMSKTBlGkOVmRBzWTphgHLrP45PRDbj80KaAOI4paei0LHxKqdIdFGCI0NcbpBmCymK+QJJwrI7LkW2ZJkmd1RyQ8oI/P84vupdv+/EQX8GxQX3Uj5tTMldWLp2OTYOpYPGHSalSbVkupAm7QR+ggAlYsrq8dP0M7PrKnOw8jyGosFblV+tfsc+XirEZ5sMG3ubzzInNhGm2hdUTo/QnynkUEPdB6qY+DyYSbEkE6pcUJ0wrH2r/I0JxGrtIoQHVCmoDE2YoCyPRa6PpUiWIZoUKH8UfNGQ9xktUlxPWrAA2RoEUHNCQvpE+U2o8hvogvsGTAqCIEkp1Om5SdEFRjm6MBIIqeqcBgS8V2gGs5DxnOmxm26MroxxaZExs7ny3fQHGWMglaZp6g0CX9+VrLWRLlPu0rYkbpClDo6McYEwXxORnG31D8z5jIRVjENMoVCxYwN60g64tvnkjb4F+4Nt9sZFyWhqpt4w6qxDtXHPDxhuPkJxzcaujMdZsZwwGS3pv0J2CUQM/4NhycKP/5sT+JCfiDEjh3iNV2gNV8bcn8aSvJwjQi+6JKNLVqZn6PIBap6DwyDpAATDTcIgiue+RrnUYPWkqENwn4tY5KxEDmUIO0QGZQPKfhYWenb0S9ix8V+ibxBDyWLzNZLhgZ2IvFLkMLqZHnZC8sp41zX4HTMOMmvsLQqxZQA45sd4DpKVR2867qaLEmhNZOw6AYaOcqhSvwJelzOHX7+ZcrO72k32OobFAXJkYkQOHWz9ExLTV+AGiDoDUTCdxddvsG/eZGEJYeiwHYKDHNq1WEVAFQS/2ZxhdKFoaBp2ylyCs1PLeKu1LcQ7xaWisblVZ5WKnIkJoLtlnF5z+l/BqzTGPJuJY2d4NGV3YZUL9LQcwr5Ad6EiH7BTpqkl6Kbey13LzG7Cj4hmw6iCXwPs1Jx26wG/P92aG6Z1MxSsjULtnFWUcXewFYKO44zMsIKnUSfGX/Ko07BbT1IfGSvodHQUxi4ctgjPLN3qYz+d3DobU5rtNm+cTe5Ko83m7W0561HYqVAdz1vEAlnQ8jMWXhsvWgSsFF/LNBj41YUnbEEIT5+hZ4D3LWgIpZ2txvcFz2WRab5kO7DeM2548DTqA84xz5K0mLJj/A7/SsB48RJl4OnB6/f9KzK0Xj+3MA7g8QoExFNpaF1sS/UFHyeggSjFDnUbZhz/dTEcQWPfJfcLAc7ZvsJiAVLkmYw/9v6AZ4GfzDZzSGp3+oPLYekC7CKS8Tvb/Jdbk2W+3fSPEpomBezy4B2ToO+Km+cuaOuEaYcszj0FPQU6YSUIAzYpFG5FCvurnMLqoX6hr4OaxycQVSvDth3Pjxhkd2NtYb1KnwqBQQhrJfmmtj0bXKToBpf9utrKXaiu7bxu7T38utqYs1ZdS46quUNlOcP9dhQLUlWWu/xuhfWxfwn4z62pJxAKJrq6GOzLqSaKR6tSvZ+OAr3s+fvw+fvFlry37xg1NvncbkUJT71k7b1/1OZudkwTePzZrWkylg+wHPxTHJ4nx37YTSdSLSZ0MBXP3fXRmuhdo5fYCf0n/rl8dgeN2AqBPaHMaPoI1i8G4batHgcis/qOG8AM+9X2pTTbJfyx+eTj1fSsbJugah+o2LEXwhbnNkwrwwn0aJ99WfBNd5KBMpZ1HagkACMBmlAKQaNRb4/R0xUC8vPbh95oX4nYgHmeRjRh9hWJbXWzta/06rHRWrbgvKxO1GdvCoWEWtxNKIzlXk2mA/Flwgd9+f2wtRCe+/S1M9OGvcG7kml1jtUBHTW8DDSo0RJxaI22pOLRvqg+VZMHs5nejQZouF9TZEF8FsDl0fckAeD/gKaoERwz8umK10TBkU8XvAaKGbmf3rXlfavcrdnUoPQmyouL3XruJse1pAmEbCeaO9u9mW5xfLLbX74Lzy30D6A6qNz1sHd+MXyB/b0J5O3wB+tMVGdSgtn3hgsKD97UMbRcsRNWv1hO9iH9FhZs5X2NYQ3qt2K9OPtrHjTfKyl8l7DjayVnu18BVDj+yyX7y/cT/GqKH1AI1VzuZP7j80R/A+Z5qt+EudpT9rfToL0CmhRrvO1pxXrZCmh64Cpg4yCheRBEyZQneBiH56FiZh8vjSMZwxNctx48IHjspK7lUDDYxu62EnyE4K9H1Vnr+nTVcLuSRgNeEXxzDqtWTOHxKmTo2/ryk33ZrYVIP3NtrMPb8iV67VCvHBEX+ZRqFu1yVtPZMmqXN97bxj72Pm/buEffkTww8OFnpgcGPthkPzDuscblgcA+Jvkdf/swEK5klhRk3NBRy9Vp9SLlM1t1ywO8jCghNZtGm+UVQ/kuVdSpJNv8JUQUvlKn5JXCA89ojWT8d3+h5HEeD7PcGw+1UrH9C4oY/x6IReFNdjEcXg5Pgb43mXf8qLSMALBTDYNi0HhSGgRQguMxnnCOx+TsjITjMa5xPDYKZJcb/B9qO76I
\ No newline at end of file
diff --git a/third_party/gpus/rocm/BUILD.tpl b/third_party/gpus/rocm/BUILD.tpl
index d2533a08de1..ecbb4b5cebc 100644
--- a/third_party/gpus/rocm/BUILD.tpl
+++ b/third_party/gpus/rocm/BUILD.tpl
@@ -147,7 +147,6 @@ filegroup(
name = "rocm_root",
srcs = [
"rocm/bin/clang-offload-bundler",
- "rocm/bin/bin2c.py",
],
)
diff --git a/third_party/gpus/rocm/rocm_config.h.tpl b/third_party/gpus/rocm/rocm_config.h.tpl
index 957413b9acd..ec26b00a5b5 100644
--- a/third_party/gpus/rocm/rocm_config.h.tpl
+++ b/third_party/gpus/rocm/rocm_config.h.tpl
@@ -18,4 +18,8 @@ limitations under the License.
#define TF_ROCM_TOOLKIT_PATH "%{rocm_toolkit_path}"
+#define TF_ROCM_VERSION %{rocm_version_number}
+#define TF_MIOPEN_VERSION %{miopen_version_number}
+#define TF_HIPRUNTIME_VERSION %{hipruntime_version_number}
+
#endif // ROCM_ROCM_CONFIG_H_
diff --git a/third_party/gpus/rocm_configure.bzl b/third_party/gpus/rocm_configure.bzl
index 05082795188..10f03bfec24 100644
--- a/third_party/gpus/rocm_configure.bzl
+++ b/third_party/gpus/rocm_configure.bzl
@@ -4,11 +4,7 @@
* `TF_NEED_ROCM`: Whether to enable building with ROCm.
* `GCC_HOST_COMPILER_PATH`: The GCC host compiler path
- * `ROCM_TOOLKIT_PATH`: The path to the ROCm toolkit. Default is
- `/opt/rocm`.
- * `TF_ROCM_VERSION`: The version of the ROCm toolkit. If this is blank, then
- use the system default.
- * `TF_MIOPEN_VERSION`: The version of the MIOpen library.
+ * `ROCM_PATH`: The path to the ROCm toolkit. Default is `/opt/rocm`.
* `TF_ROCM_AMDGPU_TARGETS`: The AMDGPU targets.
"""
@@ -27,6 +23,7 @@ load(
"get_bash_bin",
"get_cpu_value",
"get_host_environ",
+ "get_python_bin",
"raw_exec",
"realpath",
"which",
@@ -35,13 +32,9 @@ load(
_GCC_HOST_COMPILER_PATH = "GCC_HOST_COMPILER_PATH"
_GCC_HOST_COMPILER_PREFIX = "GCC_HOST_COMPILER_PREFIX"
_ROCM_TOOLKIT_PATH = "ROCM_PATH"
-_TF_ROCM_VERSION = "TF_ROCM_VERSION"
-_TF_MIOPEN_VERSION = "TF_MIOPEN_VERSION"
_TF_ROCM_AMDGPU_TARGETS = "TF_ROCM_AMDGPU_TARGETS"
_TF_ROCM_CONFIG_REPO = "TF_ROCM_CONFIG_REPO"
-_DEFAULT_ROCM_VERSION = ""
-_DEFAULT_MIOPEN_VERSION = ""
_DEFAULT_ROCM_TOOLKIT_PATH = "/opt/rocm"
def verify_build_defines(params):
@@ -193,6 +186,7 @@ def _rocm_include_path(repository_ctx, rocm_config, bash_bin):
inc_dirs.append(rocm_toolkit_path + "/llvm/lib/clang/9.0.0/include")
inc_dirs.append(rocm_toolkit_path + "/llvm/lib/clang/10.0.0/include")
inc_dirs.append(rocm_toolkit_path + "/llvm/lib/clang/11.0.0/include")
+ inc_dirs.append(rocm_toolkit_path + "/llvm/lib/clang/12.0.0/include")
# Support hcc based off clang 10.0.0 (for ROCm 3.3)
inc_dirs.append(rocm_toolkit_path + "/hcc/compiler/lib/clang/10.0.0/include/")
@@ -212,20 +206,6 @@ def _enable_rocm(repository_ctx):
return True
return False
-def _rocm_toolkit_path(repository_ctx, bash_bin):
- """Finds the rocm toolkit directory.
-
- Args:
- repository_ctx: The repository context.
-
- Returns:
- A speculative real path of the rocm toolkit install directory.
- """
- rocm_toolkit_path = get_host_environ(repository_ctx, _ROCM_TOOLKIT_PATH, _DEFAULT_ROCM_TOOLKIT_PATH)
- if files_exist(repository_ctx, [rocm_toolkit_path], bash_bin) != [True]:
- auto_configure_fail("Cannot find rocm toolkit path.")
- return rocm_toolkit_path
-
def _amdgpu_targets(repository_ctx, rocm_toolkit_path, bash_bin):
"""Returns a list of strings representing AMDGPU targets."""
amdgpu_targets_str = get_host_environ(repository_ctx, _TF_ROCM_AMDGPU_TARGETS)
@@ -236,7 +216,7 @@ def _amdgpu_targets(repository_ctx, rocm_toolkit_path, bash_bin):
amdgpu_targets_str = ",".join(targets)
amdgpu_targets = amdgpu_targets_str.split(",")
for amdgpu_target in amdgpu_targets:
- if amdgpu_target[:3] != "gfx" or not amdgpu_target[3:].isdigit():
+ if amdgpu_target[:3] != "gfx":
auto_configure_fail("Invalid AMDGPU target: %s" % amdgpu_target)
return amdgpu_targets
@@ -265,51 +245,6 @@ def _hipcc_env(repository_ctx):
hipcc_env = (hipcc_env + " " + name + "=\"" + env_value + "\";")
return hipcc_env.strip()
-def _hipcc_is_hipclang(repository_ctx, rocm_config, bash_bin):
- """Returns if hipcc is based on hip-clang toolchain.
-
- Args:
- repository_ctx: The repository context.
- rocm_config: The path to the hip compiler.
- bash_bin: the path to the bash interpreter
-
- Returns:
- A string "True" if hipcc is based on hip-clang toolchain.
- The functions returns "False" if not (ie: based on HIP/HCC toolchain).
- """
-
- # check user-defined hip-clang environment variables
- for name in ["HIP_CLANG_PATH", "HIP_VDI_HOME"]:
- if get_host_environ(repository_ctx, name):
- return "True"
-
- # grep for "HIP_COMPILER=clang" in /opt/rocm/hip/lib/.hipInfo
- cmd = "grep HIP_COMPILER=clang %s/hip/lib/.hipInfo || true" % rocm_config.rocm_toolkit_path
- grep_result = execute(repository_ctx, [bash_bin, "-c", cmd], empty_stdout_fine = True)
- result = grep_result.stdout.strip()
- if result == "HIP_COMPILER=clang":
- return "True"
- return "False"
-
-def _if_hipcc_is_hipclang(repository_ctx, rocm_config, bash_bin, if_true, if_false = []):
- """
- Returns either the if_true or if_false arg based on whether hipcc
- is based on the hip-clang toolchain
-
- Args :
- repository_ctx: The repository context.
- rocm_config: The path to the hip compiler.
- if_true : value to return if hipcc is hip-clang based
- if_false : value to return if hipcc is not hip-clang based
- (optional, defaults to empty list)
-
- Returns :
- either the if_true arg or the of_False arg
- """
- if _hipcc_is_hipclang(repository_ctx, rocm_config, bash_bin) == "True":
- return if_true
- return if_false
-
def _crosstool_verbose(repository_ctx):
"""Returns the environment variable value CROSSTOOL_VERBOSE.
@@ -402,7 +337,40 @@ def _find_libs(repository_ctx, rocm_config, bash_bin):
return _select_rocm_lib_paths(repository_ctx, libs_paths, bash_bin)
-def _get_rocm_config(repository_ctx, bash_bin):
+def _exec_find_rocm_config(repository_ctx, script_path):
+ python_bin = get_python_bin(repository_ctx)
+
+ # If used with remote execution then repository_ctx.execute() can't
+ # access files from the source tree. A trick is to read the contents
+ # of the file in Starlark and embed them as part of the command. In
+ # this case the trick is not sufficient as the find_cuda_config.py
+ # script has more than 8192 characters. 8192 is the command length
+ # limit of cmd.exe on Windows. Thus we additionally need to compress
+ # the contents locally and decompress them as part of the execute().
+ compressed_contents = repository_ctx.read(script_path)
+ decompress_and_execute_cmd = (
+ "from zlib import decompress;" +
+ "from base64 import b64decode;" +
+ "from os import system;" +
+ "script = decompress(b64decode('%s'));" % compressed_contents +
+ "f = open('script.py', 'wb');" +
+ "f.write(script);" +
+ "f.close();" +
+ "system('\"%s\" script.py');" % (python_bin)
+ )
+
+ return execute(repository_ctx, [python_bin, "-c", decompress_and_execute_cmd])
+
+def find_rocm_config(repository_ctx, script_path):
+ """Returns ROCm config dictionary from running find_rocm_config.py"""
+ exec_result = _exec_find_rocm_config(repository_ctx, script_path)
+ if exec_result.return_code:
+ auto_configure_fail("Failed to run find_rocm_config.py: %s" % err_out(exec_result))
+
+ # Parse the dict from stdout.
+ return dict([tuple(x.split(": ")) for x in exec_result.stdout.splitlines()])
+
+def _get_rocm_config(repository_ctx, bash_bin, find_rocm_config_script):
"""Detects and returns information about the ROCm installation on the system.
Args:
@@ -413,11 +381,21 @@ def _get_rocm_config(repository_ctx, bash_bin):
A struct containing the following fields:
rocm_toolkit_path: The ROCm toolkit installation directory.
amdgpu_targets: A list of the system's AMDGPU targets.
+ rocm_version_number: The version of ROCm on the system.
+ miopen_version_number: The version of MIOpen on the system.
+ hipruntime_version_number: The version of HIP Runtime on the system.
"""
- rocm_toolkit_path = _rocm_toolkit_path(repository_ctx, bash_bin)
+ config = find_rocm_config(repository_ctx, find_rocm_config_script)
+ rocm_toolkit_path = config["rocm_toolkit_path"]
+ rocm_version_number = config["rocm_version_number"]
+ miopen_version_number = config["miopen_version_number"]
+ hipruntime_version_number = config["hipruntime_version_number"]
return struct(
- rocm_toolkit_path = rocm_toolkit_path,
amdgpu_targets = _amdgpu_targets(repository_ctx, rocm_toolkit_path, bash_bin),
+ rocm_toolkit_path = rocm_toolkit_path,
+ rocm_version_number = rocm_version_number,
+ miopen_version_number = miopen_version_number,
+ hipruntime_version_number = hipruntime_version_number,
)
def _tpl_path(repository_ctx, labelname):
@@ -550,8 +528,10 @@ def _create_local_rocm_repository(repository_ctx):
"rocm:rocm_config.h",
]}
+ find_rocm_config_script = repository_ctx.path(Label("@org_tensorflow//third_party/gpus:find_rocm_config.py.gz.base64"))
+
bash_bin = get_bash_bin(repository_ctx)
- rocm_config = _get_rocm_config(repository_ctx, bash_bin)
+ rocm_config = _get_rocm_config(repository_ctx, bash_bin, find_rocm_config_script)
# Copy header and library files to execroot.
# rocm_toolkit_path
@@ -609,13 +589,7 @@ def _create_local_rocm_repository(repository_ctx):
outs = rocm_lib_outs,
))
- clang_offload_bundler_path = rocm_toolkit_path + _if_hipcc_is_hipclang(
- repository_ctx,
- rocm_config,
- bash_bin,
- "/llvm/bin/",
- "/hcc/bin/",
- ) + "clang-offload-bundler"
+ clang_offload_bundler_path = rocm_toolkit_path + "/llvm/bin/clang-offload-bundler"
# copy files mentioned in third_party/gpus/rocm/BUILD
copy_rules.append(make_copy_files_rule(
@@ -688,17 +662,7 @@ def _create_local_rocm_repository(repository_ctx):
"-DTENSORFLOW_USE_ROCM=1",
"-D__HIP_PLATFORM_HCC__",
"-DEIGEN_USE_HIP",
- ] + _if_hipcc_is_hipclang(repository_ctx, rocm_config, bash_bin, [
- #
- # define "TENSORFLOW_COMPILER_IS_HIP_CLANG" when we are using clang
- # based hipcc to compile/build tensorflow
- #
- # Note that this #define should not be used to check whether or not
- # tensorflow is being built with ROCm support
- # (only TENSORFLOW_USE_ROCM should be used for that purpose)
- #
- "-DTENSORFLOW_COMPILER_IS_HIP_CLANG=1",
- ]))
+ ])
rocm_defines["%{host_compiler_path}"] = "clang/bin/crosstool_wrapper_driver_is_not_gcc"
@@ -729,7 +693,6 @@ def _create_local_rocm_repository(repository_ctx):
"%{cpu_compiler}": str(cc),
"%{hipcc_path}": rocm_config.rocm_toolkit_path + "/hip/bin/hipcc",
"%{hipcc_env}": _hipcc_env(repository_ctx),
- "%{hipcc_is_hipclang}": _hipcc_is_hipclang(repository_ctx, rocm_config, bash_bin),
"%{rocr_runtime_path}": rocm_config.rocm_toolkit_path + "/lib",
"%{rocr_runtime_library}": "hsa-runtime64",
"%{hip_runtime_path}": rocm_config.rocm_toolkit_path + "/hip/lib",
@@ -749,6 +712,9 @@ def _create_local_rocm_repository(repository_ctx):
["\"%s\"" % c for c in rocm_config.amdgpu_targets],
),
"%{rocm_toolkit_path}": rocm_config.rocm_toolkit_path,
+ "%{rocm_version_number}": rocm_config.rocm_version_number,
+ "%{miopen_version_number}": rocm_config.miopen_version_number,
+ "%{hipruntime_version_number}": rocm_config.hipruntime_version_number,
},
)
@@ -813,8 +779,6 @@ _ENVIRONS = [
_GCC_HOST_COMPILER_PREFIX,
"TF_NEED_ROCM",
_ROCM_TOOLKIT_PATH,
- _TF_ROCM_VERSION,
- _TF_MIOPEN_VERSION,
_TF_ROCM_AMDGPU_TARGETS,
]