* In hipFree, if memory is associated with a device, synchronize that device's streams.
This changes the behavior from synchronizing the currently set TLS device.
* All devices sync in hipFree for _appId=-1 case.
* Revert "All devices sync in hipFree for _appId=-1 case."
This reverts commit 1efb34d6a8426661e45bc5f763422a1147aeac10.
* add HIP_SYNC_FREE env var
[ROCm/clr commit: cf8fb43e6b]
* Update hip_runtime_api.h
when i try to use mpicc or gcc to compile a c language code which call some hip runtime api , error occured as
> /path/to/hcc_detail/hip_runtime_api.h:2268:33: error: unknown type name ‘hipFuncAttributes’;
> hipFuncGetAttributes(hipFuncAttributes* attr, const void* func);
add ' struct ' for the first parameter of hipFuncGetAttributes will get ride of this problem.
[ROCm/clr commit: 64bdf82265]
* Add CMAKE dep to amd_comgr
* Use COMGR for read_kernarg_metadata in COV2
* Do not assume kernargs exist
* Add proper metadata destroy cleanup
* Use a process function for easier destroy
* Remove old read_kernarg_metadata
* Clean up HCC, prints, names
* Use COMGR in CMAKE by default
* Move metadata lookup for keyword values into helper
* Remove C string usage for lookup_keyword_value
* Guard COMGR for non-NVCC path
* Add hip_hcc dependency on comgr package
* Add lifetime to metadata nodes
* Find COMGR config file for amd_comgr target
* Move set_active data earlier
[ROCm/clr commit: 2c80975e9c]
Define HIP_COMPILER and HIP_RUNTIME in hip-config.cmake to
facilitate other packages identify hip-clang and HIP/VDI
runtime.
[ROCm/clr commit: a781d2e017]
+ Fix the error in cmake: clang 8.0.0 supports CUDA 10.0, not 10.1; only trunk clang 9.0.0 supports CUDA 10.1.
+ Update Readme:
1. clang's bug 36384 is gone, so latest stable release for Linux is clang 8.0.0, which supports CUDA up to 10.0;
2. update dependencies, testing and building notes.
[ROCm/clr commit: 79e6bd8426]
Due to a HCC bug in rocm-head, the test fails to compile. Disabling the test until the issue is resolved.
Change-Id: Ib4e7baf0b9c2cb5f5dbe38e6dd2bab894d28886a
[ROCm/clr commit: 27013c9af6]
__hipRegisterFunction is called during by .init functions during program initialization.
It calls hipModuleGetFunction to locate kernel symbol in code objects. hipModuleGetFunction
assumes current device when locating kernel symbols. This works for HCC but not for hip-clang,
since hip-clang needs to locate kernel symbols for different devices without switching
between devices.
This patch introduces a new hsa agent parameter to ihipModuleGetFunction, which allows
__hipRegisterFunction to choose the correct hsa agent when locating kernel symbols. By
default it uses this_agent(), therefore this patch has no impact on HCC.
[ROCm/clr commit: 8f5c812a68]
* Make hipModuleGetGlobal be in HIP runtime so it can be discovered at runtime
In HIP PR #929, quite a few HIP public APIs were made as inline functions with
hidden visibility. It was necessary to support applications with shared
libraries with GPU kernels launched via hipLaunchKernelGGL(), after HIP runtime
is initialized.
In empirical tests, the implementation has been proved to be a bit too
excessive, especially for hipModuleGetGlobal(). The function is used by another
type of client applications which relies on the existence of this function
within HIP runtime so global symbols from HSA code objects loaded dynamically
at runtime can be retrieved programmtically.
This commit moves hipModuleGetGlobal() back to src/hip_module.cpp, and makes it
visible and not inline, to fulfill requirements for applications
aforementioned. It does not change the behavior of applications depending on
hipLaunchKernelGGL().
* Add HIP_INIT_API into the implementation of hipModuleGetGlobal
Address review comments.
* Fix failing HIP unit tests
[ROCm/clr commit: 04915cea2f]
Disambiguate calling many varibles "agent".
More detail in exception message.
Create and discard map placeholders; no need to call std::vector::clear() on map value.
[ROCm/clr commit: f5e4fff6cc]
For code objects with global symbols of length 0, ROCR runtime would
ignore them even though they exist in the symbol table. Therefore the
result from read_agent_globals() can't be trusted entirely.
As a workaround to tame applications which depend on the existence of
global symbols with length 0, always return hipSuccess here.
This behavior shall be reverted once ROCR runtime has been fixed to
address SWDEV-173477
[ROCm/clr commit: cf7ad0f184]
* adding prof primitives generator
* minor change, renaming
* minor cosmetic changes, comments correcting and dead code removing
* minor changes and renaming
* minor chane, fixing comments
[ROCm/clr commit: 1229750546]
* reimplement HIP_INIT as a function, expose it as hip_impl::hip_init()
so that it could be called from hipLaunchKernelGGL and other inlined
HIP functions
* Don't call hip_init from ihipPreLaunchKernel
[ROCm/clr commit: fa9495841b]
Jeff Daily
Mr-LiuSw
Aaron Enye Shi
Evgeny Mankov
Maneesh Gupta
Yaxun (Sam) Liu
Anthony Cowley
Wen-Heng (Jack) Chung
Rahul Garg
Michael LIAO
Nico Trost
eshcherb
Siu Chi Chan