Subversion Repositories QNX 8.QNX8 LLVM/Clang compiler suite

Rev

Blame | Last modification | View Log | Download | RSS feed

  1. /*===---- __clang_hip_libdevice_declares.h - HIP device library decls -------===
  2.  *
  3.  * Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
  4.  * See https://llvm.org/LICENSE.txt for license information.
  5.  * SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
  6.  *
  7.  *===-----------------------------------------------------------------------===
  8.  */
  9.  
  10. #ifndef __CLANG_HIP_LIBDEVICE_DECLARES_H__
  11. #define __CLANG_HIP_LIBDEVICE_DECLARES_H__
  12.  
  13. #ifdef __cplusplus
  14. extern "C" {
  15. #endif
  16.  
  17. // BEGIN FLOAT
  18. __device__ __attribute__((const)) float __ocml_acos_f32(float);
  19. __device__ __attribute__((pure)) float __ocml_acosh_f32(float);
  20. __device__ __attribute__((const)) float __ocml_asin_f32(float);
  21. __device__ __attribute__((pure)) float __ocml_asinh_f32(float);
  22. __device__ __attribute__((const)) float __ocml_atan2_f32(float, float);
  23. __device__ __attribute__((const)) float __ocml_atan_f32(float);
  24. __device__ __attribute__((pure)) float __ocml_atanh_f32(float);
  25. __device__ __attribute__((pure)) float __ocml_cbrt_f32(float);
  26. __device__ __attribute__((const)) float __ocml_ceil_f32(float);
  27. __device__ __attribute__((const)) __device__ float __ocml_copysign_f32(float,
  28.                                                                        float);
  29. __device__ float __ocml_cos_f32(float);
  30. __device__ float __ocml_native_cos_f32(float);
  31. __device__ __attribute__((pure)) __device__ float __ocml_cosh_f32(float);
  32. __device__ float __ocml_cospi_f32(float);
  33. __device__ float __ocml_i0_f32(float);
  34. __device__ float __ocml_i1_f32(float);
  35. __device__ __attribute__((pure)) float __ocml_erfc_f32(float);
  36. __device__ __attribute__((pure)) float __ocml_erfcinv_f32(float);
  37. __device__ __attribute__((pure)) float __ocml_erfcx_f32(float);
  38. __device__ __attribute__((pure)) float __ocml_erf_f32(float);
  39. __device__ __attribute__((pure)) float __ocml_erfinv_f32(float);
  40. __device__ __attribute__((pure)) float __ocml_exp10_f32(float);
  41. __device__ __attribute__((pure)) float __ocml_native_exp10_f32(float);
  42. __device__ __attribute__((pure)) float __ocml_exp2_f32(float);
  43. __device__ __attribute__((pure)) float __ocml_exp_f32(float);
  44. __device__ __attribute__((pure)) float __ocml_native_exp_f32(float);
  45. __device__ __attribute__((pure)) float __ocml_expm1_f32(float);
  46. __device__ __attribute__((const)) float __ocml_fabs_f32(float);
  47. __device__ __attribute__((const)) float __ocml_fdim_f32(float, float);
  48. __device__ __attribute__((const)) float __ocml_floor_f32(float);
  49. __device__ __attribute__((const)) float __ocml_fma_f32(float, float, float);
  50. __device__ __attribute__((const)) float __ocml_fmax_f32(float, float);
  51. __device__ __attribute__((const)) float __ocml_fmin_f32(float, float);
  52. __device__ __attribute__((const)) __device__ float __ocml_fmod_f32(float,
  53.                                                                    float);
  54. __device__ float __ocml_frexp_f32(float,
  55.                                   __attribute__((address_space(5))) int *);
  56. __device__ __attribute__((const)) float __ocml_hypot_f32(float, float);
  57. __device__ __attribute__((const)) int __ocml_ilogb_f32(float);
  58. __device__ __attribute__((const)) int __ocml_isfinite_f32(float);
  59. __device__ __attribute__((const)) int __ocml_isinf_f32(float);
  60. __device__ __attribute__((const)) int __ocml_isnan_f32(float);
  61. __device__ float __ocml_j0_f32(float);
  62. __device__ float __ocml_j1_f32(float);
  63. __device__ __attribute__((const)) float __ocml_ldexp_f32(float, int);
  64. __device__ float __ocml_lgamma_f32(float);
  65. __device__ __attribute__((pure)) float __ocml_log10_f32(float);
  66. __device__ __attribute__((pure)) float __ocml_native_log10_f32(float);
  67. __device__ __attribute__((pure)) float __ocml_log1p_f32(float);
  68. __device__ __attribute__((pure)) float __ocml_log2_f32(float);
  69. __device__ __attribute__((pure)) float __ocml_native_log2_f32(float);
  70. __device__ __attribute__((const)) float __ocml_logb_f32(float);
  71. __device__ __attribute__((pure)) float __ocml_log_f32(float);
  72. __device__ __attribute__((pure)) float __ocml_native_log_f32(float);
  73. __device__ float __ocml_modf_f32(float,
  74.                                  __attribute__((address_space(5))) float *);
  75. __device__ __attribute__((const)) float __ocml_nearbyint_f32(float);
  76. __device__ __attribute__((const)) float __ocml_nextafter_f32(float, float);
  77. __device__ __attribute__((const)) float __ocml_len3_f32(float, float, float);
  78. __device__ __attribute__((const)) float __ocml_len4_f32(float, float, float,
  79.                                                         float);
  80. __device__ __attribute__((pure)) float __ocml_ncdf_f32(float);
  81. __device__ __attribute__((pure)) float __ocml_ncdfinv_f32(float);
  82. __device__ __attribute__((pure)) float __ocml_pow_f32(float, float);
  83. __device__ __attribute__((pure)) float __ocml_pown_f32(float, int);
  84. __device__ __attribute__((pure)) float __ocml_rcbrt_f32(float);
  85. __device__ __attribute__((const)) float __ocml_remainder_f32(float, float);
  86. __device__ float __ocml_remquo_f32(float, float,
  87.                                    __attribute__((address_space(5))) int *);
  88. __device__ __attribute__((const)) float __ocml_rhypot_f32(float, float);
  89. __device__ __attribute__((const)) float __ocml_rint_f32(float);
  90. __device__ __attribute__((const)) float __ocml_rlen3_f32(float, float, float);
  91. __device__ __attribute__((const)) float __ocml_rlen4_f32(float, float, float,
  92.                                                          float);
  93. __device__ __attribute__((const)) float __ocml_round_f32(float);
  94. __device__ __attribute__((pure)) float __ocml_rsqrt_f32(float);
  95. __device__ __attribute__((const)) float __ocml_scalb_f32(float, float);
  96. __device__ __attribute__((const)) float __ocml_scalbn_f32(float, int);
  97. __device__ __attribute__((const)) int __ocml_signbit_f32(float);
  98. __device__ float __ocml_sincos_f32(float,
  99.                                    __attribute__((address_space(5))) float *);
  100. __device__ float __ocml_sincospi_f32(float,
  101.                                      __attribute__((address_space(5))) float *);
  102. __device__ float __ocml_sin_f32(float);
  103. __device__ float __ocml_native_sin_f32(float);
  104. __device__ __attribute__((pure)) float __ocml_sinh_f32(float);
  105. __device__ float __ocml_sinpi_f32(float);
  106. __device__ __attribute__((const)) float __ocml_sqrt_f32(float);
  107. __device__ __attribute__((const)) float __ocml_native_sqrt_f32(float);
  108. __device__ float __ocml_tan_f32(float);
  109. __device__ __attribute__((pure)) float __ocml_tanh_f32(float);
  110. __device__ float __ocml_tgamma_f32(float);
  111. __device__ __attribute__((const)) float __ocml_trunc_f32(float);
  112. __device__ float __ocml_y0_f32(float);
  113. __device__ float __ocml_y1_f32(float);
  114.  
  115. // BEGIN INTRINSICS
  116. __device__ __attribute__((const)) float __ocml_add_rte_f32(float, float);
  117. __device__ __attribute__((const)) float __ocml_add_rtn_f32(float, float);
  118. __device__ __attribute__((const)) float __ocml_add_rtp_f32(float, float);
  119. __device__ __attribute__((const)) float __ocml_add_rtz_f32(float, float);
  120. __device__ __attribute__((const)) float __ocml_sub_rte_f32(float, float);
  121. __device__ __attribute__((const)) float __ocml_sub_rtn_f32(float, float);
  122. __device__ __attribute__((const)) float __ocml_sub_rtp_f32(float, float);
  123. __device__ __attribute__((const)) float __ocml_sub_rtz_f32(float, float);
  124. __device__ __attribute__((const)) float __ocml_mul_rte_f32(float, float);
  125. __device__ __attribute__((const)) float __ocml_mul_rtn_f32(float, float);
  126. __device__ __attribute__((const)) float __ocml_mul_rtp_f32(float, float);
  127. __device__ __attribute__((const)) float __ocml_mul_rtz_f32(float, float);
  128. __device__ __attribute__((const)) float __ocml_div_rte_f32(float, float);
  129. __device__ __attribute__((const)) float __ocml_div_rtn_f32(float, float);
  130. __device__ __attribute__((const)) float __ocml_div_rtp_f32(float, float);
  131. __device__ __attribute__((const)) float __ocml_div_rtz_f32(float, float);
  132. __device__ __attribute__((const)) float __ocml_sqrt_rte_f32(float);
  133. __device__ __attribute__((const)) float __ocml_sqrt_rtn_f32(float);
  134. __device__ __attribute__((const)) float __ocml_sqrt_rtp_f32(float);
  135. __device__ __attribute__((const)) float __ocml_sqrt_rtz_f32(float);
  136. __device__ __attribute__((const)) float __ocml_fma_rte_f32(float, float, float);
  137. __device__ __attribute__((const)) float __ocml_fma_rtn_f32(float, float, float);
  138. __device__ __attribute__((const)) float __ocml_fma_rtp_f32(float, float, float);
  139. __device__ __attribute__((const)) float __ocml_fma_rtz_f32(float, float, float);
  140.  
  141. __device__ inline __attribute__((const)) float
  142. __llvm_amdgcn_cos_f32(float __x) {
  143.   return __builtin_amdgcn_cosf(__x);
  144. }
  145. __device__ inline __attribute__((const)) float
  146. __llvm_amdgcn_rcp_f32(float __x) {
  147.   return __builtin_amdgcn_rcpf(__x);
  148. }
  149. __device__ inline __attribute__((const)) float
  150. __llvm_amdgcn_rsq_f32(float __x) {
  151.   return __builtin_amdgcn_rsqf(__x);
  152. }
  153. __device__ inline __attribute__((const)) float
  154. __llvm_amdgcn_sin_f32(float __x) {
  155.   return __builtin_amdgcn_sinf(__x);
  156. }
  157. // END INTRINSICS
  158. // END FLOAT
  159.  
  160. // BEGIN DOUBLE
  161. __device__ __attribute__((const)) double __ocml_acos_f64(double);
  162. __device__ __attribute__((pure)) double __ocml_acosh_f64(double);
  163. __device__ __attribute__((const)) double __ocml_asin_f64(double);
  164. __device__ __attribute__((pure)) double __ocml_asinh_f64(double);
  165. __device__ __attribute__((const)) double __ocml_atan2_f64(double, double);
  166. __device__ __attribute__((const)) double __ocml_atan_f64(double);
  167. __device__ __attribute__((pure)) double __ocml_atanh_f64(double);
  168. __device__ __attribute__((pure)) double __ocml_cbrt_f64(double);
  169. __device__ __attribute__((const)) double __ocml_ceil_f64(double);
  170. __device__ __attribute__((const)) double __ocml_copysign_f64(double, double);
  171. __device__ double __ocml_cos_f64(double);
  172. __device__ __attribute__((pure)) double __ocml_cosh_f64(double);
  173. __device__ double __ocml_cospi_f64(double);
  174. __device__ double __ocml_i0_f64(double);
  175. __device__ double __ocml_i1_f64(double);
  176. __device__ __attribute__((pure)) double __ocml_erfc_f64(double);
  177. __device__ __attribute__((pure)) double __ocml_erfcinv_f64(double);
  178. __device__ __attribute__((pure)) double __ocml_erfcx_f64(double);
  179. __device__ __attribute__((pure)) double __ocml_erf_f64(double);
  180. __device__ __attribute__((pure)) double __ocml_erfinv_f64(double);
  181. __device__ __attribute__((pure)) double __ocml_exp10_f64(double);
  182. __device__ __attribute__((pure)) double __ocml_exp2_f64(double);
  183. __device__ __attribute__((pure)) double __ocml_exp_f64(double);
  184. __device__ __attribute__((pure)) double __ocml_expm1_f64(double);
  185. __device__ __attribute__((const)) double __ocml_fabs_f64(double);
  186. __device__ __attribute__((const)) double __ocml_fdim_f64(double, double);
  187. __device__ __attribute__((const)) double __ocml_floor_f64(double);
  188. __device__ __attribute__((const)) double __ocml_fma_f64(double, double, double);
  189. __device__ __attribute__((const)) double __ocml_fmax_f64(double, double);
  190. __device__ __attribute__((const)) double __ocml_fmin_f64(double, double);
  191. __device__ __attribute__((const)) double __ocml_fmod_f64(double, double);
  192. __device__ double __ocml_frexp_f64(double,
  193.                                    __attribute__((address_space(5))) int *);
  194. __device__ __attribute__((const)) double __ocml_hypot_f64(double, double);
  195. __device__ __attribute__((const)) int __ocml_ilogb_f64(double);
  196. __device__ __attribute__((const)) int __ocml_isfinite_f64(double);
  197. __device__ __attribute__((const)) int __ocml_isinf_f64(double);
  198. __device__ __attribute__((const)) int __ocml_isnan_f64(double);
  199. __device__ double __ocml_j0_f64(double);
  200. __device__ double __ocml_j1_f64(double);
  201. __device__ __attribute__((const)) double __ocml_ldexp_f64(double, int);
  202. __device__ double __ocml_lgamma_f64(double);
  203. __device__ __attribute__((pure)) double __ocml_log10_f64(double);
  204. __device__ __attribute__((pure)) double __ocml_log1p_f64(double);
  205. __device__ __attribute__((pure)) double __ocml_log2_f64(double);
  206. __device__ __attribute__((const)) double __ocml_logb_f64(double);
  207. __device__ __attribute__((pure)) double __ocml_log_f64(double);
  208. __device__ double __ocml_modf_f64(double,
  209.                                   __attribute__((address_space(5))) double *);
  210. __device__ __attribute__((const)) double __ocml_nearbyint_f64(double);
  211. __device__ __attribute__((const)) double __ocml_nextafter_f64(double, double);
  212. __device__ __attribute__((const)) double __ocml_len3_f64(double, double,
  213.                                                          double);
  214. __device__ __attribute__((const)) double __ocml_len4_f64(double, double, double,
  215.                                                          double);
  216. __device__ __attribute__((pure)) double __ocml_ncdf_f64(double);
  217. __device__ __attribute__((pure)) double __ocml_ncdfinv_f64(double);
  218. __device__ __attribute__((pure)) double __ocml_pow_f64(double, double);
  219. __device__ __attribute__((pure)) double __ocml_pown_f64(double, int);
  220. __device__ __attribute__((pure)) double __ocml_rcbrt_f64(double);
  221. __device__ __attribute__((const)) double __ocml_remainder_f64(double, double);
  222. __device__ double __ocml_remquo_f64(double, double,
  223.                                     __attribute__((address_space(5))) int *);
  224. __device__ __attribute__((const)) double __ocml_rhypot_f64(double, double);
  225. __device__ __attribute__((const)) double __ocml_rint_f64(double);
  226. __device__ __attribute__((const)) double __ocml_rlen3_f64(double, double,
  227.                                                           double);
  228. __device__ __attribute__((const)) double __ocml_rlen4_f64(double, double,
  229.                                                           double, double);
  230. __device__ __attribute__((const)) double __ocml_round_f64(double);
  231. __device__ __attribute__((pure)) double __ocml_rsqrt_f64(double);
  232. __device__ __attribute__((const)) double __ocml_scalb_f64(double, double);
  233. __device__ __attribute__((const)) double __ocml_scalbn_f64(double, int);
  234. __device__ __attribute__((const)) int __ocml_signbit_f64(double);
  235. __device__ double __ocml_sincos_f64(double,
  236.                                     __attribute__((address_space(5))) double *);
  237. __device__ double
  238. __ocml_sincospi_f64(double, __attribute__((address_space(5))) double *);
  239. __device__ double __ocml_sin_f64(double);
  240. __device__ __attribute__((pure)) double __ocml_sinh_f64(double);
  241. __device__ double __ocml_sinpi_f64(double);
  242. __device__ __attribute__((const)) double __ocml_sqrt_f64(double);
  243. __device__ double __ocml_tan_f64(double);
  244. __device__ __attribute__((pure)) double __ocml_tanh_f64(double);
  245. __device__ double __ocml_tgamma_f64(double);
  246. __device__ __attribute__((const)) double __ocml_trunc_f64(double);
  247. __device__ double __ocml_y0_f64(double);
  248. __device__ double __ocml_y1_f64(double);
  249.  
  250. // BEGIN INTRINSICS
  251. __device__ __attribute__((const)) double __ocml_add_rte_f64(double, double);
  252. __device__ __attribute__((const)) double __ocml_add_rtn_f64(double, double);
  253. __device__ __attribute__((const)) double __ocml_add_rtp_f64(double, double);
  254. __device__ __attribute__((const)) double __ocml_add_rtz_f64(double, double);
  255. __device__ __attribute__((const)) double __ocml_sub_rte_f64(double, double);
  256. __device__ __attribute__((const)) double __ocml_sub_rtn_f64(double, double);
  257. __device__ __attribute__((const)) double __ocml_sub_rtp_f64(double, double);
  258. __device__ __attribute__((const)) double __ocml_sub_rtz_f64(double, double);
  259. __device__ __attribute__((const)) double __ocml_mul_rte_f64(double, double);
  260. __device__ __attribute__((const)) double __ocml_mul_rtn_f64(double, double);
  261. __device__ __attribute__((const)) double __ocml_mul_rtp_f64(double, double);
  262. __device__ __attribute__((const)) double __ocml_mul_rtz_f64(double, double);
  263. __device__ __attribute__((const)) double __ocml_div_rte_f64(double, double);
  264. __device__ __attribute__((const)) double __ocml_div_rtn_f64(double, double);
  265. __device__ __attribute__((const)) double __ocml_div_rtp_f64(double, double);
  266. __device__ __attribute__((const)) double __ocml_div_rtz_f64(double, double);
  267. __device__ __attribute__((const)) double __ocml_sqrt_rte_f64(double);
  268. __device__ __attribute__((const)) double __ocml_sqrt_rtn_f64(double);
  269. __device__ __attribute__((const)) double __ocml_sqrt_rtp_f64(double);
  270. __device__ __attribute__((const)) double __ocml_sqrt_rtz_f64(double);
  271. __device__ __attribute__((const)) double __ocml_fma_rte_f64(double, double,
  272.                                                             double);
  273. __device__ __attribute__((const)) double __ocml_fma_rtn_f64(double, double,
  274.                                                             double);
  275. __device__ __attribute__((const)) double __ocml_fma_rtp_f64(double, double,
  276.                                                             double);
  277. __device__ __attribute__((const)) double __ocml_fma_rtz_f64(double, double,
  278.                                                             double);
  279.  
  280. __device__ inline __attribute__((const)) double
  281. __llvm_amdgcn_rcp_f64(double __x) {
  282.   return __builtin_amdgcn_rcp(__x);
  283. }
  284. __device__ inline __attribute__((const)) double
  285. __llvm_amdgcn_rsq_f64(double __x) {
  286.   return __builtin_amdgcn_rsq(__x);
  287. }
  288.  
  289. __device__ __attribute__((const)) _Float16 __ocml_ceil_f16(_Float16);
  290. __device__ _Float16 __ocml_cos_f16(_Float16);
  291. __device__ __attribute__((const)) _Float16 __ocml_cvtrtn_f16_f32(float);
  292. __device__ __attribute__((const)) _Float16 __ocml_cvtrtp_f16_f32(float);
  293. __device__ __attribute__((const)) _Float16 __ocml_cvtrtz_f16_f32(float);
  294. __device__ __attribute__((pure)) _Float16 __ocml_exp_f16(_Float16);
  295. __device__ __attribute__((pure)) _Float16 __ocml_exp10_f16(_Float16);
  296. __device__ __attribute__((pure)) _Float16 __ocml_exp2_f16(_Float16);
  297. __device__ __attribute__((const)) _Float16 __ocml_floor_f16(_Float16);
  298. __device__ __attribute__((const)) _Float16 __ocml_fma_f16(_Float16, _Float16,
  299.                                                           _Float16);
  300. __device__ __attribute__((const)) _Float16 __ocml_fmax_f16(_Float16, _Float16);
  301. __device__ __attribute__((const)) _Float16 __ocml_fmin_f16(_Float16, _Float16);
  302. __device__ __attribute__((const)) _Float16 __ocml_fabs_f16(_Float16);
  303. __device__ __attribute__((const)) int __ocml_isinf_f16(_Float16);
  304. __device__ __attribute__((const)) int __ocml_isnan_f16(_Float16);
  305. __device__ __attribute__((pure)) _Float16 __ocml_log_f16(_Float16);
  306. __device__ __attribute__((pure)) _Float16 __ocml_log10_f16(_Float16);
  307. __device__ __attribute__((pure)) _Float16 __ocml_log2_f16(_Float16);
  308. __device__ __attribute__((const)) _Float16 __llvm_amdgcn_rcp_f16(_Float16);
  309. __device__ __attribute__((const)) _Float16 __ocml_rint_f16(_Float16);
  310. __device__ __attribute__((const)) _Float16 __ocml_rsqrt_f16(_Float16);
  311. __device__ _Float16 __ocml_sin_f16(_Float16);
  312. __device__ __attribute__((const)) _Float16 __ocml_sqrt_f16(_Float16);
  313. __device__ __attribute__((const)) _Float16 __ocml_trunc_f16(_Float16);
  314. __device__ __attribute__((pure)) _Float16 __ocml_pown_f16(_Float16, int);
  315.  
  316. typedef _Float16 __2f16 __attribute__((ext_vector_type(2)));
  317. typedef short __2i16 __attribute__((ext_vector_type(2)));
  318.  
  319. __device__ __attribute__((const)) float __ockl_fdot2(__2f16 a, __2f16 b,
  320.                                                      float c, bool s);
  321. __device__ __attribute__((const)) __2f16 __ocml_ceil_2f16(__2f16);
  322. __device__ __attribute__((const)) __2f16 __ocml_fabs_2f16(__2f16);
  323. __device__ __2f16 __ocml_cos_2f16(__2f16);
  324. __device__ __attribute__((pure)) __2f16 __ocml_exp_2f16(__2f16);
  325. __device__ __attribute__((pure)) __2f16 __ocml_exp10_2f16(__2f16);
  326. __device__ __attribute__((pure)) __2f16 __ocml_exp2_2f16(__2f16);
  327. __device__ __attribute__((const)) __2f16 __ocml_floor_2f16(__2f16);
  328. __device__ __attribute__((const))
  329. __2f16 __ocml_fma_2f16(__2f16, __2f16, __2f16);
  330. __device__ __attribute__((const)) __2i16 __ocml_isinf_2f16(__2f16);
  331. __device__ __attribute__((const)) __2i16 __ocml_isnan_2f16(__2f16);
  332. __device__ __attribute__((pure)) __2f16 __ocml_log_2f16(__2f16);
  333. __device__ __attribute__((pure)) __2f16 __ocml_log10_2f16(__2f16);
  334. __device__ __attribute__((pure)) __2f16 __ocml_log2_2f16(__2f16);
  335. __device__ inline __2f16
  336. __llvm_amdgcn_rcp_2f16(__2f16 __x) // Not currently exposed by ROCDL.
  337. {
  338.   return (__2f16)(__llvm_amdgcn_rcp_f16(__x.x), __llvm_amdgcn_rcp_f16(__x.y));
  339. }
  340. __device__ __attribute__((const)) __2f16 __ocml_rint_2f16(__2f16);
  341. __device__ __attribute__((const)) __2f16 __ocml_rsqrt_2f16(__2f16);
  342. __device__ __2f16 __ocml_sin_2f16(__2f16);
  343. __device__ __attribute__((const)) __2f16 __ocml_sqrt_2f16(__2f16);
  344. __device__ __attribute__((const)) __2f16 __ocml_trunc_2f16(__2f16);
  345. __device__ __attribute__((const)) __2f16 __ocml_pown_2f16(__2f16, __2i16);
  346.  
  347. #ifdef __cplusplus
  348. } // extern "C"
  349. #endif
  350.  
  351. #endif // __CLANG_HIP_LIBDEVICE_DECLARES_H__
  352.