diff --git a/openmp/libomptarget/CMakeLists.txt b/openmp/libomptarget/CMakeLists.txt --- a/openmp/libomptarget/CMakeLists.txt +++ b/openmp/libomptarget/CMakeLists.txt @@ -39,11 +39,13 @@ 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 --- a/openmp/libomptarget/plugins/amdgpu/CMakeLists.txt +++ b/openmp/libomptarget/plugins/amdgpu/CMakeLists.txt @@ -118,6 +118,6 @@ 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 --- a/openmp/libomptarget/plugins/cuda/CMakeLists.txt +++ b/openmp/libomptarget/plugins/cuda/CMakeLists.txt @@ -72,7 +72,7 @@ # 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 --- a/openmp/libomptarget/test/lit.cfg +++ b/openmp/libomptarget/test/lit.cfg @@ -106,12 +106,16 @@ 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 --- /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 --- 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