Eigen::half_impl Namespace Reference

Classes
struct	__half_raw
union	float32_bits
struct	half_base

Functions
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC __half_raw	raw_uint16_to_half (unsigned short x)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC __half_raw	float_to_half_rtne (float ff)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC float	half_to_float (__half_raw h)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	operator+ (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	operator* (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	operator- (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	operator/ (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	operator- (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half &	operator+= (half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half &	operator*= (half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half &	operator-= (half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half &	operator/= (half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool	operator== (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool	operator!= (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool	operator< (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool	operator<= (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool	operator> (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool	operator>= (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	operator/ (const half &a, Index b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool()	isinf (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool()	isnan (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool()	isfinite (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	abs (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	exp (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	log (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	log1p (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	log10 (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	sqrt (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	pow (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	sin (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	cos (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	tan (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	tanh (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	floor (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half	ceil (const half &a)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half()	min (const half &a, const half &b)
EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half()	max (const half &a, const half &b)
EIGEN_ALWAYS_INLINE std::ostream &	operator<< (std::ostream &os, const half &v)

---

Class Documentation

Eigen::half_impl::float32_bits

union Eigen::half_impl::float32_bits

Class Members
float	f
unsigned int	u

Function Documentation

abs()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::abs

(

const half &

a

)

                                                              {
  half result;
  result.x = a.x & 0x7FFF;
  return result;
}

References abs(), and Eigen::half_impl::__half_raw::x.

Referenced by abs().

Here is the call graph for this function:

Here is the caller graph for this function:

ceil()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::ceil

(

const half &

a

)

                                                               {
#if EIGEN_CUDACC_VER >= 80000 && defined EIGEN_CUDA_ARCH && EIGEN_CUDA_ARCH >= 300
  return half(hceil(a));
#else
  return half(::ceilf(float(a)));
#endif
}

cos()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::cos

(

const half &

a

)

                                                              {
  return half(::cosf(float(a)));
}

exp()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::exp

(

const half &

a

)

                                                              {
#if EIGEN_CUDACC_VER >= 80000 && defined EIGEN_CUDA_ARCH && EIGEN_CUDA_ARCH >= 530
  return half(hexp(a));
#else
   return half(::expf(float(a)));
#endif
}

float_to_half_rtne()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC __half_raw Eigen::half_impl::float_to_half_rtne

(

float

ff

)

                                                                              {
#if defined(EIGEN_HAS_CUDA_FP16) && defined(EIGEN_CUDA_ARCH) && EIGEN_CUDA_ARCH >= 300
  __half tmp_ff = __float2half(ff);
  return *(__half_raw*)&tmp_ff;
 
#elif defined(EIGEN_HAS_FP16_C)
  __half_raw h;
  h.x = _cvtss_sh(ff, 0);
  return h;
 
#else
  float32_bits f; f.f = ff;
 
  const float32_bits f32infty = { 255 << 23 };
  const float32_bits f16max = { (127 + 16) << 23 };
  const float32_bits denorm_magic = { ((127 - 15) + (23 - 10) + 1) << 23 };
  unsigned int sign_mask = 0x80000000u;
  __half_raw o;
  o.x = static_cast<unsigned short>(0x0u);
 
  unsigned int sign = f.u & sign_mask;
  f.u ^= sign;
 
  // NOTE all the integer compares in this function can be safely
  // compiled into signed compares since all operands are below
  // 0x80000000. Important if you want fast straight SSE2 code
  // (since there's no unsigned PCMPGTD).
 
  if (f.u >= f16max.u) {  // result is Inf or NaN (all exponent bits set)
    o.x = (f.u > f32infty.u) ? 0x7e00 : 0x7c00; // NaN->qNaN and Inf->Inf
  } else {  // (De)normalized number or zero
    if (f.u < (113 << 23)) {  // resulting FP16 is subnormal or zero
      // use a magic value to align our 10 mantissa bits at the bottom of
      // the float. as long as FP addition is round-to-nearest-even this
      // just works.
      f.f += denorm_magic.f;
 
      // and one integer subtract of the bias later, we have our final float!
      o.x = static_cast<unsigned short>(f.u - denorm_magic.u);
    } else {
      unsigned int mant_odd = (f.u >> 13) & 1; // resulting mantissa is odd
 
      // update exponent, rounding bias part 1
      f.u += ((unsigned int)(15 - 127) << 23) + 0xfff;
      // rounding bias part 2
      f.u += mant_odd;
      // take the bits!
      o.x = static_cast<unsigned short>(f.u >> 13);
    }
  }
 
  o.x |= static_cast<unsigned short>(sign >> 16);
  return o;
#endif
}

References Eigen::half_impl::float32_bits::f, sign(), Eigen::half_impl::float32_bits::u, and Eigen::half_impl::__half_raw::x.

Here is the call graph for this function:

floor()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::floor

(

const half &

a

)

                                                                {
#if EIGEN_CUDACC_VER >= 80000 && defined EIGEN_CUDA_ARCH && EIGEN_CUDA_ARCH >= 300
  return half(hfloor(a));
#else
  return half(::floorf(float(a)));
#endif
}

half_to_float()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC float Eigen::half_impl::half_to_float

(

__half_raw

h

)

                                                                        {
#if defined(EIGEN_HAS_CUDA_FP16) && defined(EIGEN_CUDA_ARCH) && EIGEN_CUDA_ARCH >= 300
  return __half2float(h);
 
#elif defined(EIGEN_HAS_FP16_C)
  return _cvtsh_ss(h.x);
 
#else
  const float32_bits magic = { 113 << 23 };
  const unsigned int shifted_exp = 0x7c00 << 13; // exponent mask after shift
  float32_bits o;
 
  o.u = (h.x & 0x7fff) << 13;             // exponent/mantissa bits
  unsigned int exp = shifted_exp & o.u;   // just the exponent
  o.u += (127 - 15) << 23;                // exponent adjust
 
  // handle exponent special cases
  if (exp == shifted_exp) {     // Inf/NaN?
    o.u += (128 - 16) << 23;    // extra exp adjust
  } else if (exp == 0) {        // Zero/Denormal?
    o.u += 1 << 23;             // extra exp adjust
    o.f -= magic.f;             // renormalize
  }
 
  o.u |= (h.x & 0x8000) << 16;    // sign bit
  return o.f;
#endif
}

References exp(), Eigen::half_impl::float32_bits::f, Eigen::half_impl::float32_bits::u, and Eigen::half_impl::__half_raw::x.

Referenced by Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), Eigen::half::EIGEN_EXPLICIT_CAST(), and Eigen::half::EIGEN_EXPLICIT_CAST().

Here is the call graph for this function:

Here is the caller graph for this function:

isfinite()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool() Eigen::half_impl::isfinite

(

const half &

a

)

                                                                     {
  return !(isinf EIGEN_NOT_A_MACRO (a)) && !(isnan EIGEN_NOT_A_MACRO (a));
}

References EIGEN_NOT_A_MACRO, isinf(), and isnan().

Here is the call graph for this function:

isinf()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool() Eigen::half_impl::isinf

(

const half &

a

)

                                                                  {
  return (a.x & 0x7fff) == 0x7c00;
}

Referenced by isfinite().

Here is the caller graph for this function:

isnan()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool() Eigen::half_impl::isnan

(

const half &

a

)

                                                                  {
#if defined(EIGEN_HAS_CUDA_FP16) && defined(EIGEN_CUDA_ARCH) && EIGEN_CUDA_ARCH >= 530
  return __hisnan(a);
#else
  return (a.x & 0x7fff) > 0x7c00;
#endif
}

Referenced by isfinite().

Here is the caller graph for this function:

log()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::log

(

const half &

a

)

                                                              {
#if defined(EIGEN_HAS_CUDA_FP16) && EIGEN_CUDACC_VER >= 80000 && defined(EIGEN_CUDA_ARCH) && EIGEN_CUDA_ARCH >= 530
  return half(::hlog(a));
#else
  return half(::logf(float(a)));
#endif
}

log10()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::log10

(

const half &

a

)

                                                                {
  return half(::log10f(float(a)));
}

log1p()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::log1p

(

const half &

a

)

                                                                {
  return half(numext::log1p(float(a)));
}

max()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half() Eigen::half_impl::max	(	const half &	a,
		const half &	b
	)

                                                                               {
#if defined(EIGEN_HAS_CUDA_FP16) && defined(EIGEN_CUDA_ARCH) && EIGEN_CUDA_ARCH >= 530
  return __hlt(a, b) ? b : a;
#else
  const float f1 = static_cast<float>(a);
  const float f2 = static_cast<float>(b);
  return f1 < f2 ? b : a;
#endif
}

min()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half() Eigen::half_impl::min	(	const half &	a,
		const half &	b
	)

                                                                               {
#if defined(EIGEN_HAS_CUDA_FP16) && defined(EIGEN_CUDA_ARCH) && EIGEN_CUDA_ARCH >= 530
  return __hlt(b, a) ? b : a;
#else
  const float f1 = static_cast<float>(a);
  const float f2 = static_cast<float>(b);
  return f2 < f1 ? b : a;
#endif
}

operator!=()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool Eigen::half_impl::operator!=	(	const half &	a,
		const half &	b
	)

                                                                                      {
  return numext::not_equal_strict(float(a), float(b));
}

References Eigen::numext::not_equal_strict().

Here is the call graph for this function:

operator*()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::operator*	(	const half &	a,
		const half &	b
	)

                                                                                     {
  return half(float(a) * float(b));
}

operator*=()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half & Eigen::half_impl::operator*=	(	half &	a,
		const half &	b
	)

                                                                                 {
  a = half(float(a) * float(b));
  return a;
}

operator+()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::operator+	(	const half &	a,
		const half &	b
	)

                                                                                     {
  return half(float(a) + float(b));
}

operator+=()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half & Eigen::half_impl::operator+=	(	half &	a,
		const half &	b
	)

                                                                                 {
  a = half(float(a) + float(b));
  return a;
}

operator-() [1/2]

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::operator-

(

const half &

a

)

                                                                      {
  half result;
  result.x = a.x ^ 0x8000;
  return result;
}

References Eigen::half_impl::__half_raw::x.

operator-() [2/2]

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::operator-	(	const half &	a,
		const half &	b
	)

                                                                                     {
  return half(float(a) - float(b));
}

operator-=()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half & Eigen::half_impl::operator-=	(	half &	a,
		const half &	b
	)

                                                                                 {
  a = half(float(a) - float(b));
  return a;
}

operator/() [1/2]

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::operator/	(	const half &	a,
		const half &	b
	)

                                                                                     {
  return half(float(a) / float(b));
}

operator/() [2/2]

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::operator/	(	const half &	a,
		Index	b
	)

                                                                               {
  return half(static_cast<float>(a) / static_cast<float>(b));
}

operator/=()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half & Eigen::half_impl::operator/=	(	half &	a,
		const half &	b
	)

                                                                                 {
  a = half(float(a) / float(b));
  return a;
}

operator<()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool Eigen::half_impl::operator<	(	const half &	a,
		const half &	b
	)

                                                                                     {
  return float(a) < float(b);
}

operator<<()

EIGEN_ALWAYS_INLINE std::ostream & Eigen::half_impl::operator<<	(	std::ostream &	os,
		const half &	v
	)

                                                                            {
  os << static_cast<float>(v);
  return os;
}

operator<=()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool Eigen::half_impl::operator<=	(	const half &	a,
		const half &	b
	)

                                                                                      {
  return float(a) <= float(b);
}

operator==()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool Eigen::half_impl::operator==	(	const half &	a,
		const half &	b
	)

                                                                                      {
  return numext::equal_strict(float(a),float(b));
}

References Eigen::numext::equal_strict().

Here is the call graph for this function:

operator>()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool Eigen::half_impl::operator>	(	const half &	a,
		const half &	b
	)

                                                                                     {
  return float(a) > float(b);
}

operator>=()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC bool Eigen::half_impl::operator>=	(	const half &	a,
		const half &	b
	)

                                                                                      {
  return float(a) >= float(b);
}

pow()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::pow	(	const half &	a,
		const half &	b
	)

                                                                             {
  return half(::powf(float(a), float(b)));
}

raw_uint16_to_half()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC __half_raw Eigen::half_impl::raw_uint16_to_half

(

unsigned short

x

)

                                                                                      {
  __half_raw h;
  h.x = x;
  return h;
}

References Eigen::half_impl::__half_raw::x.

Referenced by std::numeric_limits< Eigen::half >::denorm_min(), std::numeric_limits< Eigen::half >::epsilon(), Eigen::NumTraits< Eigen::half >::epsilon(), Eigen::NumTraits< Eigen::half >::highest(), std::numeric_limits< Eigen::half >::infinity(), Eigen::NumTraits< Eigen::half >::infinity(), std::numeric_limits< Eigen::half >::lowest(), Eigen::NumTraits< Eigen::half >::lowest(), std::numeric_limits< Eigen::half >::max(), std::numeric_limits< Eigen::half >::min(), std::numeric_limits< Eigen::half >::quiet_NaN(), Eigen::NumTraits< Eigen::half >::quiet_NaN(), and std::numeric_limits< Eigen::half >::signaling_NaN().

Here is the caller graph for this function:

sin()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::sin

(

const half &

a

)

                                                              {
  return half(::sinf(float(a)));
}

sqrt()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::sqrt

(

const half &

a

)

                                                               {
#if EIGEN_CUDACC_VER >= 80000 && defined EIGEN_CUDA_ARCH && EIGEN_CUDA_ARCH >= 530
  return half(hsqrt(a));
#else
    return half(::sqrtf(float(a)));
#endif
}

tan()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::tan

(

const half &

a

)

                                                              {
  return half(::tanf(float(a)));
}

tanh()

EIGEN_STRONG_INLINE EIGEN_DEVICE_FUNC half Eigen::half_impl::tanh

(

const half &

a

)

                                                               {
  return half(::tanhf(float(a)));
}