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

SWDEV-263959 - Fix device decl with long and llong

Browse Source 
Windows may expect long and ulong to be 4 bytes, while
Linux expects 8 bytes. Instead, use uint64_t for
unsigned long, and unsigned long long, and use int64_t
for long and long long to be consistent.

Change-Id: I6ed1cdde43721bcaaab0245644d607b1adbf9884
This commit is contained in:
Aaron En Ye Shi
2021-05-13 21:27:07 +00:00
committed by Aaron En Ye Shi
parent 77ba798177
commit dd643ec517
2 changed files with 5 additions and 2 deletions
Download Patch File Download Diff File Expand all files Collapse all files
hipamd/include/hip/amd_detail/device_functions.h
+1 -1
View File
@@ -59,7 +59,7 @@ __device__ static inline int __clz(int input) {
}
__device__ static inline int __clzll(long long int input) {
return __ockl_clz_u64((ullong)input);
return __ockl_clz_u64((uint64_t)input);
}
__device__ static inline unsigned int __ffs(unsigned int input) {
hipamd/include/hip/amd_detail/device_library_decls.h
+4 -1
View File
@@ -23,6 +23,9 @@ THE SOFTWARE.
/**
* @file amd_detail/device_library_decls.h
* @brief Contains declarations for types and functions in device library.
* Uses int64_t and uint64_t instead of long, long long, unsigned
* long and unsigned long long types for device library API
* declarations.
*/
#ifndef HIP_INCLUDE_HIP_AMD_DETAIL_DEVICE_LIBRARY_DECLS_H
@@ -49,7 +52,7 @@ extern "C" __device__ __attribute__((const)) uint __ockl_sadd_u32(uint, uint, ui
extern "C" __device__ __attribute__((const)) uchar __ockl_clz_u8(uchar);
extern "C" __device__ __attribute__((const)) ushort __ockl_clz_u16(ushort);
extern "C" __device__ __attribute__((const)) uint __ockl_clz_u32(uint);
extern "C" __device__ __attribute__((const)) ullong __ockl_clz_u64(ullong);
extern "C" __device__ __attribute__((const)) uint64_t __ockl_clz_u64(uint64_t);
extern "C" __device__ __attribute__((const)) float __ocml_floor_f32(float);
extern "C" __device__ __attribute__((const)) float __ocml_rint_f32(float);