diff --git a/openmp/CMakeLists.txt b/openmp/CMakeLists.txt --- a/openmp/CMakeLists.txt +++ b/openmp/CMakeLists.txt @@ -43,6 +43,8 @@ "C compiler to use for testing OpenMP runtime libraries.") set(OPENMP_TEST_CXX_COMPILER ${CMAKE_CXX_COMPILER} CACHE STRING "C++ compiler to use for testing OpenMP runtime libraries.") + set(OPENMP_TEST_Fortran_COMPILER ${CMAKE_Fortran_COMPILER} CACHE STRING + "Fortran compiler to use for testing OpenMP runtime libraries.") set(OPENMP_LLVM_TOOLS_DIR "" CACHE PATH "Path to LLVM tools for testing.") else() set(OPENMP_ENABLE_WERROR ${LLVM_ENABLE_WERROR}) @@ -56,6 +58,10 @@ set(OPENMP_TEST_C_COMPILER ${LLVM_RUNTIME_OUTPUT_INTDIR}/clang.exe) set(OPENMP_TEST_CXX_COMPILER ${LLVM_RUNTIME_OUTPUT_INTDIR}/clang++.exe) endif() + + # Check if flang is in build bin folder + find_program(OPENMP_TEST_Fortran_COMPILER NAMES flang-new flang + PATHS ${LLVM_RUNTIME_OUTPUT_INTDIR} NO_DEFAULT_PATH) endif() # Check and set up common compiler flags. diff --git a/openmp/README.rst b/openmp/README.rst --- a/openmp/README.rst +++ b/openmp/README.rst @@ -260,6 +260,10 @@ Path of the folder that contains ``libomp.so``, and ``libLLVMSupport.so`` when profiling is enabled. This is required for testing. +**OPENMP_TEST_Fortran_COMPILER** = ``${CMAKE_Fortran_COMPILER}`` + Compiler to use for testing. Defaults to the compiler that was also used for + building when building libomptarget standalone and flang for in-tree builds. + Options for ``NVPTX device RTL`` -------------------------------- 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 @@ -290,6 +290,10 @@ config.substitutions.append(("%fcheck-" + libomptarget_target, \ "echo ignored-command")) +if config.test_fortran_compiler: + config.available_features.add("flang") + config.substitutions.append(("%flang", config.test_fortran_compiler)) + config.substitutions.append(("%clangxx", config.test_cxx_compiler)) config.substitutions.append(("%clang", config.test_c_compiler)) config.substitutions.append(("%openmp_flags", config.test_openmp_flags)) @@ -297,5 +301,6 @@ config.substitutions.append(("%cuda_flags", "--cuda-path=" + config.cuda_path)) else: config.substitutions.append(("%cuda_flags", "")) +config.substitutions.append(("%llvm_lib", config.llvm_lib_directory)) config.substitutions.append(("%flags", config.test_flags)) config.substitutions.append(("%not", config.libomptarget_not)) diff --git a/openmp/libomptarget/test/lit.site.cfg.in b/openmp/libomptarget/test/lit.site.cfg.in --- a/openmp/libomptarget/test/lit.site.cfg.in +++ b/openmp/libomptarget/test/lit.site.cfg.in @@ -2,6 +2,7 @@ config.test_c_compiler = "@OPENMP_TEST_C_COMPILER@" config.test_cxx_compiler = "@OPENMP_TEST_CXX_COMPILER@" +config.test_fortran_compiler = "@OPENMP_TEST_Fortran_COMPILER@" config.test_compiler_features = @OPENMP_TEST_COMPILER_FEATURES@ config.test_openmp_flags = "@OPENMP_TEST_OPENMP_FLAGS@" config.test_extra_flags = "@OPENMP_TEST_FLAGS@" diff --git a/openmp/libomptarget/test/offloading/fortran/basic_array.c b/openmp/libomptarget/test/offloading/fortran/basic_array.c new file mode 100644 --- /dev/null +++ b/openmp/libomptarget/test/offloading/fortran/basic_array.c @@ -0,0 +1,43 @@ +// Basic offloading test for function compiled with flang +// REQUIRES: flang, amdgcn-amd-amdhsa + +// RUN: %flang -c -fopenmp -fopenmp-targets=amdgcn-amd-amdhsa \ +// RUN: %S/basic_array.f90 -o basic_array.o +// RUN: %clang-amdgcn-amd-amdhsa -L%llvm_lib basic_array.o %s -o %t +// RUN: %t | %fcheck-generic + +#include +#define TEST_ARR_LEN 10 + +#pragma omp declare target +void increment_at(int i, int *array); +#pragma omp end declare target + +void increment_array(int *b, int n) { +#pragma omp target map(tofrom : b [0:n]) + for (int i = 0; i < n; i++) { + increment_at(i, b); + } +} + +int main() { + int arr[TEST_ARR_LEN] = {0, 1, 2, 3, 4, 5, 6, 7, 8, 9}; + + increment_array(arr, TEST_ARR_LEN); + for (int i = 0; i < TEST_ARR_LEN; i++) { + printf("%d = %d\n", i, arr[i]); + } + + return 0; +} + +// CHECK: 0 = 1 +// CHECK-NEXT: 1 = 2 +// CHECK-NEXT: 2 = 3 +// CHECK-NEXT: 3 = 4 +// CHECK-NEXT: 4 = 5 +// CHECK-NEXT: 5 = 6 +// CHECK-NEXT: 6 = 7 +// CHECK-NEXT: 7 = 8 +// CHECK-NEXT: 8 = 9 +// CHECK-NEXT: 9 = 10 diff --git a/openmp/libomptarget/test/offloading/fortran/basic_array.f90 b/openmp/libomptarget/test/offloading/fortran/basic_array.f90 new file mode 100644 --- /dev/null +++ b/openmp/libomptarget/test/offloading/fortran/basic_array.f90 @@ -0,0 +1,6 @@ +subroutine increment_at(c_index, arr) bind(C, name="increment_at") + use ISO_C_BINDING + integer (C_INT), dimension(*), intent(inout) :: arr + integer (C_INT), value :: c_index + arr(c_index+1) = arr(c_index+1) + 1 +end subroutine