[Openmp-commits] [openmp] r343324 - [libomptarget-nvptx] Add testing infrastructure
Jonas Hahnfeld via Openmp-commits
openmp-commits at lists.llvm.org
Fri Sep 28 08:05:43 PDT 2018
Author: hahnfeld
Date: Fri Sep 28 08:05:43 2018
New Revision: 343324
URL: http://llvm.org/viewvc/llvm-project?rev=343324&view=rev
Log:
[libomptarget-nvptx] Add testing infrastructure
This patch also introduces testing for libomptarget-nvptx
which has been missing until now. I propose to add tests for
all bugs that are fixed in the future.
The target check-libomptarget-nvptx is not run by default because
- we can't determine if there is a GPU plugged into the system.
- it will require the latest Clang compiler. Keeping compatibility
with older releases would prevent testing newer code generation
developed in trunk.
Differential Revision: https://reviews.llvm.org/D51687
Added:
openmp/trunk/libomptarget/deviceRTLs/nvptx/test/
openmp/trunk/libomptarget/deviceRTLs/nvptx/test/CMakeLists.txt
openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.cfg
openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.site.cfg.in
openmp/trunk/libomptarget/deviceRTLs/nvptx/test/parallel/
openmp/trunk/libomptarget/deviceRTLs/nvptx/test/parallel/thread_limit.c
Modified:
openmp/trunk/cmake/OpenMPTesting.cmake
openmp/trunk/libomptarget/CMakeLists.txt
openmp/trunk/libomptarget/deviceRTLs/nvptx/CMakeLists.txt
openmp/trunk/libomptarget/test/CMakeLists.txt
openmp/trunk/libomptarget/test/lit.site.cfg.in
Modified: openmp/trunk/cmake/OpenMPTesting.cmake
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/cmake/OpenMPTesting.cmake?rev=343324&r1=343323&r2=343324&view=diff
==============================================================================
--- openmp/trunk/cmake/OpenMPTesting.cmake (original)
+++ openmp/trunk/cmake/OpenMPTesting.cmake Fri Sep 28 08:05:43 2018
@@ -147,7 +147,7 @@ function(add_openmp_testsuite target com
return()
endif()
- cmake_parse_arguments(ARG "" "" "DEPENDS" ${ARGN})
+ cmake_parse_arguments(ARG "" "" "DEPENDS;ARGS" ${ARGN})
# EXCLUDE_FROM_ALL excludes the test ${target} out of check-openmp.
if (NOT EXCLUDE_FROM_ALL)
# Register the testsuites and depends for the check-openmp rule.
@@ -156,8 +156,9 @@ function(add_openmp_testsuite target com
endif()
if (${OPENMP_STANDALONE_BUILD})
+ set(LIT_ARGS ${OPENMP_LIT_ARGS} ${ARG_ARGS})
add_custom_target(${target}
- COMMAND ${PYTHON_EXECUTABLE} ${OPENMP_LLVM_LIT_EXECUTABLE} ${OPENMP_LIT_ARGS} ${ARG_UNPARSED_ARGUMENTS}
+ COMMAND ${PYTHON_EXECUTABLE} ${OPENMP_LLVM_LIT_EXECUTABLE} ${LIT_ARGS} ${ARG_UNPARSED_ARGUMENTS}
COMMENT ${comment}
DEPENDS ${ARG_DEPENDS}
${cmake_3_2_USES_TERMINAL}
@@ -167,6 +168,7 @@ function(add_openmp_testsuite target com
${comment}
${ARG_UNPARSED_ARGUMENTS}
DEPENDS clang clang-headers FileCheck ${ARG_DEPENDS}
+ ARGS ${ARG_ARGS}
)
endif()
endfunction()
Modified: openmp/trunk/libomptarget/CMakeLists.txt
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/libomptarget/CMakeLists.txt?rev=343324&r1=343323&r2=343324&view=diff
==============================================================================
--- openmp/trunk/libomptarget/CMakeLists.txt (original)
+++ openmp/trunk/libomptarget/CMakeLists.txt Fri Sep 28 08:05:43 2018
@@ -20,6 +20,7 @@ set(CMAKE_MODULE_PATH ${CMAKE_CURRENT_SO
if(OPENMP_STANDALONE_BUILD)
# Build all libraries into a common place so that tests can find them.
+ set(CMAKE_ARCHIVE_OUTPUT_DIRECTORY ${CMAKE_CURRENT_BINARY_DIR})
set(CMAKE_LIBRARY_OUTPUT_DIRECTORY ${CMAKE_CURRENT_BINARY_DIR})
endif()
@@ -40,10 +41,6 @@ set (LIBOMPTARGET_ALL_TARGETS "${LIBOMPT
# the list of supported targets in the current system.
set (LIBOMPTARGET_SYSTEM_TARGETS "")
-# Set base directories - required for lit to locate the tests.
-set(LIBOMPTARGET_BASE_DIR ${CMAKE_CURRENT_SOURCE_DIR})
-set(LIBOMPTARGET_BINARY_DIR ${CMAKE_CURRENT_BINARY_DIR})
-
# If building this library in debug mode, we define a macro to enable
# dumping progress messages at runtime.
string( TOLOWER "${CMAKE_BUILD_TYPE}" LIBOMPTARGET_CMAKE_BUILD_TYPE)
@@ -65,6 +62,17 @@ if(NOT LIBOMPTARGET_LIBRARY_DIR)
set(LIBOMPTARGET_LIBRARY_DIR ${CMAKE_CURRENT_BINARY_DIR})
endif()
+# Definitions for testing, for reuse when testing libomptarget-nvptx.
+if(OPENMP_STANDALONE_BUILD)
+ set(LIBOMPTARGET_OPENMP_HEADER_FOLDER "${CMAKE_CURRENT_BINARY_DIR}/../runtime/src" CACHE STRING
+ "Path to folder containing omp.h")
+ set(LIBOMPTARGET_OPENMP_HOST_RTL_FOLDER "${CMAKE_CURRENT_BINARY_DIR}/../runtime/src" CACHE STRING
+ "Path to folder containing libomp.so")
+else()
+ set(LIBOMPTARGET_OPENMP_HEADER_FOLDER "${CMAKE_CURRENT_BINARY_DIR}/../runtime/src")
+endif()
+
+
# Build offloading plugins and device RTLs if they are available.
add_subdirectory(plugins)
add_subdirectory(deviceRTLs)
Modified: openmp/trunk/libomptarget/deviceRTLs/nvptx/CMakeLists.txt
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/libomptarget/deviceRTLs/nvptx/CMakeLists.txt?rev=343324&r1=343323&r2=343324&view=diff
==============================================================================
--- openmp/trunk/libomptarget/deviceRTLs/nvptx/CMakeLists.txt (original)
+++ openmp/trunk/libomptarget/deviceRTLs/nvptx/CMakeLists.txt Fri Sep 28 08:05:43 2018
@@ -132,6 +132,9 @@ if(LIBOMPTARGET_DEP_CUDA_FOUND)
set(bc_flags ${bc_flags} -Dnv_weak=weak)
endif()
+ # Create target to build all Bitcode libraries.
+ add_custom_target(omptarget-nvptx-bc)
+
# Generate a Bitcode library for all the compute capabilities the user requested.
foreach(sm ${nvptx_sm_list})
set(cuda_arch --cuda-gpu-arch=sm_${sm})
@@ -165,6 +168,7 @@ if(LIBOMPTARGET_DEP_CUDA_FOUND)
set_property(DIRECTORY APPEND PROPERTY ADDITIONAL_MAKE_CLEAN_FILES libomptarget-nvptx-sm_${sm}.bc)
add_custom_target(omptarget-nvptx-${sm}-bc ALL DEPENDS ${CMAKE_CURRENT_BINARY_DIR}/libomptarget-nvptx-sm_${sm}.bc)
+ add_dependencies(omptarget-nvptx-bc omptarget-nvptx-${sm}-bc)
# Copy library to destination.
add_custom_command(TARGET omptarget-nvptx-${sm}-bc POST_BUILD
@@ -176,6 +180,7 @@ if(LIBOMPTARGET_DEP_CUDA_FOUND)
endforeach()
endif()
+ add_subdirectory(test)
else()
libomptarget_say("Not building CUDA offloading device RTL: CUDA tools not found in the system.")
endif()
Added: openmp/trunk/libomptarget/deviceRTLs/nvptx/test/CMakeLists.txt
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/libomptarget/deviceRTLs/nvptx/test/CMakeLists.txt?rev=343324&view=auto
==============================================================================
--- openmp/trunk/libomptarget/deviceRTLs/nvptx/test/CMakeLists.txt (added)
+++ openmp/trunk/libomptarget/deviceRTLs/nvptx/test/CMakeLists.txt Fri Sep 28 08:05:43 2018
@@ -0,0 +1,26 @@
+if(NOT OPENMP_TEST_COMPILER_ID STREQUAL "Clang")
+ # Silently return, no need to annoy the user.
+ return()
+endif()
+
+set(deps omptarget-nvptx omptarget omp)
+if(LIBOMPTARGET_NVPTX_ENABLE_BCLIB)
+ set(deps ${deps} omptarget-nvptx-bc)
+endif()
+
+# Don't run by default.
+set(EXCLUDE_FROM_ALL True)
+# Run with only one thread to only launch one application to the GPU at a time.
+add_openmp_testsuite(check-libomptarget-nvptx
+ "Running libomptarget-nvptx tests" ${CMAKE_CURRENT_BINARY_DIR}
+ DEPENDS ${deps} ARGS -j1)
+
+set(LIBOMPTARGET_NVPTX_TEST_FLAGS "" CACHE STRING
+ "Extra compiler flags to send to the test compiler.")
+set(LIBOMPTARGET_NVPTX_TEST_OPENMP_FLAGS
+ "-fopenmp -fopenmp-targets=nvptx64-nvidia-cuda" CACHE STRING
+ "OpenMP compiler flags to use for testing libomptarget-nvptx.")
+
+# Configure the lit.site.cfg.in file
+set(AUTO_GEN_COMMENT "## Autogenerated by libomptarget-nvptx configuration.\n# Do not edit!")
+configure_file(lit.site.cfg.in lit.site.cfg @ONLY)
Added: openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.cfg
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.cfg?rev=343324&view=auto
==============================================================================
--- openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.cfg (added)
+++ openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.cfg Fri Sep 28 08:05:43 2018
@@ -0,0 +1,69 @@
+# -*- Python -*- vim: set ft=python ts=4 sw=4 expandtab tw=79:
+# Configuration file for the 'lit' test runner.
+
+import os
+import lit.formats
+
+# Tell pylint that we know config and lit_config exist somewhere.
+if 'PYLINT_IMPORT' in os.environ:
+ config = object()
+ lit_config = object()
+
+def prepend_library_path(name, value, sep):
+ if name in config.environment:
+ config.environment[name] = value + sep + config.environment[name]
+ else:
+ config.environment[name] = value
+
+# name: The name of this test suite.
+config.name = 'libomptarget-nvptx'
+
+# suffixes: A list of file extensions to treat as test files.
+config.suffixes = ['.c', '.cpp', '.cc']
+
+# test_source_root: The root path where tests are located.
+config.test_source_root = os.path.dirname(__file__)
+
+# test_exec_root: The root object directory where output is placed
+config.test_exec_root = config.binary_dir
+
+# test format
+config.test_format = lit.formats.ShTest()
+
+# compiler flags
+config.test_flags = " -I " + config.omp_header_directory + \
+ " -L " + config.library_dir + \
+ " --libomptarget-nvptx-path=" + config.library_dir;
+
+if config.omp_host_rtl_directory:
+ config.test_flags = config.test_flags + \
+ " -L " + config.omp_host_rtl_directory
+
+config.test_flags = config.test_flags + " " + config.test_extra_flags
+
+# Setup environment to find dynamic library at runtime.
+prepend_library_path('LD_LIBRARY_PATH', config.library_dir, ":")
+prepend_library_path('LD_LIBRARY_PATH', config.omp_host_rtl_directory, ":")
+
+# Forbid fallback to host.
+config.environment["OMP_TARGET_OFFLOAD"] = "MANDATORY"
+
+# substitutions
+config.substitutions.append(("%compilexx-run-and-check",
+ "%compilexx-and-run | " + config.libomptarget_filecheck + " %s"))
+config.substitutions.append(("%compile-run-and-check",
+ "%compile-and-run | " + config.libomptarget_filecheck + " %s"))
+config.substitutions.append(("%compilexx-and-run", "%compilexx && %run"))
+config.substitutions.append(("%compile-and-run", "%compile && %run"))
+
+config.substitutions.append(("%compilexx",
+ "%clangxx %openmp_flags %flags %s -o %t"))
+config.substitutions.append(("%compile",
+ "%clang %openmp_flags %flags %s -o %t"))
+
+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))
+config.substitutions.append(("%flags", config.test_flags))
+
+config.substitutions.append(("%run", "%t"))
Added: openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.site.cfg.in
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.site.cfg.in?rev=343324&view=auto
==============================================================================
--- openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.site.cfg.in (added)
+++ openmp/trunk/libomptarget/deviceRTLs/nvptx/test/lit.site.cfg.in Fri Sep 28 08:05:43 2018
@@ -0,0 +1,14 @@
+ at AUTO_GEN_COMMENT@
+
+config.test_c_compiler = "@OPENMP_TEST_C_COMPILER@"
+config.test_cxx_compiler = "@OPENMP_TEST_CXX_COMPILER@"
+config.test_openmp_flags = "@LIBOMPTARGET_NVPTX_TEST_OPENMP_FLAGS@"
+config.test_extra_flags = "@LIBOMPTARGET_NVPTX_TEST_FLAGS@"
+config.binary_dir = "@CMAKE_CURRENT_BINARY_DIR@"
+config.library_dir = "@LIBOMPTARGET_LIBRARY_DIR@"
+config.omp_header_directory = "@LIBOMPTARGET_OPENMP_HEADER_FOLDER@"
+config.omp_host_rtl_directory = "@LIBOMPTARGET_OPENMP_HOST_RTL_FOLDER@"
+config.libomptarget_filecheck = "@OPENMP_FILECHECK_EXECUTABLE@"
+
+# Let the main config do the real work.
+lit_config.load_config(config, "@CMAKE_CURRENT_SOURCE_DIR@/lit.cfg")
Added: openmp/trunk/libomptarget/deviceRTLs/nvptx/test/parallel/thread_limit.c
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/libomptarget/deviceRTLs/nvptx/test/parallel/thread_limit.c?rev=343324&view=auto
==============================================================================
--- openmp/trunk/libomptarget/deviceRTLs/nvptx/test/parallel/thread_limit.c (added)
+++ openmp/trunk/libomptarget/deviceRTLs/nvptx/test/parallel/thread_limit.c Fri Sep 28 08:05:43 2018
@@ -0,0 +1,77 @@
+// RUN: %compile-run-and-check
+
+#include <stdio.h>
+#include <omp.h>
+
+const int WarpSize = 32;
+const int ThreadLimit = 1 * WarpSize;
+const int NumThreads2 = 2 * WarpSize;
+const int NumThreads3 = 3 * WarpSize;
+const int MaxThreads = 1024;
+
+int main(int argc, char *argv[]) {
+ int check1[MaxThreads];
+ int check2[MaxThreads];
+ int check3[MaxThreads];
+ for (int i = 0; i < MaxThreads; i++) {
+ check1[i] = check2[i] = check3[i] = 0;
+ }
+
+ int threadLimit = -1;
+
+ #pragma omp target teams num_teams(1) thread_limit(ThreadLimit) \
+ map(check1[:], check2[:], check3[:], threadLimit)
+ {
+ threadLimit = omp_get_thread_limit();
+
+ // All parallel regions should get as many threads as specified by the
+ // thread_limit() clause.
+ #pragma omp parallel
+ {
+ check1[omp_get_thread_num()] += omp_get_num_threads();
+ }
+
+ omp_set_num_threads(NumThreads2);
+ #pragma omp parallel
+ {
+ check2[omp_get_thread_num()] += omp_get_num_threads();
+ }
+
+ #pragma omp parallel num_threads(NumThreads3)
+ {
+ check3[omp_get_thread_num()] += omp_get_num_threads();
+ }
+ }
+
+ // CHECK: threadLimit = 32
+ printf("threadLimit = %d\n", threadLimit);
+
+ // CHECK-NOT: invalid
+ for (int i = 0; i < MaxThreads; i++) {
+ if (i < ThreadLimit) {
+ if (check1[i] != ThreadLimit) {
+ printf("invalid: check1[%d] should be %d, is %d\n", i, ThreadLimit, check1[i]);
+ }
+ } else if (check1[i] != 0) {
+ printf("invalid: check1[%d] should be 0, is %d\n", i, check1[i]);
+ }
+
+ if (i < ThreadLimit) {
+ if (check2[i] != ThreadLimit) {
+ printf("invalid: check2[%d] should be %d, is %d\n", i, ThreadLimit, check2[i]);
+ }
+ } else if (check2[i] != 0) {
+ printf("invalid: check2[%d] should be 0, is %d\n", i, check2[i]);
+ }
+
+ if (i < ThreadLimit) {
+ if (check3[i] != ThreadLimit) {
+ printf("invalid: check3[%d] should be %d, is %d\n", i, ThreadLimit, check3[i]);
+ }
+ } else if (check3[i] != 0) {
+ printf("invalid: check3[%d] should be 0, is %d\n", i, check3[i]);
+ }
+ }
+
+ return 0;
+}
Modified: openmp/trunk/libomptarget/test/CMakeLists.txt
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/libomptarget/test/CMakeLists.txt?rev=343324&r1=343323&r2=343324&view=diff
==============================================================================
--- openmp/trunk/libomptarget/test/CMakeLists.txt (original)
+++ openmp/trunk/libomptarget/test/CMakeLists.txt Fri Sep 28 08:05:43 2018
@@ -14,15 +14,6 @@ endif()
add_openmp_testsuite(check-libomptarget "Running libomptarget tests" ${CMAKE_CURRENT_BINARY_DIR} DEPENDS omptarget omp)
-if(${OPENMP_STANDALONE_BUILD})
- set(LIBOMPTARGET_OPENMP_HEADER_FOLDER "${CMAKE_CURRENT_BINARY_DIR}/../../runtime/src" CACHE STRING
- "Path to folder containing omp.h")
- set(LIBOMPTARGET_OPENMP_HOST_RTL_FOLDER "${CMAKE_CURRENT_BINARY_DIR}/../../runtime/src" CACHE STRING
- "Path to folder containing libomp.so")
-else()
- set(LIBOMPTARGET_OPENMP_HEADER_FOLDER "${LIBOMPTARGET_BINARY_DIR}/../runtime/src")
-endif()
-
# Configure the lit.site.cfg.in file
set(AUTO_GEN_COMMENT "## Autogenerated by libomptarget configuration.\n# Do not edit!")
configure_file(lit.site.cfg.in lit.site.cfg @ONLY)
Modified: openmp/trunk/libomptarget/test/lit.site.cfg.in
URL: http://llvm.org/viewvc/llvm-project/openmp/trunk/libomptarget/test/lit.site.cfg.in?rev=343324&r1=343323&r2=343324&view=diff
==============================================================================
--- openmp/trunk/libomptarget/test/lit.site.cfg.in (original)
+++ openmp/trunk/libomptarget/test/lit.site.cfg.in Fri Sep 28 08:05:43 2018
@@ -16,4 +16,4 @@ config.libomptarget_filecheck = "@OPENMP
config.libomptarget_debug = @LIBOMPTARGET_DEBUG@
# Let the main config do the real work.
-lit_config.load_config(config, "@LIBOMPTARGET_BASE_DIR@/test/lit.cfg")
+lit_config.load_config(config, "@CMAKE_CURRENT_SOURCE_DIR@/lit.cfg")
More information about the Openmp-commits
mailing list