From a06d43114ea231cc41ad6fe57a56b7a0ee9a7750 Mon Sep 17 00:00:00 2001 From: Jatin Chaudhary Date: Thu, 20 Aug 2020 06:22:09 -0400 Subject: [PATCH] Adding dim3 conversion operator to __hip_builtin_threadIdx_t Change-Id: Ib7f9e45872ec6d48c0e5b7fc93ccecaafaeac754 [ROCm/clr commit: 9d4bdd466a041a951a72e43f750ed0510a18b861] --- .../clr/hipamd/include/hip/hcc_detail/hip_runtime.h | 12 ++++++++++++ 1 file changed, 12 insertions(+) diff --git a/projects/clr/hipamd/include/hip/hcc_detail/hip_runtime.h b/projects/clr/hipamd/include/hip/hcc_detail/hip_runtime.h index b277ba038c..7c5f7c7287 100644 --- a/projects/clr/hipamd/include/hip/hcc_detail/hip_runtime.h +++ b/projects/clr/hipamd/include/hip/hcc_detail/hip_runtime.h @@ -478,24 +478,36 @@ struct __hip_builtin_threadIdx_t { __HIP_DEVICE_BUILTIN(x,__hip_get_thread_idx_x()); __HIP_DEVICE_BUILTIN(y,__hip_get_thread_idx_y()); __HIP_DEVICE_BUILTIN(z,__hip_get_thread_idx_z()); +#ifdef __cplusplus + __device__ operator dim3() const { return dim3(x, y, z); } +#endif }; struct __hip_builtin_blockIdx_t { __HIP_DEVICE_BUILTIN(x,__hip_get_block_idx_x()); __HIP_DEVICE_BUILTIN(y,__hip_get_block_idx_y()); __HIP_DEVICE_BUILTIN(z,__hip_get_block_idx_z()); +#ifdef __cplusplus + __device__ operator dim3() const { return dim3(x, y, z); } +#endif }; struct __hip_builtin_blockDim_t { __HIP_DEVICE_BUILTIN(x,__hip_get_block_dim_x()); __HIP_DEVICE_BUILTIN(y,__hip_get_block_dim_y()); __HIP_DEVICE_BUILTIN(z,__hip_get_block_dim_z()); +#ifdef __cplusplus + __device__ operator dim3() const { return dim3(x, y, z); } +#endif }; struct __hip_builtin_gridDim_t { __HIP_DEVICE_BUILTIN(x,__hip_get_grid_dim_x()); __HIP_DEVICE_BUILTIN(y,__hip_get_grid_dim_y()); __HIP_DEVICE_BUILTIN(z,__hip_get_grid_dim_z()); +#ifdef __cplusplus + __device__ operator dim3() const { return dim3(x, y, z); } +#endif }; #undef __HIP_DEVICE_BUILTIN