From 32ac03519faddacab93fde7bf856f2ca8846b150 Mon Sep 17 00:00:00 2001 From: Pak Lui <5041261+paklui@users.noreply.github.com> Date: Mon, 11 Nov 2019 21:07:13 -0800 Subject: [PATCH 1/5] TOOLS/PERF: changes to enable the rocm perf modules --- config/m4/rocm.m4 | 12 +- src/tools/perf/Makefile.am | 2 +- src/tools/perf/configure.m4 | 1 + src/tools/perf/perftest.c | 14 +++ src/tools/perf/rocm/Makefile.am | 17 +++ src/tools/perf/rocm/configure.m4 | 11 ++ src/tools/perf/rocm/rocm_alloc.c | 189 +++++++++++++++++++++++++++++++ 7 files changed, 239 insertions(+), 7 deletions(-) create mode 100644 src/tools/perf/rocm/Makefile.am create mode 100644 src/tools/perf/rocm/configure.m4 create mode 100644 src/tools/perf/rocm/rocm_alloc.c diff --git a/config/m4/rocm.m4 b/config/m4/rocm.m4 index 567e50ed095..949ac96e942 100644 --- a/config/m4/rocm.m4 +++ b/config/m4/rocm.m4 @@ -42,14 +42,14 @@ AS_IF([test "x$with_rocm" != "xno"], [x|xguess|xyes], [AC_MSG_NOTICE([ROCm path was not specified. Guessing ...]) with_rocm=/opt/rocm - ROCM_CPPFLAGS="-I$with_rocm/include/hsa -I$with_rocm/include" - ROCM_LDFLAGS="-L$with_rocm/hsa/lib -L$with_rocm/lib" - ROCM_LIBS="-lhsa-runtime64"], + ROCM_CPPFLAGS="-I$with_rocm/include/hsa -D__HIP_PLATFORM_HCC__= -I$with_rocm/hip/include -I$with_rocm/hcc/include -I$with_rocm/hsa/include" + ROCM_LDFLAGS="-L$with_rocm/hsa/lib -L$with_rocm/lib -lhip_hcc -lhsa-runtime64" + ROCM_LIBS="-lhip_hcc -lhsa-runtime64"], [x/*], [AC_MSG_NOTICE([ROCm path given as $with_rocm ...]) - ROCM_CPPFLAGS="-I$with_rocm/include/hsa -I$with_rocm/include" - ROCM_LDFLAGS="-L$with_rocm/hsa/lib -L$with_rocm/lib" - ROCM_LIBS="-lhsa-runtime64"], + ROCM_CPPFLAGS="-I$with_rocm/include/hsa -D__HIP_PLATFORM_HCC__= -I$with_rocm/hip/include -I$with_rocm/hcc/include -I$with_rocm/hsa/include" + ROCM_LDFLAGS="-L$with_rocm/hsa/lib -L$with_rocm/lib -lhip_hcc -lhsa-runtime64" + ROCM_LIBS="-lhip_hcc -lhsa-runtime64"], [AC_MSG_NOTICE([ROCm flags given ...]) ROCM_PARSE_FLAGS([with_rocm], [ROCM_LIBS], [ROCM_LDFLAGS], [ROCM_CPPFLAGS])]) diff --git a/src/tools/perf/Makefile.am b/src/tools/perf/Makefile.am index 5cd2153802b..c0388239d8a 100644 --- a/src/tools/perf/Makefile.am +++ b/src/tools/perf/Makefile.am @@ -8,7 +8,7 @@ # See file LICENSE for terms. # -SUBDIRS = cuda lib +SUBDIRS = cuda rocm lib CC = $(UCX_PERFTEST_CC) noinst_HEADERS = api/libperf.h diff --git a/src/tools/perf/configure.m4 b/src/tools/perf/configure.m4 index 181006bd330..509c38331f3 100644 --- a/src/tools/perf/configure.m4 +++ b/src/tools/perf/configure.m4 @@ -7,6 +7,7 @@ ucx_perftest_modules="" m4_include([src/tools/perf/lib/configure.m4]) m4_include([src/tools/perf/cuda/configure.m4]) +m4_include([src/tools/perf/rocm/configure.m4]) AC_DEFINE_UNQUOTED([ucx_perftest_MODULES], ["${ucx_perftest_modules}"], [Perftest loadable modules]) diff --git a/src/tools/perf/perftest.c b/src/tools/perf/perftest.c index ad190cfb5ca..e4fbe920c47 100644 --- a/src/tools/perf/perftest.c +++ b/src/tools/perf/perftest.c @@ -370,9 +370,15 @@ static void usage(const struct perftest_context *ctx, const char *program) if (ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_CUDA] != NULL) { printf(" cuda - NVIDIA GPU memory\n"); } + if (ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM] != NULL) { + printf(" rocm - ROCm GPU memory\n"); + } if (ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_CUDA_MANAGED] != NULL) { printf(" cuda-managed - NVIDIA cuda managed/unified memory\n"); } + if (ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM_MANAGED] != NULL) { + printf(" rocm-managed - AMD ROCm managed/unified memory\n"); + } printf(" -n number of iterations to run (%ld)\n", ctx->params.max_iter); printf(" -w number of warm-up iterations (%zu)\n", ctx->params.warmup_iter); @@ -669,10 +675,18 @@ static ucs_status_t parse_test_params(ucx_perf_params_t *params, char opt, const (ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_CUDA] != NULL)) { params->mem_type = UCS_MEMORY_TYPE_CUDA; return UCS_OK; + } else if (!strcmp(optarg, "rocm") && + (ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM] != NULL)) { + params->mem_type = UCS_MEMORY_TYPE_ROCM; + return UCS_OK; } else if (!strcmp(optarg, "cuda-managed") && (ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_CUDA_MANAGED] != NULL)) { params->mem_type = UCS_MEMORY_TYPE_CUDA_MANAGED; return UCS_OK; + } else if (!strcmp(optarg, "rocm-managed") && + (ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM_MANAGED] != NULL)) { + params->mem_type = UCS_MEMORY_TYPE_ROCM_MANAGED; + return UCS_OK; } ucs_error("Unsupported memory type: \"%s\"", optarg); diff --git a/src/tools/perf/rocm/Makefile.am b/src/tools/perf/rocm/Makefile.am new file mode 100644 index 00000000000..b03480d4b12 --- /dev/null +++ b/src/tools/perf/rocm/Makefile.am @@ -0,0 +1,17 @@ +# +# Copyright (C) Mellanox Technologies Ltd. 2001-2018. ALL RIGHTS RESERVED. +# +# See file LICENSE for terms. +# + +if HAVE_ROCM + +module_LTLIBRARIES = libucx_perftest_rocm.la +libucx_perftest_rocm_la_CPPFLAGS = $(BASE_CPPFLAGS) $(ROCM_CPPFLAGS) -D__HIP_PLATFORM_HCC__= -I/opt/rocm/hip/include -I/opt/rocm/hcc/include -I/opt/rocm/hsa/include +libucx_perftest_rocm_la_CFLAGS = $(BASE_CFLAGS) $(ROCM_CFLAGS) -D__HIP_PLATFORM_HCC__= -I/opt/rocm/hip/include -I/opt/rocm/hcc/include -I/opt/rocm/hsa/include +libucx_perftest_rocm_la_LDFLAGS = $(ROCM_LDFLAGS) -version-info $(SOVERSION) +libucx_perftest_rocm_la_SOURCES = rocm_alloc.c + +include $(top_srcdir)/config/module.am + +endif diff --git a/src/tools/perf/rocm/configure.m4 b/src/tools/perf/rocm/configure.m4 new file mode 100644 index 00000000000..e23376672f1 --- /dev/null +++ b/src/tools/perf/rocm/configure.m4 @@ -0,0 +1,11 @@ +# +# Copyright (C) Mellanox Technologies Ltd. 2001-2018. ALL RIGHTS RESERVED. +# +# See file LICENSE for terms. +# + +UCX_CHECK_ROCM + +AS_IF([test "x$rocm_happy" = "xyes"], [ucx_perftest_modules="${ucx_perftest_modules}:rocm"]) + +AC_CONFIG_FILES([src/tools/perf/rocm/Makefile]) diff --git a/src/tools/perf/rocm/rocm_alloc.c b/src/tools/perf/rocm/rocm_alloc.c new file mode 100644 index 00000000000..43e9b0d6b03 --- /dev/null +++ b/src/tools/perf/rocm/rocm_alloc.c @@ -0,0 +1,189 @@ +/** + * Copyright (C) Mellanox Technologies Ltd. 2001-2018. ALL RIGHTS RESERVED. + * + * See file LICENSE for terms. + */ + +#include + +#include "hip/hip_runtime.h" +#include + + +static ucs_status_t ucx_perf_rocm_init(ucx_perf_context_t *perf) +{ + hipError_t hiperr; + unsigned group_index; + int num_gpus; + int gpu_index; + + group_index = rte_call(perf, group_index); + + hiperr = hipGetDeviceCount(&num_gpus); + if (hiperr != hipSuccess) { + return UCS_ERR_NO_DEVICE; + } + + gpu_index = group_index % num_gpus; + + hiperr = hipSetDevice(gpu_index); + if (hiperr != hipSuccess) { + return UCS_ERR_NO_DEVICE; + } + + return UCS_OK; +} + +static inline ucs_status_t ucx_perf_rocm_alloc(size_t length, + ucs_memory_type_t mem_type, + void **address_p) +{ + hipError_t hiperr; + + ucs_assert((mem_type == UCS_MEMORY_TYPE_ROCM) || + (mem_type == UCS_MEMORY_TYPE_ROCM_MANAGED)); + + hiperr = ((mem_type == UCS_MEMORY_TYPE_ROCM) ? + hipMalloc(address_p, length) : + hipMallocManaged(address_p, length, hipMemAttachGlobal)); + if (hiperr != hipSuccess) { + ucs_error("failed to allocate memory"); + return UCS_ERR_NO_MEMORY; + } + + return UCS_OK; +} + +static ucs_status_t ucp_perf_rocm_alloc(const ucx_perf_context_t *perf, size_t length, + void **address_p, ucp_mem_h *memh_p, + int non_blk_flag) +{ + return ucx_perf_rocm_alloc(length, UCS_MEMORY_TYPE_ROCM, address_p); +} + +static ucs_status_t ucp_perf_rocm_alloc_managed(const ucx_perf_context_t *perf, + size_t length, void **address_p, + ucp_mem_h *memh_p, int non_blk_flag) +{ + return ucx_perf_rocm_alloc(length, UCS_MEMORY_TYPE_ROCM_MANAGED, address_p); +} + +static void ucp_perf_rocm_free(const ucx_perf_context_t *perf, + void *address, ucp_mem_h memh) +{ + hipFree(address); +} + +static inline ucs_status_t +uct_perf_rocm_alloc_reg_mem(const ucx_perf_context_t *perf, + size_t length, + ucs_memory_type_t mem_type, + unsigned flags, + uct_allocated_memory_t *alloc_mem) +{ + ucs_status_t status; + + status = ucx_perf_rocm_alloc(length, mem_type, &alloc_mem->address); + if (status != UCS_OK) { + return status; + } + + status = uct_md_mem_reg(perf->uct.md, alloc_mem->address, + length, flags, &alloc_mem->memh); + if (status != UCS_OK) { + hipFree(alloc_mem->address); + ucs_error("failed to register memory"); + return status; + } + + alloc_mem->mem_type = mem_type; + alloc_mem->md = perf->uct.md; + + return UCS_OK; +} + +static ucs_status_t uct_perf_rocm_alloc(const ucx_perf_context_t *perf, + size_t length, unsigned flags, + uct_allocated_memory_t *alloc_mem) +{ + return uct_perf_rocm_alloc_reg_mem(perf, length, UCS_MEMORY_TYPE_ROCM, + flags, alloc_mem); +} + +static ucs_status_t uct_perf_rocm_managed_alloc(const ucx_perf_context_t *perf, + size_t length, unsigned flags, + uct_allocated_memory_t *alloc_mem) +{ + return uct_perf_rocm_alloc_reg_mem(perf, length, UCS_MEMORY_TYPE_ROCM_MANAGED, + flags, alloc_mem); +} + +static void uct_perf_rocm_free(const ucx_perf_context_t *perf, + uct_allocated_memory_t *alloc_mem) +{ + ucs_status_t status; + + ucs_assert(alloc_mem->md == perf->uct.md); + + status = uct_md_mem_dereg(perf->uct.md, alloc_mem->memh); + if (status != UCS_OK) { + ucs_error("failed to deregister memory"); + } + + hipFree(alloc_mem->address); +} + +static void ucx_perf_rocm_memcpy(void *dst, ucs_memory_type_t dst_mem_type, + const void *src, ucs_memory_type_t src_mem_type, + size_t count) +{ + hipError_t hiperr; + + hiperr = hipMemcpy(dst, src, count, hipMemcpyDefault); + if (hiperr != hipSuccess) { + ucs_error("failed to copy memory: %s", hipGetErrorString(hiperr)); + } +} + +static void* ucx_perf_rocm_memset(void *dst, int value, size_t count) +{ + hipError_t hiperr; + + hiperr = hipMemset(dst, value, count); + if (hiperr != hipSuccess) { + ucs_error("failed to set memory: %s", hipGetErrorString(hiperr)); + } + + return dst; +} + +UCS_STATIC_INIT { + static ucx_perf_allocator_t hip_allocator = { + .mem_type = UCS_MEMORY_TYPE_ROCM, + .init = ucx_perf_rocm_init, + .ucp_alloc = ucp_perf_rocm_alloc, + .ucp_free = ucp_perf_rocm_free, + .uct_alloc = uct_perf_rocm_alloc, + .uct_free = uct_perf_rocm_free, + .memcpy = ucx_perf_rocm_memcpy, + .memset = ucx_perf_rocm_memset + }; + static ucx_perf_allocator_t hip_managed_allocator = { + .mem_type = UCS_MEMORY_TYPE_ROCM_MANAGED, + .init = ucx_perf_rocm_init, + .ucp_alloc = ucp_perf_rocm_alloc_managed, + .ucp_free = ucp_perf_rocm_free, + .uct_alloc = uct_perf_rocm_managed_alloc, + .uct_free = uct_perf_rocm_free, + .memcpy = ucx_perf_rocm_memcpy, + .memset = ucx_perf_rocm_memset + }; + + ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM] = &hip_allocator; + ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM_MANAGED] = &hip_managed_allocator; +} +UCS_STATIC_CLEANUP { + ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM] = NULL; + ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM_MANAGED] = NULL; + +} From 877b6b4e31a7312054ea77b696716ae15e726c0f Mon Sep 17 00:00:00 2001 From: Pak Lui <5041261+paklui@users.noreply.github.com> Date: Mon, 11 Nov 2019 23:02:00 -0800 Subject: [PATCH 2/5] TOOLS/PERF: remove empty lines and add copyright as suggested. --- src/tools/perf/rocm/rocm_alloc.c | 18 +----------------- 1 file changed, 1 insertion(+), 17 deletions(-) diff --git a/src/tools/perf/rocm/rocm_alloc.c b/src/tools/perf/rocm/rocm_alloc.c index 43e9b0d6b03..1507d63ea27 100644 --- a/src/tools/perf/rocm/rocm_alloc.c +++ b/src/tools/perf/rocm/rocm_alloc.c @@ -1,36 +1,30 @@ /** * Copyright (C) Mellanox Technologies Ltd. 2001-2018. ALL RIGHTS RESERVED. + * Copyright (C) Advanced Micro Devices, Inc. 2019. ALL RIGHTS RESERVED. * * See file LICENSE for terms. */ #include - #include "hip/hip_runtime.h" #include - static ucs_status_t ucx_perf_rocm_init(ucx_perf_context_t *perf) { hipError_t hiperr; unsigned group_index; int num_gpus; int gpu_index; - group_index = rte_call(perf, group_index); - hiperr = hipGetDeviceCount(&num_gpus); if (hiperr != hipSuccess) { return UCS_ERR_NO_DEVICE; } - gpu_index = group_index % num_gpus; - hiperr = hipSetDevice(gpu_index); if (hiperr != hipSuccess) { return UCS_ERR_NO_DEVICE; } - return UCS_OK; } @@ -39,10 +33,8 @@ static inline ucs_status_t ucx_perf_rocm_alloc(size_t length, void **address_p) { hipError_t hiperr; - ucs_assert((mem_type == UCS_MEMORY_TYPE_ROCM) || (mem_type == UCS_MEMORY_TYPE_ROCM_MANAGED)); - hiperr = ((mem_type == UCS_MEMORY_TYPE_ROCM) ? hipMalloc(address_p, length) : hipMallocManaged(address_p, length, hipMemAttachGlobal)); @@ -50,7 +42,6 @@ static inline ucs_status_t ucx_perf_rocm_alloc(size_t length, ucs_error("failed to allocate memory"); return UCS_ERR_NO_MEMORY; } - return UCS_OK; } @@ -87,7 +78,6 @@ uct_perf_rocm_alloc_reg_mem(const ucx_perf_context_t *perf, if (status != UCS_OK) { return status; } - status = uct_md_mem_reg(perf->uct.md, alloc_mem->address, length, flags, &alloc_mem->memh); if (status != UCS_OK) { @@ -95,10 +85,8 @@ uct_perf_rocm_alloc_reg_mem(const ucx_perf_context_t *perf, ucs_error("failed to register memory"); return status; } - alloc_mem->mem_type = mem_type; alloc_mem->md = perf->uct.md; - return UCS_OK; } @@ -129,7 +117,6 @@ static void uct_perf_rocm_free(const ucx_perf_context_t *perf, if (status != UCS_OK) { ucs_error("failed to deregister memory"); } - hipFree(alloc_mem->address); } @@ -153,7 +140,6 @@ static void* ucx_perf_rocm_memset(void *dst, int value, size_t count) if (hiperr != hipSuccess) { ucs_error("failed to set memory: %s", hipGetErrorString(hiperr)); } - return dst; } @@ -178,12 +164,10 @@ UCS_STATIC_INIT { .memcpy = ucx_perf_rocm_memcpy, .memset = ucx_perf_rocm_memset }; - ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM] = &hip_allocator; ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM_MANAGED] = &hip_managed_allocator; } UCS_STATIC_CLEANUP { ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM] = NULL; ucx_perf_mem_type_allocators[UCS_MEMORY_TYPE_ROCM_MANAGED] = NULL; - } From 0a9ed31776d5841eabec6243c64888d025de6740 Mon Sep 17 00:00:00 2001 From: Pak Lui <5041261+paklui@users.noreply.github.com> Date: Wed, 27 Nov 2019 16:21:55 -0800 Subject: [PATCH 3/5] TOOLS/PERF: HIP version check to avoid issue for older HIP version --- src/tools/perf/rocm/configure.m4 | 27 ++++++++++++++++++++++++++- 1 file changed, 26 insertions(+), 1 deletion(-) diff --git a/src/tools/perf/rocm/configure.m4 b/src/tools/perf/rocm/configure.m4 index e23376672f1..effb4d484cc 100644 --- a/src/tools/perf/rocm/configure.m4 +++ b/src/tools/perf/rocm/configure.m4 @@ -1,11 +1,36 @@ # # Copyright (C) Mellanox Technologies Ltd. 2001-2018. ALL RIGHTS RESERVED. +# Copyright (C) Advanced Micro Devices, Inc. 2019. ALL RIGHTS RESERVED. # # See file LICENSE for terms. # UCX_CHECK_ROCM -AS_IF([test "x$rocm_happy" = "xyes"], [ucx_perftest_modules="${ucx_perftest_modules}:rocm"]) +# HIP version check for tools/perf/rocm. In HIP version 1.5, hip_runtime.h can +# trigger error due to missing braces around initializer in hip_vector_type.h + +AC_CHECK_PROG([HIPCONFIG_CHECK], [hipconfig], [yes], [], [$with_rocm/hip/bin]) + +AC_MSG_CHECKING([HIP version for ROCm perftest]) +hip_happy=no +if test x"${HIPCONFIG_CHECK}" == x"yes" ; then + HIPCONFIG=${with_rocm}/hip/bin/hipconfig + HIP_VER_MAJOR=$($HIPCONFIG -v | cut -d '.' -f1) + HIP_VER_MINOR=$($HIPCONFIG -v | cut -d '.' -f2) + AS_VERSION_COMPARE([$HIP_VER_MAJOR.$HIP_VER_MINOR], [2.0], + [AC_MSG_RESULT( + [HIP v${HIP_VER_MAJOR}.${HIP_VER_MINOR} is old, skipping ROCm perftest])], + [hip_happy=yes], + [hip_happy=yes] + ) +else + AC_MSG_RESULT([no hipconfig detected, skipping ROCm perftest]) +fi + +AS_IF([test "x$rocm_happy" = "xyes" && test "x$hip_happy" = "xyes"], + [AC_MSG_RESULT([yes]) + ucx_perftest_modules="${ucx_perftest_modules}:rocm"] +) AC_CONFIG_FILES([src/tools/perf/rocm/Makefile]) From 0c39d0f293f2ef963db3e2274dad32384fe3a9ff Mon Sep 17 00:00:00 2001 From: Pak Lui <5041261+paklui@users.noreply.github.com> Date: Wed, 27 Nov 2019 22:31:12 -0800 Subject: [PATCH 4/5] TOOLS/PERF: Use HAVE_HIP to control whether to build ROCm perftest --- src/tools/perf/rocm/Makefile.am | 3 ++- src/tools/perf/rocm/configure.m4 | 2 ++ 2 files changed, 4 insertions(+), 1 deletion(-) diff --git a/src/tools/perf/rocm/Makefile.am b/src/tools/perf/rocm/Makefile.am index b03480d4b12..321ba44efab 100644 --- a/src/tools/perf/rocm/Makefile.am +++ b/src/tools/perf/rocm/Makefile.am @@ -1,10 +1,11 @@ # # Copyright (C) Mellanox Technologies Ltd. 2001-2018. ALL RIGHTS RESERVED. +# Copyright (C) Advanced Micro Devices, Inc. 2019. ALL RIGHTS RESERVED. # # See file LICENSE for terms. # -if HAVE_ROCM +if HAVE_HIP module_LTLIBRARIES = libucx_perftest_rocm.la libucx_perftest_rocm_la_CPPFLAGS = $(BASE_CPPFLAGS) $(ROCM_CPPFLAGS) -D__HIP_PLATFORM_HCC__= -I/opt/rocm/hip/include -I/opt/rocm/hcc/include -I/opt/rocm/hsa/include diff --git a/src/tools/perf/rocm/configure.m4 b/src/tools/perf/rocm/configure.m4 index effb4d484cc..918ece63169 100644 --- a/src/tools/perf/rocm/configure.m4 +++ b/src/tools/perf/rocm/configure.m4 @@ -33,4 +33,6 @@ AS_IF([test "x$rocm_happy" = "xyes" && test "x$hip_happy" = "xyes"], ucx_perftest_modules="${ucx_perftest_modules}:rocm"] ) +AM_CONDITIONAL([HAVE_HIP], [test "x$hip_happy" != xno]) + AC_CONFIG_FILES([src/tools/perf/rocm/Makefile]) From cf087f6d369e26bf1d9f5775c2dd955f180d47d1 Mon Sep 17 00:00:00 2001 From: Pak Lui <5041261+paklui@users.noreply.github.com> Date: Thu, 28 Nov 2019 07:46:09 -0800 Subject: [PATCH 5/5] TOOLS/PERF: small fix to define HAVE_HIP in configure.ac --- configure.ac | 2 ++ 1 file changed, 2 insertions(+) diff --git a/configure.ac b/configure.ac index 96df7e9d1e6..581ae0e13ba 100644 --- a/configure.ac +++ b/configure.ac @@ -4,6 +4,7 @@ # Copyright (C) The University of Tennessee and The University # of Tennessee Research Foundation. 2016. ALL RIGHTS RESERVED. # Copyright (C) ARM Ltd. 2016-2019. ALL RIGHTS RESERVED. +# Copyright (C) Advanced Micro Devices, Inc. 2019. ALL RIGHTS RESERVED. # See file LICENSE for terms. # AC_PREREQ([2.63]) @@ -182,6 +183,7 @@ AS_IF([test "x$with_docs_only" = xyes], AM_CONDITIONAL([HAVE_CUDA], [false]) AM_CONDITIONAL([HAVE_GDR_COPY], [false]) AM_CONDITIONAL([HAVE_ROCM], [false]) + AM_CONDITIONAL([HAVE_HIP], [false]) AM_CONDITIONAL([HAVE_XPMEM], [false]) AM_CONDITIONAL([HAVE_CMA], [false]) AM_CONDITIONAL([HAVE_KNEM], [false])