added host for float2half and hlaf2float

このコミットが含まれているのは:
lcskrishna
2018-09-06 15:02:52 -07:00
コミット fc2c8de816
+13 -13
ファイルの表示
@@ -635,37 +635,37 @@ THE SOFTWARE.
// TODO: rounding behaviour is not correct. // TODO: rounding behaviour is not correct.
// float -> half | half2 // float -> half | half2
inline inline
__device__ __device__ __host__
__half __float2half(float x) __half __float2half(float x)
{ {
return __half_raw{static_cast<_Float16>(x)}; return __half_raw{static_cast<_Float16>(x)};
} }
inline inline
__device__ __device__ __host__
__half __float2half_rn(float x) __half __float2half_rn(float x)
{ {
return __half_raw{static_cast<_Float16>(x)}; return __half_raw{static_cast<_Float16>(x)};
} }
inline inline
__device__ __device__ __host__
__half __float2half_rz(float x) __half __float2half_rz(float x)
{ {
return __half_raw{static_cast<_Float16>(x)}; return __half_raw{static_cast<_Float16>(x)};
} }
inline inline
__device__ __device__ __host__
__half __float2half_rd(float x) __half __float2half_rd(float x)
{ {
return __half_raw{static_cast<_Float16>(x)}; return __half_raw{static_cast<_Float16>(x)};
} }
inline inline
__device__ __device__ __host__
__half __float2half_ru(float x) __half __float2half_ru(float x)
{ {
return __half_raw{static_cast<_Float16>(x)}; return __half_raw{static_cast<_Float16>(x)};
} }
inline inline
__device__ __device__ __host__
__half2 __float2half2_rn(float x) __half2 __float2half2_rn(float x)
{ {
return __half2_raw{ return __half2_raw{
@@ -673,14 +673,14 @@ THE SOFTWARE.
static_cast<_Float16>(x), static_cast<_Float16>(x)}}; static_cast<_Float16>(x), static_cast<_Float16>(x)}};
} }
inline inline
__device__ __device__ __host__
__half2 __floats2half2_rn(float x, float y) __half2 __floats2half2_rn(float x, float y)
{ {
return __half2_raw{_Float16_2{ return __half2_raw{_Float16_2{
static_cast<_Float16>(x), static_cast<_Float16>(y)}}; static_cast<_Float16>(x), static_cast<_Float16>(y)}};
} }
inline inline
__device__ __device__ __host__
__half2 __float22half2_rn(float2 x) __half2 __float22half2_rn(float2 x)
{ {
return __floats2half2_rn(x.x, x.y); return __floats2half2_rn(x.x, x.y);
@@ -688,25 +688,25 @@ THE SOFTWARE.
// half | half2 -> float // half | half2 -> float
inline inline
__device__ __device__ __host__
float __half2float(__half x) float __half2float(__half x)
{ {
return static_cast<__half_raw>(x).data; return static_cast<__half_raw>(x).data;
} }
inline inline
__device__ __device__ __host__
float __low2float(__half2 x) float __low2float(__half2 x)
{ {
return static_cast<__half2_raw>(x).data.x; return static_cast<__half2_raw>(x).data.x;
} }
inline inline
__device__ __device__ __host__
float __high2float(__half2 x) float __high2float(__half2 x)
{ {
return static_cast<__half2_raw>(x).data.y; return static_cast<__half2_raw>(x).data.y;
} }
inline inline
__device__ __device__ __host__
float2 __half22float2(__half2 x) float2 __half22float2(__half2 x)
{ {
return make_float2( return make_float2(
@@ -1633,4 +1633,4 @@ THE SOFTWARE.
#endif // defined(__cplusplus) #endif // defined(__cplusplus)
#elif defined(__GNUC__) #elif defined(__GNUC__)
#include "hip_fp16_gcc.h" #include "hip_fp16_gcc.h"
#endif // !defined(__clang__) && defined(__GNUC__) #endif // !defined(__clang__) && defined(__GNUC__)