GitHub-Proxy/eigen
1
0
Fork 0
mirror of https://gitlab.com/libeigen/eigen.git synced 2025-08-15 13:15:57 +08:00
Code Issues Packages Projects Releases Wiki Activity

Added more conversion operators for half floats

Browse Source
This commit is contained in:
Benoit Steiner 2016-03-23 09:44:52 -07:00
parent 9642fd7a93
commit 6971146ca9
1 changed files with 48 additions and 7 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

55
Eigen/src/Core/arch/CUDA/Half.h
View File

@ -65,20 +65,61 @@ static inline EIGEN_DEVICE_FUNC float half_to_float(__half h);
struct half : public __half { struct half : public __half {
EIGEN_DEVICE_FUNC half() : __half(internal::raw_uint16_to_half(0)) {} EIGEN_DEVICE_FUNC half() : __half(internal::raw_uint16_to_half(0)) {}
// TODO(sesse): Should these conversions be marked as explicit?
EIGEN_DEVICE_FUNC half(float f) : __half(internal::float_to_half_rtne(f)) {}
EIGEN_DEVICE_FUNC half(int i) : __half(internal::float_to_half_rtne(static_cast<float>(i))) {}
EIGEN_DEVICE_FUNC half(double d) : __half(internal::float_to_half_rtne(static_cast<float>(d))) {}
EIGEN_DEVICE_FUNC half(bool b)
: __half(internal::raw_uint16_to_half(b ? 0x3c00 : 0)) {}
EIGEN_DEVICE_FUNC half(const __half& h) : __half(h) {} EIGEN_DEVICE_FUNC half(const __half& h) : __half(h) {}
EIGEN_DEVICE_FUNC half(const half& h) : __half(h) {} EIGEN_DEVICE_FUNC half(const half& h) : __half(h) {}
explicit EIGEN_DEVICE_FUNC half(bool b)
: __half(internal::raw_uint16_to_half(b ? 0x3c00 : 0)) {}
explicit EIGEN_DEVICE_FUNC half(int i)
: __half(internal::float_to_half_rtne(static_cast<float>(i))) {}
explicit EIGEN_DEVICE_FUNC half(long l)
: __half(internal::float_to_half_rtne(static_cast<float>(l))) {}
explicit EIGEN_DEVICE_FUNC half(long long ll)
: __half(internal::float_to_half_rtne(static_cast<float>(ll))) {}
explicit EIGEN_DEVICE_FUNC half(float f)
: __half(internal::float_to_half_rtne(f)) {}
explicit EIGEN_DEVICE_FUNC half(double d)
: __half(internal::float_to_half_rtne(static_cast<float>(d))) {}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(bool) const {
// +0.0 and -0.0 become false, everything else becomes true.
return static_cast<bool>(x & 0x7fff);
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(signed char) const {
return static_cast<signed char>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(unsigned char) const {
return static_cast<unsigned char>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(short) const {
return static_cast<short>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(unsigned short) const {
return static_cast<unsigned short>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(int) const {
return static_cast<int>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(unsigned int) const {
return static_cast<unsigned int>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(long) const {
return static_cast<long>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(unsigned long) const {
return static_cast<unsigned long>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(long long) const {
return static_cast<long long>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(unsigned long long) const {
return static_cast<unsigned long long>(internal::half_to_float(*this));
}
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(float) const { EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(float) const {
return internal::half_to_float(*this); return internal::half_to_float(*this);
} }
EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(double) const { EIGEN_DEVICE_FUNC EIGEN_EXPLICIT_CAST(double) const {
return internal::half_to_float(*this); return static_cast<double>(internal::half_to_float(*this));
} }
EIGEN_DEVICE_FUNC half& operator=(const half& other) { EIGEN_DEVICE_FUNC half& operator=(const half& other) {