https://bugs.freedesktop.org/show_bug.cgi?id=70779
Priority: medium Bug ID: 70779 Assignee: dri-devel@lists.freedesktop.org Summary: OpenCL hangs with big kernels Severity: normal Classification: Unclassified OS: All Reporter: mustrumr97@gmail.com Hardware: Other Status: NEW Version: git Component: Drivers/Gallium/radeonsi Product: Mesa
Radeon 7870 XT. A 50-word kernel works, a 67-word one hangs. Other tests confirm this. See bug 60879
https://bugs.freedesktop.org/show_bug.cgi?id=70779
--- Comment #1 from Hristo Venev mustrumr97@gmail.com --- Bump.
https://bugs.freedesktop.org/show_bug.cgi?id=70779
--- Comment #2 from Hristo Venev mustrumr97@gmail.com --- Bump.
https://bugs.freedesktop.org/show_bug.cgi?id=70779
EoD EoD@xmw.de changed:
What |Removed |Added ---------------------------------------------------------------------------- CC| |EoD@xmw.de
--- Comment #3 from EoD EoD@xmw.de --- Do you have an example of such a kernel?
https://bugs.freedesktop.org/show_bug.cgi?id=70779
--- Comment #4 from Hristo Venev mustrumr97@gmail.com --- Here are two kernels that fail: __kernel void uint_div(__global const uint *a, __global const uint *b, __global uint *c){ c[0]=a[0]/b[0]; } __kernel void uint_add16(__global const uint *a, __global const uint *b, __global uint *c){ for(uint i=0;i<16;i++) c[i]=a[i]+b[i]; }
This one works: __kernel void uint_add(__global const uint *a, __global const uint *b, __global uint *c){ c[0]=a[0]+b[0]; }
Sadly I don't have access to the hardware anymore (it's probably in a dump somewhere).
https://bugs.freedesktop.org/show_bug.cgi?id=70779
--- Comment #5 from EoD EoD@xmw.de --- Created attachment 121742 --> https://bugs.freedesktop.org/attachment.cgi?id=121742&action=edit dmesg after running the uint_add16() kernel
(In reply to Hristo Venev from comment #4)
Here are two kernels that fail: __kernel void uint_div(__global const uint *a, __global const uint *b, __global uint *c){ c[0]=a[0]/b[0]; } __kernel void uint_add16(__global const uint *a, __global const uint *b, __global uint *c){ for(uint i=0;i<16;i++) c[i]=a[i]+b[i]; }
I actually can confirm that the 2nd kernel does cause a GPU stall in the radeon driver (r600/Barts). It causes no problem with amdgpu (radeonsi/Tonga).
I am using kernel 4.5.0-rc3, current llvm 3.8 branch and current mesa git.
As I am not overly good with OpenCL, is this kernel somehow problematic?
https://bugs.freedesktop.org/show_bug.cgi?id=70779
Michel Dänzer michel@daenzer.net changed:
What |Removed |Added ---------------------------------------------------------------------------- Attachment #121742|text/x-log |text/plain mime type| |
https://bugs.freedesktop.org/show_bug.cgi?id=70779
--- Comment #6 from Michel Dänzer michel@daenzer.net --- I think this is probably a duplicate of bug 60879 and that it was basically luck that "small kernels" didn't hang on Hristo's card.
https://bugs.freedesktop.org/show_bug.cgi?id=70779
Vedran Miletić vedran@miletic.net changed:
What |Removed |Added ---------------------------------------------------------------------------- Resolution|--- |DUPLICATE Status|NEW |RESOLVED
--- Comment #7 from Vedran Miletić vedran@miletic.net --- (In reply to EoD from comment #5)
I actually can confirm that the 2nd kernel does cause a GPU stall in the radeon driver (r600/Barts). It causes no problem with amdgpu (radeonsi/Tonga).
I am using kernel 4.5.0-rc3, current llvm 3.8 branch and current mesa git.
As I am not overly good with OpenCL, is this kernel somehow problematic?
EoD, if these stalls on r600 are still reproducible, please open a separate bug.
*** This bug has been marked as a duplicate of bug 60879 ***
dri-devel@lists.freedesktop.org
-
bugzilla-daemon＠freedesktop.org