[Beignet] clEnqueueNDRangeKernel and kernel completion

[Zou, Nanhai]

Hi,
         Thanks for reporting this. It looks like a bug. The driver use wait_rendering on last_batch.
But nobody is actually rendering into last_batch.
We will fix that.
Thanks
Zou Nanhai

From: beignet-bounces+nanhai.zou=intel.com at lists.freedesktop.org [mailto:beignet-bounces+nanhai.zou=intel.com at lists.freedesktop.org] On Behalf Of Edward Ching
Sent: Thursday, June 06, 2013 8:03 AM
To: beignet at lists.freedesktop.org
Subject: [Beignet] clEnqueueNDRangeKernel and kernel completion

I hope this is the right forum to post comments/questions on Beignet OpenCL API behaviour. If not, please ignore and excuse the disruption.
I'm running Beignet on an Ivy Bridge machine and noticed that clFinish would return before the GPU has complete processing of previously submitted commands.

e.g:
I submitted an OpenCL kernel via clEnqueueNDRangeKernel, followed by clFinish, and expected the GPU to have finished all processing when clFinish returns.

But clFinish returned right away, and when I then call clEnqueueMapBuffer to access data, the call blocks, so I traced the logic all the way into the Ivy Bridge GPU device driver (~/drivers/gpu/drm/i915/*), and it looks like every IvyBr GPU batchbuffer used to submit an OpenCL kernel has an associated sequence number which the GPU would write to a special location which has to be monitored in order to tell if the GPU has finished processing the submtted kernel (details in Intel HD graphics PRM and i915/GEM design notes).  Neither clFinish nor clEnqueueNDRangeKernel does this monitoring, and clEnqueueMapBuffer happened to do it when it tried to transfer the buffer objects from GPU to CPU domain.
Does this make sense? It seems to me that clFinish should instead be monitoring and blocking if the GPU is still busy executing an OpenCL kernel.
/Ed
-------------- next part --------------
An HTML attachment was scrubbed...
URL: <http://lists.freedesktop.org/archives/beignet/attachments/20130606/1f43734f/attachment.html>