31#ifndef HIP_INCLUDE_HIP_AMD_DETAIL_DEVICE_LIBRARY_DECLS_H
32#define HIP_INCLUDE_HIP_AMD_DETAIL_DEVICE_LIBRARY_DECLS_H
36typedef unsigned char uchar;
37typedef unsigned short ushort;
38typedef unsigned int uint;
39typedef unsigned long ulong;
40typedef unsigned long long ullong;
42extern "C" __device__ __attribute__((
const)) bool __ockl_wfany_i32(
int);
43extern "C" __device__ __attribute__((const))
bool __ockl_wfall_i32(
int);
44extern "C" __device__ uint __ockl_activelane_u32(
void);
46extern "C" __device__ __attribute__((const)) uint __ockl_mul24_u32(uint, uint);
47extern "C" __device__ __attribute__((const))
int __ockl_mul24_i32(
int,
int);
48extern "C" __device__ __attribute__((const)) uint __ockl_mul_hi_u32(uint, uint);
49extern "C" __device__ __attribute__((const))
int __ockl_mul_hi_i32(
int,
int);
50extern "C" __device__ __attribute__((const)) uint __ockl_sadd_u32(uint, uint, uint);
52extern "C" __device__ __attribute__((const)) uchar __ockl_clz_u8(uchar);
53extern "C" __device__ __attribute__((const)) ushort __ockl_clz_u16(ushort);
54extern "C" __device__ __attribute__((const)) uint __ockl_clz_u32(uint);
55extern "C" __device__ __attribute__((const)) uint64_t __ockl_clz_u64(uint64_t);
57extern "C" __device__ __attribute__((const))
float __ocml_floor_f32(
float);
58extern "C" __device__ __attribute__((const))
float __ocml_rint_f32(
float);
59extern "C" __device__ __attribute__((const))
float __ocml_ceil_f32(
float);
60extern "C" __device__ __attribute__((const))
float __ocml_trunc_f32(
float);
62extern "C" __device__ __attribute__((const))
float __ocml_fmin_f32(
float,
float);
63extern "C" __device__ __attribute__((const))
float __ocml_fmax_f32(
float,
float);
65extern "C" __device__ __attribute__((const))
float __ocml_cvtrtn_f32_f64(
double);
66extern "C" __device__ __attribute__((const))
float __ocml_cvtrtp_f32_f64(
double);
67extern "C" __device__ __attribute__((const))
float __ocml_cvtrtz_f32_f64(
double);
69extern "C" __device__ __attribute__((const)) _Float16 __ocml_cvtrtn_f16_f32(
float);
70extern "C" __device__ __attribute__((const)) _Float16 __ocml_cvtrtp_f16_f32(
float);
71extern "C" __device__ __attribute__((const)) _Float16 __ocml_cvtrtz_f16_f32(
float);
73extern "C" __device__ __attribute__((const))
float __ocml_cvtrtn_f32_s32(
int);
74extern "C" __device__ __attribute__((const))
float __ocml_cvtrtp_f32_s32(
int);
75extern "C" __device__ __attribute__((const))
float __ocml_cvtrtz_f32_s32(
int);
76extern "C" __device__ __attribute__((const))
float __ocml_cvtrtn_f32_u32(uint32_t);
77extern "C" __device__ __attribute__((const))
float __ocml_cvtrtp_f32_u32(uint32_t);
78extern "C" __device__ __attribute__((const))
float __ocml_cvtrtz_f32_u32(uint32_t);
79extern "C" __device__ __attribute__((const))
float __ocml_cvtrtn_f32_s64(int64_t);
80extern "C" __device__ __attribute__((const))
float __ocml_cvtrtp_f32_s64(int64_t);
81extern "C" __device__ __attribute__((const))
float __ocml_cvtrtz_f32_s64(int64_t);
82extern "C" __device__ __attribute__((const))
float __ocml_cvtrtn_f32_u64(uint64_t);
83extern "C" __device__ __attribute__((const))
float __ocml_cvtrtp_f32_u64(uint64_t);
84extern "C" __device__ __attribute__((const))
float __ocml_cvtrtz_f32_u64(uint64_t);
85extern "C" __device__ __attribute__((const))
double __ocml_cvtrtn_f64_s64(int64_t);
86extern "C" __device__ __attribute__((const))
double __ocml_cvtrtp_f64_s64(int64_t);
87extern "C" __device__ __attribute__((const))
double __ocml_cvtrtz_f64_s64(int64_t);
88extern "C" __device__ __attribute__((const))
double __ocml_cvtrtn_f64_u64(uint64_t);
89extern "C" __device__ __attribute__((const))
double __ocml_cvtrtp_f64_u64(uint64_t);
90extern "C" __device__ __attribute__((const))
double __ocml_cvtrtz_f64_u64(uint64_t);
92extern "C" __device__ __attribute__((convergent))
void __ockl_gws_init(uint nwm1, uint rid);
93extern "C" __device__ __attribute__((convergent))
void __ockl_gws_barrier(uint nwm1, uint rid);
95extern "C" __device__ __attribute__((const)) uint32_t __ockl_lane_u32();
96extern "C" __device__ __attribute__((const))
int __ockl_grid_is_valid(
void);
97extern "C" __device__ __attribute__((convergent))
void __ockl_grid_sync(
void);
98extern "C" __device__ __attribute__((const)) uint __ockl_multi_grid_num_grids(
void);
99extern "C" __device__ __attribute__((const)) uint __ockl_multi_grid_grid_rank(
void);
100extern "C" __device__ __attribute__((const)) uint __ockl_multi_grid_size(
void);
101extern "C" __device__ __attribute__((const)) uint __ockl_multi_grid_thread_rank(
void);
102extern "C" __device__ __attribute__((const))
int __ockl_multi_grid_is_valid(
void);
103extern "C" __device__ __attribute__((convergent))
void __ockl_multi_grid_sync(
void);
105extern "C" __device__
void __ockl_atomic_add_noret_f32(
float*,
float);
107extern "C" __device__ __attribute__((convergent))
int __ockl_wgred_add_i32(
int a);
108extern "C" __device__ __attribute__((convergent))
int __ockl_wgred_and_i32(
int a);
109extern "C" __device__ __attribute__((convergent))
int __ockl_wgred_or_i32(
int a);
111extern "C" __device__ uint64_t __ockl_fprintf_stderr_begin();
112extern "C" __device__ uint64_t __ockl_fprintf_append_args(uint64_t msg_desc, uint32_t num_args,
113 uint64_t value0, uint64_t value1,
114 uint64_t value2, uint64_t value3,
115 uint64_t value4, uint64_t value5,
116 uint64_t value6, uint32_t is_last);
117extern "C" __device__ uint64_t __ockl_fprintf_append_string_n(uint64_t msg_desc, const
char* data,
118 uint64_t length, uint32_t is_last);
121#define __local __attribute__((address_space(3)))
123#ifdef __HIP_DEVICE_COMPILE__
124__device__
inline static __local
void* __to_local(
unsigned x) {
return (__local
void*)x; }
128#define __CLK_LOCAL_MEM_FENCE 0x01
129typedef unsigned __cl_mem_fence_flags;