diff --git a/.gitignore b/.gitignore index 3f79bce4e..5e6e56aea 100644 --- a/.gitignore +++ b/.gitignore @@ -2,9 +2,10 @@ *~ out/ *.pyc +gen/ # Files generated from cl.xml -generated/api +generated/ # Files generated by extraction from spec source man/*.txt diff --git a/api/cl_khr_command_buffer_mutable_dispatch.asciidoc b/api/cl_khr_command_buffer_mutable_dispatch.asciidoc index 1ef08b76e..980c2dc1f 100644 --- a/api/cl_khr_command_buffer_mutable_dispatch.asciidoc +++ b/api/cl_khr_command_buffer_mutable_dispatch.asciidoc @@ -111,6 +111,8 @@ void pointer using {cl_command_buffer_update_type_khr_TYPE}. ** {CL_COMMAND_BUFFER_MUTABLE_DISPATCH_ASSERTS_KHR} * {cl_command_buffer_update_type_khr_TYPE} ** {CL_STRUCTURE_TYPE_MUTABLE_DISPATCH_CONFIG_KHR} + * {cl_command_buffer_state_khr_TYPE} + ** {CL_COMMAND_BUFFER_STATE_FINALIZED_KHR} * New Error Codes ** {CL_INVALID_MUTABLE_COMMAND_KHR} @@ -379,3 +381,6 @@ may be a introduced as a stand alone extension. ** Move `CL_COMMAND_BUFFER_SIMULTANEOUS_USE_KHR` and `CL_COMMAND_BUFFER_CAPABILITY_SIMULTANEOUS_USE_KHR` in this extension from the base extension (experimental). + * Revision 0.9.5 2025-08-28 + ** Permitting recording ND-range kernel commands without having set all of + their arguments (experimental). diff --git a/api/opencl_runtime_layer.asciidoc b/api/opencl_runtime_layer.asciidoc index a28c4b7f8..80c501c8f 100644 --- a/api/opencl_runtime_layer.asciidoc +++ b/api/opencl_runtime_layer.asciidoc @@ -14768,9 +14768,16 @@ the recording state. Recording:: Initial state of a command-buffer on creation, where commands can be recorded to the command-buffer. +[[finalized]] +Finalized:: State after command recording has finished with +{clFinalizeCommandBufferKHR}, but there is at least one command for which not +all arguments or parameters have been set (this is a valid state only for +mutable command buffers). + [[executable]] Executable:: State after command recording has finished with -{clFinalizeCommandBufferKHR} and the command-buffer may be enqueued. +{clFinalizeCommandBufferKHR} and all the arguments and parameters of all +commands have been set. In this state the command-buffer may be enqueued. // Image generated from the following mermaid diagram description using https://mermaid.live // Ideally we'd use the asciidoctor-diagram extension to generate the rendered diagram, but @@ -14780,7 +14787,9 @@ Executable:: State after command recording has finished with // .... // stateDiagram-v2 // [*] --> Recording: Create -// Recording -->Executable: Finalize +// Recording -->Finalized: Finalize +// Finalized --> Executable: All commands arguments/parameters set +// Executable // .... image::images/commandbuffer_lifecycle.png[align="center", title="Lifecycle of a command-buffer."] @@ -15026,9 +15035,10 @@ include::{generated}/api/version-notes/clFinalizeCommandBufferKHR.asciidoc[] [NOTE] ==== -{clFinalizeCommandBufferKHR} places the command-buffer in the -<> state where commands can no longer be recorded, at -this point the command-buffer is ready to be enqueued. +{clFinalizeCommandBufferKHR} places the command-buffer either in the +<> or <> states where commands can +no longer be recorded. If the new state is <>, the +command-buffer is ready to be enqueued. ==== // refError @@ -16035,7 +16045,14 @@ ifdef::cl_khr_command_buffer_mutable_dispatch[] in the <> table. endif::cl_khr_command_buffer_mutable_dispatch[] +ifndef::cl_khr_command_buffer_mutable_dispatch[] * _kernel_ is a valid kernel object which **must** have its arguments set. +endif::cl_khr_command_buffer_mutable_dispatch[] +ifdef::cl_khr_command_buffer_mutable_dispatch[] + * _kernel_ is a valid kernel object, which **must** have its arguments set + unless the command has the {CL_MUTABLE_DISPATCH_ARGUMENTS_KHR} property set + (see description of argument _mutable_handle_). +endif::cl_khr_command_buffer_mutable_dispatch[] Any changes to _kernel_ after calling {clCommandNDRangeKernelKHR}, such as with {clSetKernelArg} or {clSetKernelExecInfo}, have no effect on the recorded command. @@ -16213,6 +16230,14 @@ The starting local ID is always (0, 0, ... 0). successfully. Otherwise, it returns the errors defined by {clEnqueueNDRangeKernel} except: +ifdef::cl_khr_command_buffer_mutable_dispatch[] +{CL_INVALID_KERNEL_ARGS} is replaced with: + + * {CL_INVALID_KERNEL_ARGS} if the kernel argument values have not been + specified and the {CL_MUTABLE_DISPATCH_ARGUMENTS_KHR} flag was not set in + the _properties_ parameter. +endif::cl_khr_command_buffer_mutable_dispatch[] + {CL_INVALID_COMMAND_QUEUE} is replaced with: * {CL_INVALID_COMMAND_QUEUE} if the @@ -16653,6 +16678,13 @@ device occupancy high by avoiding blocking in host code. [open,refpage='clUpdateMutableCommandsKHR',desc='Modify configuration of mutable-command handles to update behavior for future enqueues',type='protos'] -- + +When recording an ND-range kernel command, the kernel's arguments do not have to +be set, and setting them may be postponed to after the command buffer's +finalization (in which case, the command buffer is in Finalized state). +All the arguments must be set before the command buffer can be enqueued +for execution (Executable state). + To modify the configuration of mutable-command handles returned during _command_buffer_ recording, updating the behavior of those commands in future enqueues of _command_buffer_, call the function @@ -16909,8 +16941,15 @@ include::{generated}/api/version-notes/CL_COMMAND_BUFFER_STATE_KHR.asciidoc[] include::{generated}/api/version-notes/CL_COMMAND_BUFFER_STATE_RECORDING_KHR.asciidoc[] + {CL_COMMAND_BUFFER_STATE_FINALIZED_KHR_anchor} is returned when an + instance of _command_buffer_ has been finalized, but there is at least + one command for which not all arguments or parameters have been set. + +include::{generated}/api/version-notes/CL_COMMAND_BUFFER_STATE_FINALIZED_KHR.asciidoc[] + {CL_COMMAND_BUFFER_STATE_EXECUTABLE_KHR_anchor} is returned when - _command_buffer_ has been finalized. + _command_buffer_ has been finalized, all the arguments and parameters of + all commands have been set. include::{generated}/api/version-notes/CL_COMMAND_BUFFER_STATE_EXECUTABLE_KHR.asciidoc[] diff --git a/images/commandbuffer_lifecycle.png b/images/commandbuffer_lifecycle.png index 8d82abdf1..41c789db4 100644 Binary files a/images/commandbuffer_lifecycle.png and b/images/commandbuffer_lifecycle.png differ diff --git a/xml/cl.xml b/xml/cl.xml index 66210fcc9..fa5a4afd9 100644 --- a/xml/cl.xml +++ b/xml/cl.xml @@ -1371,6 +1371,7 @@ server's OpenCL/api-docs repository. + @@ -7507,7 +7508,7 @@ server's OpenCL/api-docs repository. - + @@ -7573,6 +7574,9 @@ server's OpenCL/api-docs repository. + + +