From 0ac799b5c977b7fea7f27d0f9086d16eb105f630 Mon Sep 17 00:00:00 2001 From: Joseph Huber Date: Mon, 31 Jan 2022 14:31:54 -0500 Subject: [PATCH] [Libomptarget] Run GPU offloading tests using the new drvier This patch adds a new target to the tests to run using the new driver as the method for generating offloading code. Depends on D116541 Differential Revision: https://reviews.llvm.org/D118637 --- openmp/libomptarget/CMakeLists.txt | 2 ++ openmp/libomptarget/plugins/amdgpu/CMakeLists.txt | 2 +- openmp/libomptarget/plugins/cuda/CMakeLists.txt | 2 +- openmp/libomptarget/test/lit.cfg | 6 ++++- .../libomptarget/test/offloading/static_linking.c | 29 ++++++++++++++++++++++ .../libomptarget/test/unified_shared_memory/api.c | 2 ++ 6 files changed, 40 insertions(+), 3 deletions(-) create mode 100644 openmp/libomptarget/test/offloading/static_linking.c diff --git a/openmp/libomptarget/CMakeLists.txt b/openmp/libomptarget/CMakeLists.txt index 8d5a7bd..63f0801 100644 --- a/openmp/libomptarget/CMakeLists.txt +++ b/openmp/libomptarget/CMakeLists.txt @@ -39,11 +39,13 @@ endif() set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} aarch64-unknown-linux-gnu") set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} amdgcn-amd-amdhsa") set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} amdgcn-amd-amdhsa-newRTL") +set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} amdgcn-amd-amdhsa-newDriver") set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} powerpc64le-ibm-linux-gnu") set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} powerpc64-ibm-linux-gnu") set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} x86_64-pc-linux-gnu") set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} nvptx64-nvidia-cuda") set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} nvptx64-nvidia-cuda-newRTL") +set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPTARGET_ALL_TARGETS} nvptx64-nvidia-cuda-newDriver") # Once the plugins for the different targets are validated, they will be added to # the list of supported targets in the current system. diff --git a/openmp/libomptarget/plugins/amdgpu/CMakeLists.txt b/openmp/libomptarget/plugins/amdgpu/CMakeLists.txt index 67a892f..3375877 100644 --- a/openmp/libomptarget/plugins/amdgpu/CMakeLists.txt +++ b/openmp/libomptarget/plugins/amdgpu/CMakeLists.txt @@ -118,6 +118,6 @@ if (${amdgpu_arch_result}) libomptarget_say("Not generating amdgcn test targets as amdgpu-arch exited with ${amdgpu_arch_result}") else() # Report to the parent scope that we are building a plugin for amdgpu - set(LIBOMPTARGET_SYSTEM_TARGETS "${LIBOMPTARGET_SYSTEM_TARGETS} amdgcn-amd-amdhsa-newRTL" PARENT_SCOPE) + set(LIBOMPTARGET_SYSTEM_TARGETS "${LIBOMPTARGET_SYSTEM_TARGETS} amdgcn-amd-amdhsa-newRTL amdgcn-amd-amdhsa-newDriver " PARENT_SCOPE) endif() diff --git a/openmp/libomptarget/plugins/cuda/CMakeLists.txt b/openmp/libomptarget/plugins/cuda/CMakeLists.txt index 02311a2..4fa8b99 100644 --- a/openmp/libomptarget/plugins/cuda/CMakeLists.txt +++ b/openmp/libomptarget/plugins/cuda/CMakeLists.txt @@ -72,7 +72,7 @@ target_link_libraries(omptarget.rtl.cuda # Otherwise this plugin is being built speculatively and there may be no cuda available if (LIBOMPTARGET_CAN_LINK_LIBCUDA OR LIBOMPTARGET_FORCE_DLOPEN_LIBCUDA) libomptarget_say("Enable tests using CUDA plugin") - set(LIBOMPTARGET_SYSTEM_TARGETS "${LIBOMPTARGET_SYSTEM_TARGETS} nvptx64-nvidia-cuda-newRTL" PARENT_SCOPE) + set(LIBOMPTARGET_SYSTEM_TARGETS "${LIBOMPTARGET_SYSTEM_TARGETS} nvptx64-nvidia-cuda-newRTL nvptx64-nvidia-cuda-newDriver" PARENT_SCOPE) else() libomptarget_say("Disabling tests using CUDA plugin as cuda may not be available") endif() diff --git a/openmp/libomptarget/test/lit.cfg b/openmp/libomptarget/test/lit.cfg index 353466b6..63d3278 100644 --- a/openmp/libomptarget/test/lit.cfg +++ b/openmp/libomptarget/test/lit.cfg @@ -106,12 +106,16 @@ else: # Unices config.test_flags += " --libomptarget-nvptx-bc-path=" + config.library_dir if config.libomptarget_current_target.endswith('-newRTL'): config.test_flags += " -fopenmp-target-new-runtime" - else: + elif not config.libomptarget_current_target.endswith('-newDriver'): config.test_flags += " -fno-openmp-target-new-runtime" + if config.libomptarget_current_target.endswith('-newDriver'): + config.test_flags += " -fopenmp-new-driver" def remove_newRTL_suffix_if_present(name): if name.endswith('-newRTL'): return name[:-7] + elif name.endswith('-newDriver'): + return name[:-10] else: return name diff --git a/openmp/libomptarget/test/offloading/static_linking.c b/openmp/libomptarget/test/offloading/static_linking.c new file mode 100644 index 0000000..2af28a3 --- /dev/null +++ b/openmp/libomptarget/test/offloading/static_linking.c @@ -0,0 +1,29 @@ +// RUN: %libomptarget-compile-generic -DLIBRARY -c -o %t.o +// RUN: llvm-ar rcs %t.a %t.o +// RUN: %libomptarget-compile-generic %t.a && %libomptarget-run-generic 2>&1 | %fcheck-generic + +// REQUIRES: nvptx64-nvidia-cuda-newDriver +// REQUIRES: amdgcn-amd-amdhsa-newDriver + +#ifdef LIBRARY +int x = 42; +#pragma omp declare target(x) + +int foo() { + int value; +#pragma omp target map(from : value) + value = x; + return value; +} +#else +#include +int foo(); + +int main() { + int x = foo(); + + // CHECK: PASS + if (x == 42) + printf("PASS\n"); +} +#endif diff --git a/openmp/libomptarget/test/unified_shared_memory/api.c b/openmp/libomptarget/test/unified_shared_memory/api.c index f2882f2..c373a2d 100644 --- a/openmp/libomptarget/test/unified_shared_memory/api.c +++ b/openmp/libomptarget/test/unified_shared_memory/api.c @@ -1,10 +1,12 @@ // RUN: %libomptarget-compile-run-and-check-generic // XFAIL: nvptx64-nvidia-cuda // XFAIL: nvptx64-nvidia-cuda-newRTL +// XFAIL: nvptx64-nvidia-cuda-newDriver // Fails on amdgpu with error: GPU Memory Error // XFAIL: amdgcn-amd-amdhsa // XFAIL: amdgcn-amd-amdhsa-newRTL +// XFAIL: amdgcn-amd-amdhsa-newDriver #include #include -- 2.7.4