Author: jlebar Date: Sat Oct 8 17:16:03 2016 New Revision: 283678 URL: http://llvm.org/viewvc/llvm-project?rev=283678&view=rev Log: [CUDA] Declare our __device__ math functions in the same inline namespace as our standard library.
Summary: Currently we declare our inline __device__ math functions in namespace std. But libstdc++ and libc++ declare these functions in an inline namespace inside namespace std. We need to match this because, in a later patch, we want to get e.g. <complex> to use our device overloads, and it only will if those overloads are in the right inline namespace. Reviewers: tra Subscribers: cfe-commits, jhen Differential Revision: https://reviews.llvm.org/D24977 Modified: cfe/trunk/lib/Headers/__clang_cuda_cmath.h cfe/trunk/lib/Headers/__clang_cuda_math_forward_declares.h Modified: cfe/trunk/lib/Headers/__clang_cuda_cmath.h URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Headers/__clang_cuda_cmath.h?rev=283678&r1=283677&r2=283678&view=diff ============================================================================== --- cfe/trunk/lib/Headers/__clang_cuda_cmath.h (original) +++ cfe/trunk/lib/Headers/__clang_cuda_cmath.h Sat Oct 8 17:16:03 2016 @@ -316,7 +316,19 @@ scalbn(__T __x, int __exp) { return std::scalbn((double)__x, __exp); } +// We need to define these overloads in exactly the namespace our standard +// library uses (including the right inline namespace), otherwise they won't be +// picked up by other functions in the standard library (e.g. functions in +// <complex>). Thus the ugliness below. +#ifdef _LIBCPP_BEGIN_NAMESPACE_STD +_LIBCPP_BEGIN_NAMESPACE_STD +#else namespace std { +#ifdef _GLIBCXX_BEGIN_NAMESPACE_VERSION +_GLIBCXX_BEGIN_NAMESPACE_VERSION +#endif +#endif + // Pull the new overloads we defined above into namespace std. using ::acos; using ::acosh; @@ -451,7 +463,15 @@ using ::tanf; using ::tanhf; using ::tgammaf; using ::truncf; -} + +#ifdef _LIBCPP_END_NAMESPACE_STD +_LIBCPP_END_NAMESPACE_STD +#else +#ifdef _GLIBCXX_BEGIN_NAMESPACE_VERSION +_GLIBCXX_END_NAMESPACE_VERSION +#endif +} // namespace std +#endif #undef __DEVICE__ Modified: cfe/trunk/lib/Headers/__clang_cuda_math_forward_declares.h URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Headers/__clang_cuda_math_forward_declares.h?rev=283678&r1=283677&r2=283678&view=diff ============================================================================== --- cfe/trunk/lib/Headers/__clang_cuda_math_forward_declares.h (original) +++ cfe/trunk/lib/Headers/__clang_cuda_math_forward_declares.h Sat Oct 8 17:16:03 2016 @@ -185,7 +185,19 @@ __DEVICE__ float tgamma(float); __DEVICE__ double trunc(double); __DEVICE__ float trunc(float); +// We need to define these overloads in exactly the namespace our standard +// library uses (including the right inline namespace), otherwise they won't be +// picked up by other functions in the standard library (e.g. functions in +// <complex>). Thus the ugliness below. +#ifdef _LIBCPP_BEGIN_NAMESPACE_STD +_LIBCPP_BEGIN_NAMESPACE_STD +#else namespace std { +#ifdef _GLIBCXX_BEGIN_NAMESPACE_VERSION +_GLIBCXX_BEGIN_NAMESPACE_VERSION +#endif +#endif + using ::abs; using ::acos; using ::acosh; @@ -259,7 +271,15 @@ using ::tan; using ::tanh; using ::tgamma; using ::trunc; + +#ifdef _LIBCPP_END_NAMESPACE_STD +_LIBCPP_END_NAMESPACE_STD +#else +#ifdef _GLIBCXX_BEGIN_NAMESPACE_VERSION +_GLIBCXX_END_NAMESPACE_VERSION +#endif } // namespace std +#endif #pragma pop_macro("__DEVICE__") _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
