OpenCL latency spikes during concurrent out-of-order queue submissions
0 reputation · 20 Sept 2025, 13:34 UTC
When utilizing multiple OpenCL command queues configured with CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, a measurable latency gap often emerges between clEnqueueNDRangeKernel and the CL_PROFILING_COMMAND_START timestamp. This delay appears to manifest specifically when multiple queues submit kernels simultaneously, whereas single-queue operations maintain consistent start times.
The OpenCL specification does not define a maximum bound for kernel launch scheduling, leaving the behavior to the driver implementation. It is unclear if this latency is a necessary result of the runtime serializing requests across a shared device context or an overhead from managing event-based synchronization under high contention.
How do different OpenCL runtimes prioritize kernel scheduling when multiple out-of-order queues target the same device? Is there a documented mechanism to minimize the enqueue-to-start delay in concurrent-request environments without switching to blocking in-order queues?