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

Add more hip_bug.md entry, regarding hang after hipLaunchKernel

Browse Source 
Change-Id: I5800cb627179ec0e913cd36d332fb8c2994ab71e


[ROCm/clr commit: a8fcbb2584]
This commit is contained in:
pensun
2017-01-24 22:43:25 -06:00
parent 5d356d12de
commit 8bcdd56fa0
1 changed files with 10 additions and 1 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/clr/hipamd/docs/markdown/hip_bugs.md
+10 -1
View File
@@ -2,7 +2,8 @@
<!-- toc -->
- [Errors related to undefined reference to `__hcLaunchKernel__***__grid_launch_parm**](#error-undefined-reference)
- [Errors related to undefined reference to `__hcLaunchKernel__***__grid_launch_parm**](#errors-related-to-undefined-reference-to-hclaunchkernel__grid_launch_parm)
- [Application hangs after a hipLaunchKernel call](#what-if-i-see-application-hangs-after-a-hiplaunchkernel-call)
<!-- tocstop -->
@@ -22,3 +23,11 @@ namespace {
__global__ MyKernel …
- Avoid calling member functions
### What if I see application hangs after a hipLaunchKernel call?
If hipLaunchKernel takes parameters that request explicitly memcpy, then it will cause application hang.
Reason is that the hipLaunchKernel macro locks the stream.
If kernel paramters are actually function calls which invoke other hip apis (i.e. memcpy) to the same stream, then deadlock occurs.
To workaround, try:
Move the function calls so they occur outside the hipLaunchKernel macro, store results in temps, then use the tems inside the kernel.