From d6065d4875d6e09eff366ca1e6aba1dd30f66023 Mon Sep 17 00:00:00 2001 From: Gheorghe-Teodor Bercea Date: Mon, 13 May 2019 22:11:44 +0000 Subject: [PATCH] [OpenMP][Clang][BugFix] Split declares and math functions inclusion. Summary: This patches fixes an issue in which the __clang_cuda_cmath.h header is being included even when cmath or math.h headers are not included. Reviewers: jdoerfert, ABataev, hfinkel, caomhin, tra Reviewed By: tra Subscribers: tra, mgorny, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D61765 git-svn-id: https://llvm.org/svn/llvm-project/cfe/trunk@360626 91177308-0d34-0410-b5e6-96231b3b80d8 --- lib/Driver/ToolChains/Clang.cpp | 2 +- lib/Headers/CMakeLists.txt | 1 + lib/Headers/__clang_cuda_cmath.h | 2 ++ lib/Headers/__clang_cuda_device_functions.h | 2 ++ .../__clang_cuda_math_forward_declares.h | 8 +++-- .../openmp_wrappers/__clang_openmp_math.h | 9 ----- .../__clang_openmp_math_declares.h | 33 +++++++++++++++++++ lib/Headers/openmp_wrappers/cmath | 2 -- lib/Headers/openmp_wrappers/math.h | 2 -- test/Headers/Inputs/include/cstdlib | 16 +++++++++ test/Headers/nvptx_device_cmath_functions.c | 2 +- test/Headers/nvptx_device_cmath_functions.cpp | 3 +- test/Headers/nvptx_device_math_functions.c | 2 +- test/Headers/nvptx_device_math_functions.cpp | 3 +- 14 files changed, 66 insertions(+), 21 deletions(-) create mode 100644 lib/Headers/openmp_wrappers/__clang_openmp_math_declares.h create mode 100644 test/Headers/Inputs/include/cstdlib diff --git a/lib/Driver/ToolChains/Clang.cpp b/lib/Driver/ToolChains/Clang.cpp index 2e17b4fc69..303c6a96ee 100644 --- a/lib/Driver/ToolChains/Clang.cpp +++ b/lib/Driver/ToolChains/Clang.cpp @@ -1166,7 +1166,7 @@ void Clang::AddPreprocessingOptions(Compilation &C, const JobAction &JA, } CmdArgs.push_back("-include"); - CmdArgs.push_back("__clang_openmp_math.h"); + CmdArgs.push_back("__clang_openmp_math_declares.h"); } // Add -i* options, and automatically translate to diff --git a/lib/Headers/CMakeLists.txt b/lib/Headers/CMakeLists.txt index 5db3c7cd17..392ca2ae39 100644 --- a/lib/Headers/CMakeLists.txt +++ b/lib/Headers/CMakeLists.txt @@ -132,6 +132,7 @@ set(openmp_wrapper_files openmp_wrappers/math.h openmp_wrappers/cmath openmp_wrappers/__clang_openmp_math.h + openmp_wrappers/__clang_openmp_math_declares.h ) set(output_dir ${LLVM_LIBRARY_OUTPUT_INTDIR}/clang/${CLANG_VERSION}/include) diff --git a/lib/Headers/__clang_cuda_cmath.h b/lib/Headers/__clang_cuda_cmath.h index 82e52d1466..98d0c723f1 100644 --- a/lib/Headers/__clang_cuda_cmath.h +++ b/lib/Headers/__clang_cuda_cmath.h @@ -36,8 +36,10 @@ #define __DEVICE__ static __device__ __inline__ __attribute__((always_inline)) #endif +#if !(defined(_OPENMP) && defined(__cplusplus)) __DEVICE__ long long abs(long long __n) { return ::llabs(__n); } __DEVICE__ long abs(long __n) { return ::labs(__n); } +#endif __DEVICE__ float abs(float __x) { return ::fabsf(__x); } __DEVICE__ double abs(double __x) { return ::fabs(__x); } __DEVICE__ float acos(float __x) { return ::acosf(__x); } diff --git a/lib/Headers/__clang_cuda_device_functions.h b/lib/Headers/__clang_cuda_device_functions.h index a4c385b4b7..af3cea739d 100644 --- a/lib/Headers/__clang_cuda_device_functions.h +++ b/lib/Headers/__clang_cuda_device_functions.h @@ -1493,8 +1493,10 @@ __DEVICE__ double cbrt(double __a) { return __nv_cbrt(__a); } __DEVICE__ float cbrtf(float __a) { return __nv_cbrtf(__a); } __DEVICE__ double ceil(double __a) { return __nv_ceil(__a); } __DEVICE__ float ceilf(float __a) { return __nv_ceilf(__a); } +#ifndef _OPENMP __DEVICE__ int clock() { return __nvvm_read_ptx_sreg_clock(); } __DEVICE__ long long clock64() { return __nvvm_read_ptx_sreg_clock64(); } +#endif __DEVICE__ double copysign(double __a, double __b) { return __nv_copysign(__a, __b); } diff --git a/lib/Headers/__clang_cuda_math_forward_declares.h b/lib/Headers/__clang_cuda_math_forward_declares.h index e1a4e9fe1f..4bc6b9f058 100644 --- a/lib/Headers/__clang_cuda_math_forward_declares.h +++ b/lib/Headers/__clang_cuda_math_forward_declares.h @@ -27,11 +27,13 @@ static __inline__ __attribute__((always_inline)) __attribute__((device)) #endif -__DEVICE__ double abs(double); -__DEVICE__ float abs(float); -__DEVICE__ int abs(int); +#if !(defined(_OPENMP) && defined(__cplusplus)) __DEVICE__ long abs(long); __DEVICE__ long long abs(long long); +#endif +__DEVICE__ int abs(int); +__DEVICE__ double abs(double); +__DEVICE__ float abs(float); __DEVICE__ double acos(double); __DEVICE__ float acos(float); __DEVICE__ double acosh(double); diff --git a/lib/Headers/openmp_wrappers/__clang_openmp_math.h b/lib/Headers/openmp_wrappers/__clang_openmp_math.h index 4cfeb397da..5d7ce9a965 100644 --- a/lib/Headers/openmp_wrappers/__clang_openmp_math.h +++ b/lib/Headers/openmp_wrappers/__clang_openmp_math.h @@ -22,15 +22,6 @@ #define __CUDA__ -#if defined(__cplusplus) - #include <__clang_cuda_math_forward_declares.h> -#endif - -/// Include declarations for libdevice functions. -#include <__clang_cuda_libdevice_declares.h> -/// Provide definitions for these functions. -#include <__clang_cuda_device_functions.h> - #if defined(__cplusplus) #include <__clang_cuda_cmath.h> #endif diff --git a/lib/Headers/openmp_wrappers/__clang_openmp_math_declares.h b/lib/Headers/openmp_wrappers/__clang_openmp_math_declares.h new file mode 100644 index 0000000000..a422c98bf9 --- /dev/null +++ b/lib/Headers/openmp_wrappers/__clang_openmp_math_declares.h @@ -0,0 +1,33 @@ +/*===---- __clang_openmp_math_declares.h - OpenMP math declares ------------=== + * + * Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. + * See https://llvm.org/LICENSE.txt for license information. + * SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception + * + *===-----------------------------------------------------------------------=== + */ + +#ifndef __CLANG_OPENMP_MATH_DECLARES_H__ +#define __CLANG_OPENMP_MATH_DECLARES_H__ + +#ifndef _OPENMP +#error "This file is for OpenMP compilation only." +#endif + +#if defined(__NVPTX__) && defined(_OPENMP) + +#define __CUDA__ + +#if defined(__cplusplus) + #include <__clang_cuda_math_forward_declares.h> +#endif + +/// Include declarations for libdevice functions. +#include <__clang_cuda_libdevice_declares.h> +/// Provide definitions for these functions. +#include <__clang_cuda_device_functions.h> + +#undef __CUDA__ + +#endif +#endif diff --git a/lib/Headers/openmp_wrappers/cmath b/lib/Headers/openmp_wrappers/cmath index 0f34375d45..a5183a1d8d 100644 --- a/lib/Headers/openmp_wrappers/cmath +++ b/lib/Headers/openmp_wrappers/cmath @@ -7,9 +7,7 @@ *===-----------------------------------------------------------------------=== */ -#ifndef __cplusplus #include <__clang_openmp_math.h> -#endif #ifndef __CLANG_NO_HOST_MATH__ #include_next diff --git a/lib/Headers/openmp_wrappers/math.h b/lib/Headers/openmp_wrappers/math.h index d9f798943c..d2786ecb24 100644 --- a/lib/Headers/openmp_wrappers/math.h +++ b/lib/Headers/openmp_wrappers/math.h @@ -7,9 +7,7 @@ *===-----------------------------------------------------------------------=== */ -#ifndef __cplusplus #include <__clang_openmp_math.h> -#endif #ifndef __CLANG_NO_HOST_MATH__ #include_next diff --git a/test/Headers/Inputs/include/cstdlib b/test/Headers/Inputs/include/cstdlib new file mode 100644 index 0000000000..4dc1ff6cee --- /dev/null +++ b/test/Headers/Inputs/include/cstdlib @@ -0,0 +1,16 @@ +#pragma once + +extern int abs (int __x) __attribute__ ((__const__)) ; +extern long int labs (long int __x) __attribute__ ((__const__)) ; + +namespace std +{ + +using ::abs; + +inline long +abs(long __i) { return __builtin_labs(__i); } + +inline long long +abs(long long __x) { return __builtin_llabs (__x); } +} diff --git a/test/Headers/nvptx_device_cmath_functions.c b/test/Headers/nvptx_device_cmath_functions.c index 828df04823..54e60a1da2 100644 --- a/test/Headers/nvptx_device_cmath_functions.c +++ b/test/Headers/nvptx_device_cmath_functions.c @@ -4,7 +4,7 @@ // REQUIRES: nvptx-registered-target // RUN: %clang_cc1 -internal-isystem %S/Inputs/include -include cmath -fopenmp -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc -// RUN: %clang_cc1 -internal-isystem %S/../../lib/Headers/openmp_wrappers -include __clang_openmp_math.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -include cmath -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck -check-prefix CHECK-YES %s +// RUN: %clang_cc1 -internal-isystem %S/../../lib/Headers/openmp_wrappers -include __clang_openmp_math_declares.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -include cmath -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck -check-prefix CHECK-YES %s #include diff --git a/test/Headers/nvptx_device_cmath_functions.cpp b/test/Headers/nvptx_device_cmath_functions.cpp index cdf93d89c5..48795feb9b 100644 --- a/test/Headers/nvptx_device_cmath_functions.cpp +++ b/test/Headers/nvptx_device_cmath_functions.cpp @@ -4,9 +4,10 @@ // REQUIRES: nvptx-registered-target // RUN: %clang_cc1 -internal-isystem %S/Inputs/include -include cmath -x c++ -fopenmp -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc -// RUN: %clang_cc1 -internal-isystem %S/../../lib/Headers/openmp_wrappers -include __clang_openmp_math.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -include cmath -internal-isystem %S/Inputs/include -include stdlib.h -x c++ -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck -check-prefix CHECK-YES %s +// RUN: %clang_cc1 -internal-isystem %S/../../lib/Headers/openmp_wrappers -include __clang_openmp_math_declares.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -include cmath -internal-isystem %S/Inputs/include -include stdlib.h -x c++ -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck -check-prefix CHECK-YES %s #include +#include void test_sqrt(double a1) { #pragma omp target diff --git a/test/Headers/nvptx_device_math_functions.c b/test/Headers/nvptx_device_math_functions.c index 428da244a1..22767d0cde 100644 --- a/test/Headers/nvptx_device_math_functions.c +++ b/test/Headers/nvptx_device_math_functions.c @@ -4,7 +4,7 @@ // REQUIRES: nvptx-registered-target // RUN: %clang_cc1 -internal-isystem %S/Inputs/include -include math.h -fopenmp -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc -// RUN: %clang_cc1 -internal-isystem %S/../../lib/Headers/openmp_wrappers -include __clang_openmp_math.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -include math.h -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck -check-prefix CHECK-YES %s +// RUN: %clang_cc1 -internal-isystem %S/../../lib/Headers/openmp_wrappers -include __clang_openmp_math_declares.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -include math.h -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck -check-prefix CHECK-YES %s #include diff --git a/test/Headers/nvptx_device_math_functions.cpp b/test/Headers/nvptx_device_math_functions.cpp index eadf7b2403..a4f637ba6b 100644 --- a/test/Headers/nvptx_device_math_functions.cpp +++ b/test/Headers/nvptx_device_math_functions.cpp @@ -4,8 +4,9 @@ // REQUIRES: nvptx-registered-target // RUN: %clang_cc1 -internal-isystem %S/Inputs/include -include math.h -x c++ -fopenmp -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc -// RUN: %clang_cc1 -internal-isystem %S/../../lib/Headers/openmp_wrappers -include __clang_openmp_math.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -include math.h -internal-isystem %S/Inputs/include -include stdlib.h -include limits -x c++ -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck -check-prefix CHECK-YES %s +// RUN: %clang_cc1 -internal-isystem %S/../../lib/Headers/openmp_wrappers -include __clang_openmp_math_declares.h -internal-isystem %S/../../lib/Headers/openmp_wrappers -include math.h -internal-isystem %S/Inputs/include -include stdlib.h -include limits -include cstdlib -x c++ -fopenmp -triple nvptx64-nvidia-cuda -aux-triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck -check-prefix CHECK-YES %s +#include #include void test_sqrt(double a1) { -- 2.40.0