aboutsummaryrefslogtreecommitdiff
path: root/lib/clang/11.0.0/include/__clang_cuda_cmath.h
diff options
context:
space:
mode:
Diffstat (limited to 'lib/clang/11.0.0/include/__clang_cuda_cmath.h')
-rw-r--r--lib/clang/11.0.0/include/__clang_cuda_cmath.h485
1 files changed, 0 insertions, 485 deletions
diff --git a/lib/clang/11.0.0/include/__clang_cuda_cmath.h b/lib/clang/11.0.0/include/__clang_cuda_cmath.h
deleted file mode 100644
index 834a2e3..0000000
--- a/lib/clang/11.0.0/include/__clang_cuda_cmath.h
+++ /dev/null
@@ -1,485 +0,0 @@
-/*===---- __clang_cuda_cmath.h - Device-side CUDA cmath support ------------===
- *
- * 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_CUDA_CMATH_H__
-#define __CLANG_CUDA_CMATH_H__
-#ifndef __CUDA__
-#error "This file is for CUDA compilation only."
-#endif
-
-#include <limits>
-
-// CUDA lets us use various std math functions on the device side. This file
-// works in concert with __clang_cuda_math_forward_declares.h to make this work.
-//
-// Specifically, the forward-declares header declares __device__ overloads for
-// these functions in the global namespace, then pulls them into namespace std
-// with 'using' statements. Then this file implements those functions, after
-// their implementations have been pulled in.
-//
-// It's important that we declare the functions in the global namespace and pull
-// them into namespace std with using statements, as opposed to simply declaring
-// these functions in namespace std, because our device functions need to
-// overload the standard library functions, which may be declared in the global
-// namespace or in std, depending on the degree of conformance of the stdlib
-// implementation. Declaring in the global namespace and pulling into namespace
-// std covers all of the known knowns.
-
-#ifdef _OPENMP
-#define __DEVICE__ static __attribute__((always_inline))
-#else
-#define __DEVICE__ static __device__ __inline__ __attribute__((always_inline))
-#endif
-
-// For C++ 17 we need to include noexcept attribute to be compatible
-// with the header-defined version. This may be removed once
-// variant is supported.
-#if defined(_OPENMP) && defined(__cplusplus) && __cplusplus >= 201703L
-#define __NOEXCEPT noexcept
-#else
-#define __NOEXCEPT
-#endif
-
-#if !(defined(_OPENMP) && defined(__cplusplus))
-__DEVICE__ long long abs(long long __n) { return ::llabs(__n); }
-__DEVICE__ long abs(long __n) { return ::labs(__n); }
-__DEVICE__ float abs(float __x) { return ::fabsf(__x); }
-__DEVICE__ double abs(double __x) { return ::fabs(__x); }
-#endif
-// TODO: remove once variat is supported.
-#if defined(_OPENMP) && defined(__cplusplus)
-__DEVICE__ const float abs(const float __x) { return ::fabsf((float)__x); }
-__DEVICE__ const double abs(const double __x) { return ::fabs((double)__x); }
-#endif
-__DEVICE__ float acos(float __x) { return ::acosf(__x); }
-__DEVICE__ float asin(float __x) { return ::asinf(__x); }
-__DEVICE__ float atan(float __x) { return ::atanf(__x); }
-__DEVICE__ float atan2(float __x, float __y) { return ::atan2f(__x, __y); }
-__DEVICE__ float ceil(float __x) { return ::ceilf(__x); }
-__DEVICE__ float cos(float __x) { return ::cosf(__x); }
-__DEVICE__ float cosh(float __x) { return ::coshf(__x); }
-__DEVICE__ float exp(float __x) { return ::expf(__x); }
-__DEVICE__ float fabs(float __x) __NOEXCEPT { return ::fabsf(__x); }
-__DEVICE__ float floor(float __x) { return ::floorf(__x); }
-__DEVICE__ float fmod(float __x, float __y) { return ::fmodf(__x, __y); }
-// TODO: remove when variant is supported
-#ifndef _OPENMP
-__DEVICE__ int fpclassify(float __x) {
- return __builtin_fpclassify(FP_NAN, FP_INFINITE, FP_NORMAL, FP_SUBNORMAL,
- FP_ZERO, __x);
-}
-__DEVICE__ int fpclassify(double __x) {
- return __builtin_fpclassify(FP_NAN, FP_INFINITE, FP_NORMAL, FP_SUBNORMAL,
- FP_ZERO, __x);
-}
-#endif
-__DEVICE__ float frexp(float __arg, int *__exp) {
- return ::frexpf(__arg, __exp);
-}
-
-// For inscrutable reasons, the CUDA headers define these functions for us on
-// Windows.
-#ifndef _MSC_VER
-__DEVICE__ bool isinf(float __x) { return ::__isinff(__x); }
-__DEVICE__ bool isinf(double __x) { return ::__isinf(__x); }
-__DEVICE__ bool isfinite(float __x) { return ::__finitef(__x); }
-// For inscrutable reasons, __finite(), the double-precision version of
-// __finitef, does not exist when compiling for MacOS. __isfinited is available
-// everywhere and is just as good.
-__DEVICE__ bool isfinite(double __x) { return ::__isfinited(__x); }
-__DEVICE__ bool isnan(float __x) { return ::__isnanf(__x); }
-__DEVICE__ bool isnan(double __x) { return ::__isnan(__x); }
-#endif
-
-__DEVICE__ bool isgreater(float __x, float __y) {
- return __builtin_isgreater(__x, __y);
-}
-__DEVICE__ bool isgreater(double __x, double __y) {
- return __builtin_isgreater(__x, __y);
-}
-__DEVICE__ bool isgreaterequal(float __x, float __y) {
- return __builtin_isgreaterequal(__x, __y);
-}
-__DEVICE__ bool isgreaterequal(double __x, double __y) {
- return __builtin_isgreaterequal(__x, __y);
-}
-__DEVICE__ bool isless(float __x, float __y) {
- return __builtin_isless(__x, __y);
-}
-__DEVICE__ bool isless(double __x, double __y) {
- return __builtin_isless(__x, __y);
-}
-__DEVICE__ bool islessequal(float __x, float __y) {
- return __builtin_islessequal(__x, __y);
-}
-__DEVICE__ bool islessequal(double __x, double __y) {
- return __builtin_islessequal(__x, __y);
-}
-__DEVICE__ bool islessgreater(float __x, float __y) {
- return __builtin_islessgreater(__x, __y);
-}
-__DEVICE__ bool islessgreater(double __x, double __y) {
- return __builtin_islessgreater(__x, __y);
-}
-__DEVICE__ bool isnormal(float __x) { return __builtin_isnormal(__x); }
-__DEVICE__ bool isnormal(double __x) { return __builtin_isnormal(__x); }
-__DEVICE__ bool isunordered(float __x, float __y) {
- return __builtin_isunordered(__x, __y);
-}
-__DEVICE__ bool isunordered(double __x, double __y) {
- return __builtin_isunordered(__x, __y);
-}
-__DEVICE__ float ldexp(float __arg, int __exp) {
- return ::ldexpf(__arg, __exp);
-}
-__DEVICE__ float log(float __x) { return ::logf(__x); }
-__DEVICE__ float log10(float __x) { return ::log10f(__x); }
-__DEVICE__ float modf(float __x, float *__iptr) { return ::modff(__x, __iptr); }
-__DEVICE__ float pow(float __base, float __exp) {
- return ::powf(__base, __exp);
-}
-__DEVICE__ float pow(float __base, int __iexp) {
- return ::powif(__base, __iexp);
-}
-__DEVICE__ double pow(double __base, int __iexp) {
- return ::powi(__base, __iexp);
-}
-__DEVICE__ bool signbit(float __x) { return ::__signbitf(__x); }
-__DEVICE__ bool signbit(double __x) { return ::__signbitd(__x); }
-__DEVICE__ float sin(float __x) { return ::sinf(__x); }
-__DEVICE__ float sinh(float __x) { return ::sinhf(__x); }
-__DEVICE__ float sqrt(float __x) { return ::sqrtf(__x); }
-__DEVICE__ float tan(float __x) { return ::tanf(__x); }
-__DEVICE__ float tanh(float __x) { return ::tanhf(__x); }
-
-// Notably missing above is nexttoward. We omit it because
-// libdevice doesn't provide an implementation, and we don't want to be in the
-// business of implementing tricky libm functions in this header.
-
-// Now we've defined everything we promised we'd define in
-// __clang_cuda_math_forward_declares.h. We need to do two additional things to
-// fix up our math functions.
-//
-// 1) Define __device__ overloads for e.g. sin(int). The CUDA headers define
-// only sin(float) and sin(double), which means that e.g. sin(0) is
-// ambiguous.
-//
-// 2) Pull the __device__ overloads of "foobarf" math functions into namespace
-// std. These are defined in the CUDA headers in the global namespace,
-// independent of everything else we've done here.
-
-// We can't use std::enable_if, because we want to be pre-C++11 compatible. But
-// we go ahead and unconditionally define functions that are only available when
-// compiling for C++11 to match the behavior of the CUDA headers.
-template<bool __B, class __T = void>
-struct __clang_cuda_enable_if {};
-
-template <class __T> struct __clang_cuda_enable_if<true, __T> {
- typedef __T type;
-};
-
-// Defines an overload of __fn that accepts one integral argument, calls
-// __fn((double)x), and returns __retty.
-#define __CUDA_CLANG_FN_INTEGER_OVERLOAD_1(__retty, __fn) \
- template <typename __T> \
- __DEVICE__ \
- typename __clang_cuda_enable_if<std::numeric_limits<__T>::is_integer, \
- __retty>::type \
- __fn(__T __x) { \
- return ::__fn((double)__x); \
- }
-
-// Defines an overload of __fn that accepts one two arithmetic arguments, calls
-// __fn((double)x, (double)y), and returns a double.
-//
-// Note this is different from OVERLOAD_1, which generates an overload that
-// accepts only *integral* arguments.
-#define __CUDA_CLANG_FN_INTEGER_OVERLOAD_2(__retty, __fn) \
- template <typename __T1, typename __T2> \
- __DEVICE__ typename __clang_cuda_enable_if< \
- std::numeric_limits<__T1>::is_specialized && \
- std::numeric_limits<__T2>::is_specialized, \
- __retty>::type \
- __fn(__T1 __x, __T2 __y) { \
- return __fn((double)__x, (double)__y); \
- }
-
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, acos)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, acosh)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, asin)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, asinh)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, atan)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, atan2);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, atanh)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, cbrt)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, ceil)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, copysign);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, cos)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, cosh)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, erf)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, erfc)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, exp)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, exp2)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, expm1)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, fabs)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, fdim);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, floor)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, fmax);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, fmin);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, fmod);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(int, fpclassify)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, hypot);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(int, ilogb)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(bool, isfinite)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(bool, isgreater);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(bool, isgreaterequal);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(bool, isinf);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(bool, isless);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(bool, islessequal);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(bool, islessgreater);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(bool, isnan);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(bool, isnormal)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(bool, isunordered);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, lgamma)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, log)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, log10)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, log1p)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, log2)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, logb)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(long long, llrint)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(long long, llround)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(long, lrint)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(long, lround)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, nearbyint);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, nextafter);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, pow);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_2(double, remainder);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, rint);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, round);
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(bool, signbit)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, sin)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, sinh)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, sqrt)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, tan)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, tanh)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, tgamma)
-__CUDA_CLANG_FN_INTEGER_OVERLOAD_1(double, trunc);
-
-#undef __CUDA_CLANG_FN_INTEGER_OVERLOAD_1
-#undef __CUDA_CLANG_FN_INTEGER_OVERLOAD_2
-
-// Overloads for functions that don't match the patterns expected by
-// __CUDA_CLANG_FN_INTEGER_OVERLOAD_{1,2}.
-template <typename __T1, typename __T2, typename __T3>
-__DEVICE__ typename __clang_cuda_enable_if<
- std::numeric_limits<__T1>::is_specialized &&
- std::numeric_limits<__T2>::is_specialized &&
- std::numeric_limits<__T3>::is_specialized,
- double>::type
-fma(__T1 __x, __T2 __y, __T3 __z) {
- return std::fma((double)__x, (double)__y, (double)__z);
-}
-
-template <typename __T>
-__DEVICE__ typename __clang_cuda_enable_if<std::numeric_limits<__T>::is_integer,
- double>::type
-frexp(__T __x, int *__exp) {
- return std::frexp((double)__x, __exp);
-}
-
-template <typename __T>
-__DEVICE__ typename __clang_cuda_enable_if<std::numeric_limits<__T>::is_integer,
- double>::type
-ldexp(__T __x, int __exp) {
- return std::ldexp((double)__x, __exp);
-}
-
-template <typename __T1, typename __T2>
-__DEVICE__ typename __clang_cuda_enable_if<
- std::numeric_limits<__T1>::is_specialized &&
- std::numeric_limits<__T2>::is_specialized,
- double>::type
-remquo(__T1 __x, __T2 __y, int *__quo) {
- return std::remquo((double)__x, (double)__y, __quo);
-}
-
-template <typename __T>
-__DEVICE__ typename __clang_cuda_enable_if<std::numeric_limits<__T>::is_integer,
- double>::type
-scalbln(__T __x, long __exp) {
- return std::scalbln((double)__x, __exp);
-}
-
-template <typename __T>
-__DEVICE__ typename __clang_cuda_enable_if<std::numeric_limits<__T>::is_integer,
- double>::type
-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;
-using ::asin;
-using ::asinh;
-using ::atan;
-using ::atan2;
-using ::atanh;
-using ::cbrt;
-using ::ceil;
-using ::copysign;
-using ::cos;
-using ::cosh;
-using ::erf;
-using ::erfc;
-using ::exp;
-using ::exp2;
-using ::expm1;
-using ::fabs;
-using ::fdim;
-using ::floor;
-using ::fma;
-using ::fmax;
-using ::fmin;
-using ::fmod;
-using ::fpclassify;
-using ::frexp;
-using ::hypot;
-using ::ilogb;
-using ::isfinite;
-using ::isgreater;
-using ::isgreaterequal;
-using ::isless;
-using ::islessequal;
-using ::islessgreater;
-using ::isnormal;
-using ::isunordered;
-using ::ldexp;
-using ::lgamma;
-using ::llrint;
-using ::llround;
-using ::log;
-using ::log10;
-using ::log1p;
-using ::log2;
-using ::logb;
-using ::lrint;
-using ::lround;
-using ::nearbyint;
-using ::nextafter;
-using ::pow;
-using ::remainder;
-using ::remquo;
-using ::rint;
-using ::round;
-using ::scalbln;
-using ::scalbn;
-using ::signbit;
-using ::sin;
-using ::sinh;
-using ::sqrt;
-using ::tan;
-using ::tanh;
-using ::tgamma;
-using ::trunc;
-
-// Well this is fun: We need to pull these symbols in for libc++, but we can't
-// pull them in with libstdc++, because its ::isinf and ::isnan are different
-// than its std::isinf and std::isnan.
-#ifndef __GLIBCXX__
-using ::isinf;
-using ::isnan;
-#endif
-
-// Finally, pull the "foobarf" functions that CUDA defines in its headers into
-// namespace std.
-using ::acosf;
-using ::acoshf;
-using ::asinf;
-using ::asinhf;
-using ::atan2f;
-using ::atanf;
-using ::atanhf;
-using ::cbrtf;
-using ::ceilf;
-using ::copysignf;
-using ::cosf;
-using ::coshf;
-using ::erfcf;
-using ::erff;
-using ::exp2f;
-using ::expf;
-using ::expm1f;
-using ::fabsf;
-using ::fdimf;
-using ::floorf;
-using ::fmaf;
-using ::fmaxf;
-using ::fminf;
-using ::fmodf;
-using ::frexpf;
-using ::hypotf;
-using ::ilogbf;
-using ::ldexpf;
-using ::lgammaf;
-using ::llrintf;
-using ::llroundf;
-using ::log10f;
-using ::log1pf;
-using ::log2f;
-using ::logbf;
-using ::logf;
-using ::lrintf;
-using ::lroundf;
-using ::modff;
-using ::nearbyintf;
-using ::nextafterf;
-using ::powf;
-using ::remainderf;
-using ::remquof;
-using ::rintf;
-using ::roundf;
-// TODO: remove once variant is supported
-#ifndef _OPENMP
-using ::scalblnf;
-#endif
-using ::scalbnf;
-using ::sinf;
-using ::sinhf;
-using ::sqrtf;
-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 __NOEXCEPT
-#undef __DEVICE__
-
-#endif
generated by cgit v1.2.3 (git 2.25.1) at 2025-12-20 13:52:08 +0000