[background] There is currently a compilation issue when both hip/hip_runtime.h and hip/hip_cooperative_groups.h included
in a file and hipcc used to compile it on NV platform. The issue is that an abort is defined in hip/nvcc_detail/hip_runtime.h
and it is also defined in the CUDA cooperative groups header (/cuda/include/cooperative_groups/details/helpers.h).
this is problematic and leads to a compilation issue in hipcc on NV platform.
Change-Id: I2ab6982ac4103822a1a4a0ced942cd604d6c19c1
[ROCm/hip commit: 3c72e7beea]
The static members __HIP_Coordinates::x, ::y and ::z must be defined
outside the class. Otherwise, linker throws `undefined reference error`
when these definitions are needed in the HIP application.
Change-Id: Iabc09744b478c22e4b13cf9824877ec9cfdd4f7a
[ROCm/hip commit: b8640f1f0d]
This makes hipLaunchKernelGGL take a variable argument list, that will be
expanded before being fed to hipLaunchKernelGGLInternal.
This is different from b2edee4693.
We try to accomodate the case when a kernel template has multiple
type parameters.
Change-Id: I87577d402c92b0f3b51e298f8293f4065e1f6de8
[ROCm/hip commit: affe9ab9b5]
- HIP-Clang follows the standard assert definition by providing
`__assert_fail`. But, `assert` macro is added as an HCC-specific
workaround due to the missing implementation. Only enable that on the
HCC compilation to avoid unexpected behaviors on HIP-Clang
compilation.
Change-Id: I1c9a707baff9b85c30faef58c52ebfe07e3fc3fc
[ROCm/hip commit: cea6b99a28]
This makes hipLaunchKernelGGL take a variable argument list, that will be
expanded before being fed to hipLaunchKernelGGLInternal.
Change-Id: Id76e2bf91acd5d68f56a24fc39f219f2eeb06d33
[ROCm/hip commit: 961717879d]
Add hipHostMallocNumaUser flag to hipHostMalloc() in order to support
numa policy set by user.
Change-Id: I6d70ed539a5f97f27187f2242b68849c0e27e4d6
[ROCm/hip commit: 63051ca2e1]
Currently std::complex and some other std functions require uses to
include hip_runtime.h before any other headers to work, which is not
reliable.
changes are made in clang to fix this issue:
https://reviews.llvm.org/D81176
which requires hipcc and HIP headers to make corresponding changes.
This patch will make sure the clang change will not break
HIP/ROCclr during this transition.
After the transition is done, we can remove explicitly setting
include path for HIP-Clang and HIP header in hipcc and hip config
cmake files and rely on clang driver to set it automatically.
Change-Id: I5d226861c2560ffa6c5ab17343a43cc378048061
[ROCm/hip commit: 087c579625]
1. Updated FAQ with shft*sync not supported hip_faq.md
2. Corrected some of input parameter description in hcc_details/hip_runtime_api.h
3. Redirect shfl*() to shfl_*_sync() for nvcc path where CUDA > 9.0
Change-Id: I3d8184db5fcc622852c9bad96b706348e8dfc16c
[ROCm/hip commit: 83b11f9a61]
This technique should never be used, and only accessed through
__builtins.
There's currently no builtin for groupstaticsize. I left ds_swizzle
since for some reason it switches to the builtin based on __HCC__ or
not.
Change-Id: If1e1394221dba83ea4add6db5e94d6b715552044
[ROCm/hip commit: d2dd307c7d]
Aryan Salmanpour
mshivama
Sarbojit Sarkar
agodavar
Rahul Garg
Siu Chi Chan
Jatin Chaudhary
Todd tiantuo Li
kjayapra-amd
Yaxun (Sam) Liu
Ronak Chauhan
Daniil Fukalov
Michael LIAO
Tao Sang
Jason Tang
German Andryeyev
Dittakavi Satyanvesh
Evgeny
Matt Arsenault
Payam
Vlad Sytchenko
Alex Xie