AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

Add __device__ to device functions in hip_fp16_math_fwd.h

Browse Source
This commit is contained in:
Yaxun (Sam) Liu
2018-06-08 07:44:02 -04:00
parent cb642f14ab
commit 9566616dd2
1 changed files with 36 additions and 36 deletions
Download Patch File Download Diff File Expand all files Collapse all files
hipamd/include/hip/hcc_detail/hip_fp16_math_fwd.h
+36 -36
View File
@@ -28,49 +28,49 @@ THE SOFTWARE.
extern "C" extern "C"
{ {
__attribute__((const)) _Float16 __ocml_ceil_f16(_Float16); __device__ __attribute__((const)) _Float16 __ocml_ceil_f16(_Float16);
_Float16 __ocml_cos_f16(_Float16); __device__ _Float16 __ocml_cos_f16(_Float16);
__attribute__((pure)) _Float16 __ocml_exp_f16(_Float16); __device__ __attribute__((pure)) _Float16 __ocml_exp_f16(_Float16);
__attribute__((pure)) _Float16 __ocml_exp10_f16(_Float16); __device__ __attribute__((pure)) _Float16 __ocml_exp10_f16(_Float16);
__attribute__((pure)) _Float16 __ocml_exp2_f16(_Float16); __device__ __attribute__((pure)) _Float16 __ocml_exp2_f16(_Float16);
__attribute__((const)) _Float16 __ocml_floor_f16(_Float16); __device__ __attribute__((const)) _Float16 __ocml_floor_f16(_Float16);
__attribute__((const)) __device__ __attribute__((const))
_Float16 __ocml_fma_f16(_Float16, _Float16, _Float16); _Float16 __ocml_fma_f16(_Float16, _Float16, _Float16);
__attribute__((const)) int __ocml_isinf_f16(_Float16); __device__ __attribute__((const)) int __ocml_isinf_f16(_Float16);
__attribute__((const)) int __ocml_isnan_f16(_Float16); __device__ __attribute__((const)) int __ocml_isnan_f16(_Float16);
__attribute__((pure)) _Float16 __ocml_log_f16(_Float16); __device__ __attribute__((pure)) _Float16 __ocml_log_f16(_Float16);
__attribute__((pure)) _Float16 __ocml_log10_f16(_Float16); __device__ __attribute__((pure)) _Float16 __ocml_log10_f16(_Float16);
__attribute__((pure)) _Float16 __ocml_log2_f16(_Float16); __device__ __attribute__((pure)) _Float16 __ocml_log2_f16(_Float16);
__attribute__((const)) _Float16 __llvm_amdgcn_rcp_f16(_Float16); __device__ __attribute__((const)) _Float16 __llvm_amdgcn_rcp_f16(_Float16);
__attribute__((const)) _Float16 __ocml_rint_f16(_Float16); __device__ __attribute__((const)) _Float16 __ocml_rint_f16(_Float16);
__attribute__((const)) _Float16 __ocml_rsqrt_f16(_Float16); __device__ __attribute__((const)) _Float16 __ocml_rsqrt_f16(_Float16);
_Float16 __ocml_sin_f16(_Float16); __device__ _Float16 __ocml_sin_f16(_Float16);
__attribute__((const)) _Float16 __ocml_sqrt_f16(_Float16); __device__ __attribute__((const)) _Float16 __ocml_sqrt_f16(_Float16);
__attribute__((const)) _Float16 __ocml_trunc_f16(_Float16); __device__ __attribute__((const)) _Float16 __ocml_trunc_f16(_Float16);
typedef _Float16 __2f16 __attribute__((ext_vector_type(2))); typedef _Float16 __2f16 __attribute__((ext_vector_type(2)));
typedef short __2i16 __attribute__((ext_vector_type(2))); typedef short __2i16 __attribute__((ext_vector_type(2)));
__attribute__((const)) __2f16 __ocml_ceil_2f16(__2f16); __device__ __attribute__((const)) __2f16 __ocml_ceil_2f16(__2f16);
__2f16 __ocml_cos_2f16(__2f16); __device__ __2f16 __ocml_cos_2f16(__2f16);
__attribute__((pure)) __2f16 __ocml_exp_2f16(__2f16); __device__ __attribute__((pure)) __2f16 __ocml_exp_2f16(__2f16);
__attribute__((pure)) __2f16 __ocml_exp10_2f16(__2f16); __device__ __attribute__((pure)) __2f16 __ocml_exp10_2f16(__2f16);
__attribute__((pure)) __2f16 __ocml_exp2_2f16(__2f16); __device__ __attribute__((pure)) __2f16 __ocml_exp2_2f16(__2f16);
__attribute__((const)) __2f16 __ocml_floor_2f16(__2f16); __device__ __attribute__((const)) __2f16 __ocml_floor_2f16(__2f16);
__attribute__((const)) __2f16 __ocml_fma_2f16(__2f16, __2f16, __2f16); __device__ __attribute__((const)) __2f16 __ocml_fma_2f16(__2f16, __2f16, __2f16);
__attribute__((const)) __2i16 __ocml_isinf_2f16(__2f16); __device__ __attribute__((const)) __2i16 __ocml_isinf_2f16(__2f16);
__attribute__((const)) __2i16 __ocml_isnan_2f16(__2f16); __device__ __attribute__((const)) __2i16 __ocml_isnan_2f16(__2f16);
__attribute__((pure)) __2f16 __ocml_log_2f16(__2f16); __device__ __attribute__((pure)) __2f16 __ocml_log_2f16(__2f16);
__attribute__((pure)) __2f16 __ocml_log10_2f16(__2f16); __device__ __attribute__((pure)) __2f16 __ocml_log10_2f16(__2f16);
__attribute__((pure)) __2f16 __ocml_log2_2f16(__2f16); __device__ __attribute__((pure)) __2f16 __ocml_log2_2f16(__2f16);
inline __device__ inline
__2f16 __llvm_amdgcn_rcp_2f16(__2f16 x) // Not currently exposed by ROCDL. __2f16 __llvm_amdgcn_rcp_2f16(__2f16 x) // Not currently exposed by ROCDL.
{ {
return __2f16{__llvm_amdgcn_rcp_f16(x.x), __llvm_amdgcn_rcp_f16(x.y)}; return __2f16{__llvm_amdgcn_rcp_f16(x.x), __llvm_amdgcn_rcp_f16(x.y)};
} }
__attribute__((const)) __2f16 __ocml_rint_2f16(__2f16); __device__ __attribute__((const)) __2f16 __ocml_rint_2f16(__2f16);
__attribute__((const)) __2f16 __ocml_rsqrt_2f16(__2f16); __device__ __attribute__((const)) __2f16 __ocml_rsqrt_2f16(__2f16);
__2f16 __ocml_sin_2f16(__2f16); __device__ __2f16 __ocml_sin_2f16(__2f16);
__attribute__((const)) __2f16 __ocml_sqrt_2f16(__2f16); __device__ __attribute__((const)) __2f16 __ocml_sqrt_2f16(__2f16);
__attribute__((const)) __2f16 __ocml_trunc_2f16(__2f16); __device__ __attribute__((const)) __2f16 __ocml_trunc_2f16(__2f16);
} }