309 lines
9.1 KiB
C++
309 lines
9.1 KiB
C++
/*===- __clang_math_forward_declares.h - Prototypes of __device__ math fns --===
|
|
*
|
|
* 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_MATH_FORWARD_DECLARES_H__
|
|
#define __CLANG__CUDA_MATH_FORWARD_DECLARES_H__
|
|
#ifndef __CUDA__
|
|
#error "This file is for CUDA compilation only."
|
|
#endif
|
|
|
|
// This file forward-declares of some math functions we (or the CUDA headers)
|
|
// will define later. We need to do this, and do it before cmath is included,
|
|
// because the standard library may have constexpr math functions. In the
|
|
// absence of a prior __device__ decl, those constexpr functions may become
|
|
// implicitly host+device. host+device functions can't be overloaded, so that
|
|
// would preclude the use of our own __device__ overloads for these functions.
|
|
|
|
#pragma push_macro("__DEVICE__")
|
|
#ifdef _OPENMP
|
|
#define __DEVICE__ static __inline__ __attribute__((always_inline))
|
|
#else
|
|
#define __DEVICE__ \
|
|
static __inline__ __attribute__((always_inline)) __attribute__((device))
|
|
#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 abs(long);
|
|
__DEVICE__ long long abs(long long);
|
|
__DEVICE__ double abs(double);
|
|
__DEVICE__ float abs(float);
|
|
#endif
|
|
// While providing the CUDA declarations and definitions for math functions,
|
|
// we may manually define additional functions.
|
|
// TODO: Once variant is supported the additional functions will have
|
|
// to be removed.
|
|
#if defined(_OPENMP) && defined(__cplusplus)
|
|
__DEVICE__ const double abs(const double);
|
|
__DEVICE__ const float abs(const float);
|
|
#endif
|
|
__DEVICE__ int abs(int) __NOEXCEPT;
|
|
__DEVICE__ double acos(double);
|
|
__DEVICE__ float acos(float);
|
|
__DEVICE__ double acosh(double);
|
|
__DEVICE__ float acosh(float);
|
|
__DEVICE__ double asin(double);
|
|
__DEVICE__ float asin(float);
|
|
__DEVICE__ double asinh(double);
|
|
__DEVICE__ float asinh(float);
|
|
__DEVICE__ double atan2(double, double);
|
|
__DEVICE__ float atan2(float, float);
|
|
__DEVICE__ double atan(double);
|
|
__DEVICE__ float atan(float);
|
|
__DEVICE__ double atanh(double);
|
|
__DEVICE__ float atanh(float);
|
|
__DEVICE__ double cbrt(double);
|
|
__DEVICE__ float cbrt(float);
|
|
__DEVICE__ double ceil(double);
|
|
__DEVICE__ float ceil(float);
|
|
__DEVICE__ double copysign(double, double);
|
|
__DEVICE__ float copysign(float, float);
|
|
__DEVICE__ double cos(double);
|
|
__DEVICE__ float cos(float);
|
|
__DEVICE__ double cosh(double);
|
|
__DEVICE__ float cosh(float);
|
|
__DEVICE__ double erfc(double);
|
|
__DEVICE__ float erfc(float);
|
|
__DEVICE__ double erf(double);
|
|
__DEVICE__ float erf(float);
|
|
__DEVICE__ double exp2(double);
|
|
__DEVICE__ float exp2(float);
|
|
__DEVICE__ double exp(double);
|
|
__DEVICE__ float exp(float);
|
|
__DEVICE__ double expm1(double);
|
|
__DEVICE__ float expm1(float);
|
|
__DEVICE__ double fabs(double) __NOEXCEPT;
|
|
__DEVICE__ float fabs(float) __NOEXCEPT;
|
|
__DEVICE__ double fdim(double, double);
|
|
__DEVICE__ float fdim(float, float);
|
|
__DEVICE__ double floor(double);
|
|
__DEVICE__ float floor(float);
|
|
__DEVICE__ double fma(double, double, double);
|
|
__DEVICE__ float fma(float, float, float);
|
|
__DEVICE__ double fmax(double, double);
|
|
__DEVICE__ float fmax(float, float);
|
|
__DEVICE__ double fmin(double, double);
|
|
__DEVICE__ float fmin(float, float);
|
|
__DEVICE__ double fmod(double, double);
|
|
__DEVICE__ float fmod(float, float);
|
|
__DEVICE__ int fpclassify(double);
|
|
__DEVICE__ int fpclassify(float);
|
|
__DEVICE__ double frexp(double, int *);
|
|
__DEVICE__ float frexp(float, int *);
|
|
__DEVICE__ double hypot(double, double);
|
|
__DEVICE__ float hypot(float, float);
|
|
__DEVICE__ int ilogb(double);
|
|
__DEVICE__ int ilogb(float);
|
|
#ifdef _MSC_VER
|
|
__DEVICE__ bool isfinite(long double);
|
|
#endif
|
|
__DEVICE__ bool isfinite(double);
|
|
__DEVICE__ bool isfinite(float);
|
|
__DEVICE__ bool isgreater(double, double);
|
|
__DEVICE__ bool isgreaterequal(double, double);
|
|
__DEVICE__ bool isgreaterequal(float, float);
|
|
__DEVICE__ bool isgreater(float, float);
|
|
#ifdef _MSC_VER
|
|
__DEVICE__ bool isinf(long double);
|
|
#endif
|
|
__DEVICE__ bool isinf(double);
|
|
__DEVICE__ bool isinf(float);
|
|
__DEVICE__ bool isless(double, double);
|
|
__DEVICE__ bool islessequal(double, double);
|
|
__DEVICE__ bool islessequal(float, float);
|
|
__DEVICE__ bool isless(float, float);
|
|
__DEVICE__ bool islessgreater(double, double);
|
|
__DEVICE__ bool islessgreater(float, float);
|
|
#ifdef _MSC_VER
|
|
__DEVICE__ bool isnan(long double);
|
|
#endif
|
|
__DEVICE__ bool isnan(double);
|
|
__DEVICE__ bool isnan(float);
|
|
__DEVICE__ bool isnormal(double);
|
|
__DEVICE__ bool isnormal(float);
|
|
__DEVICE__ bool isunordered(double, double);
|
|
__DEVICE__ bool isunordered(float, float);
|
|
__DEVICE__ long labs(long) __NOEXCEPT;
|
|
__DEVICE__ double ldexp(double, int);
|
|
__DEVICE__ float ldexp(float, int);
|
|
__DEVICE__ double lgamma(double);
|
|
__DEVICE__ float lgamma(float);
|
|
__DEVICE__ long long llabs(long long) __NOEXCEPT;
|
|
__DEVICE__ long long llrint(double);
|
|
__DEVICE__ long long llrint(float);
|
|
__DEVICE__ double log10(double);
|
|
__DEVICE__ float log10(float);
|
|
__DEVICE__ double log1p(double);
|
|
__DEVICE__ float log1p(float);
|
|
__DEVICE__ double log2(double);
|
|
__DEVICE__ float log2(float);
|
|
__DEVICE__ double logb(double);
|
|
__DEVICE__ float logb(float);
|
|
#if defined(_OPENMP) && defined(__cplusplus)
|
|
__DEVICE__ long double log(long double);
|
|
#endif
|
|
__DEVICE__ double log(double);
|
|
__DEVICE__ float log(float);
|
|
__DEVICE__ long lrint(double);
|
|
__DEVICE__ long lrint(float);
|
|
__DEVICE__ long lround(double);
|
|
__DEVICE__ long lround(float);
|
|
__DEVICE__ long long llround(float); // No llround(double).
|
|
__DEVICE__ double modf(double, double *);
|
|
__DEVICE__ float modf(float, float *);
|
|
__DEVICE__ double nan(const char *);
|
|
__DEVICE__ float nanf(const char *);
|
|
__DEVICE__ double nearbyint(double);
|
|
__DEVICE__ float nearbyint(float);
|
|
__DEVICE__ double nextafter(double, double);
|
|
__DEVICE__ float nextafter(float, float);
|
|
__DEVICE__ double pow(double, double);
|
|
__DEVICE__ double pow(double, int);
|
|
__DEVICE__ float pow(float, float);
|
|
__DEVICE__ float pow(float, int);
|
|
__DEVICE__ double remainder(double, double);
|
|
__DEVICE__ float remainder(float, float);
|
|
__DEVICE__ double remquo(double, double, int *);
|
|
__DEVICE__ float remquo(float, float, int *);
|
|
__DEVICE__ double rint(double);
|
|
__DEVICE__ float rint(float);
|
|
__DEVICE__ double round(double);
|
|
__DEVICE__ float round(float);
|
|
__DEVICE__ double scalbln(double, long);
|
|
__DEVICE__ float scalbln(float, long);
|
|
__DEVICE__ double scalbn(double, int);
|
|
__DEVICE__ float scalbn(float, int);
|
|
__DEVICE__ bool signbit(double);
|
|
__DEVICE__ bool signbit(float);
|
|
__DEVICE__ double sin(double);
|
|
__DEVICE__ float sin(float);
|
|
__DEVICE__ double sinh(double);
|
|
__DEVICE__ float sinh(float);
|
|
__DEVICE__ double sqrt(double);
|
|
__DEVICE__ float sqrt(float);
|
|
__DEVICE__ double tan(double);
|
|
__DEVICE__ float tan(float);
|
|
__DEVICE__ double tanh(double);
|
|
__DEVICE__ float tanh(float);
|
|
__DEVICE__ double tgamma(double);
|
|
__DEVICE__ float tgamma(float);
|
|
__DEVICE__ double trunc(double);
|
|
__DEVICE__ float trunc(float);
|
|
|
|
// Notably missing above is nexttoward, which we don't define on
|
|
// the device side because libdevice doesn't give us an implementation, and we
|
|
// don't want to be in the business of writing one ourselves.
|
|
|
|
// 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;
|
|
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 ::isinf;
|
|
using ::isless;
|
|
using ::islessequal;
|
|
using ::islessgreater;
|
|
using ::isnan;
|
|
using ::isnormal;
|
|
using ::isunordered;
|
|
using ::labs;
|
|
using ::ldexp;
|
|
using ::lgamma;
|
|
using ::llabs;
|
|
using ::llrint;
|
|
using ::log;
|
|
using ::log10;
|
|
using ::log1p;
|
|
using ::log2;
|
|
using ::logb;
|
|
using ::lrint;
|
|
using ::lround;
|
|
using ::llround;
|
|
using ::modf;
|
|
using ::nan;
|
|
using ::nanf;
|
|
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;
|
|
|
|
#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
|
|
#pragma pop_macro("__DEVICE__")
|
|
|
|
#endif
|