diff --git a/OpenCL_Ext.txt b/OpenCL_Ext.txt index 06cb08bc..b0972853 100644 --- a/OpenCL_Ext.txt +++ b/OpenCL_Ext.txt @@ -84,6 +84,11 @@ include::ext/cl_khr_suggested_local_work_size.asciidoc[] include::ext/cl_khr_integer_dot_product.asciidoc[] +include::ext/cl_khr_semaphore.asciidoc[] + +include::ext/cl_khr_external_semaphore.asciidoc[] +include::ext/cl_khr_external_memory.asciidoc[] + // NOTE: To keep meaningful section numbers, new // extension documents should be added above here! diff --git a/ext/cl_khr_external_memory.asciidoc b/ext/cl_khr_external_memory.asciidoc new file mode 100644 index 00000000..f1928b50 --- /dev/null +++ b/ext/cl_khr_external_memory.asciidoc @@ -0,0 +1,588 @@ +// Copyright 2021 The Khronos Group. This work is licensed under a +// Creative Commons Attribution 4.0 International License; see +// http://creativecommons.org/licenses/by/4.0/ + +[[cl_khr_external_memory]] +== External Memory (Provisional) + +This extension defines a generic mechanism to share buffer and image objects between OpenCL and many other APIs. + +In particular, the `cl_khr_external_memory` extension defines: + +* Optional properties to import external memory exported by other APIs into OpenCL for a set of devices. + +* Routines to explicitly hand off memory ownership between OpenCL and other APIs. + +Other related extensions define specific external memory types that may be imported into OpenCL. + +=== General Information + +==== Name Strings + +`cl_khr_external_memory` + +`cl_khr_external_memory_dma_buf` + +`cl_khr_external_memory_dx` + +`cl_khr_external_memory_opaque_fd` + +`cl_khr_external_memory_win32` + +==== Version History + +[cols="1,1,3",options="header",] +|==== +| *Date* | *Version* | *Description* +| 2021-09-10 | 0.9.0 | Initial version (provisional). +|==== + +NOTE: This is a preview of an OpenCL provisional extension specification that has been Ratified under the Khronos Intellectual Property Framework. It is being made publicly available prior to being uploaded to the Khronos registry to enable review and feedback from the community. If you have feedback please create an issue on https://github.com/KhronosGroup/OpenCL-Docs/ + +==== Dependencies + +This extension is written against the OpenCL Specification Version 3.0.8. + +Because this extension adds new properties for {clCreateBufferWithProperties} +and {clCreateImageWithProperties} this extension requires OpenCL 3.0. + +==== Contributors + +// spell-checker: disable +Ajit Hakke-Patil, NVIDIA + +Amit Rao, NVIDIA + +Balaji Calidas, QUALCOMM + +Ben Ashbaugh, INTEL + +Carsten Rohde, NVIDIA + +Christoph Kubisch, NVIDIA + +Debalina Bhattacharjee, NVIDIA + +James Jones, NVIDIA + +Jason Ekstrand, INTEL + +Jeremy Kemp, IMAGINATION + +Joshua Kelly, QUALCOMM + +Karthik Raghavan Ravi, NVIDIA + +Kedar Patil, NVIDIA + +Kevin Petit, ARM + +Nikhil Joshi, NVIDIA + +Sharan Ashwathnarayan, NVIDIA + +Vivek Kini, NVIDIA + +// spell-checker: enable + +=== New Types + +[source] +---- +typedef cl_uint cl_external_memory_handle_type_khr; +---- + +=== New API Functions + +[source] +---- +cl_int clEnqueueAcquireExternalMemObjectsKHR( + cl_command_queue command_queue, + cl_uint num_mem_objects, + const cl_mem *mem_objects, + cl_uint num_events_in_wait_list, + const cl_event *event_wait_list, + cl_event *event); + +cl_int clEnqueueReleaseExternalMemObjectsKHR( + cl_command_queue command_queue, + cl_uint num_mem_objects, + const cl_mem *mem_objects, + cl_uint num_events_in_wait_list, + const cl_event *event_wait_list, + cl_event *event); +---- + +=== New API Enums + +Accepted value for the _param_name_ parameter to {clGetPlatformInfo} to query external memory handle types that may be imported by all devices in an OpenCL platform: + +[source] +---- +CL_PLATFORM_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR 0x2044 +---- + +Accepted value for the _param_name_ parameter to {clGetDeviceInfo} to query external memory handle types that may be imported by an OpenCL device: + +[source] +---- +CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR 0x204F +---- + +New properties accepted as _properties_ to {clCreateBufferWithProperties} and {clCreateImageWithProperties}: + +[source] +---- +CL_DEVICE_HANDLE_LIST_KHR 0x2051 +CL_DEVICE_HANDLE_LIST_END_KHR 0 +---- + +New return values from {clGetEventInfo} when _param_name_ is {CL_EVENT_COMMAND_TYPE}: + +[source] +---- +CL_COMMAND_ACQUIRE_EXTERNAL_MEM_OBJECTS_KHR 0x2047 +CL_COMMAND_RELEASE_EXTERNAL_MEM_OBJECTS_KHR 0x2048 +---- + +External memory handle type added by `cl_khr_external_memory_dma_buf`: + +[source] +---- +CL_EXTERNAL_MEMORY_HANDLE_DMA_BUF_KHR 0x2067 +---- + +External memory handle types added by `cl_khr_external_memory_dx`: + +[source] +---- +CL_EXTERNAL_MEMORY_HANDLE_D3D11_TEXTURE_KHR 0x2063 +CL_EXTERNAL_MEMORY_HANDLE_D3D11_TEXTURE_KMT_KHR 0x2064 +CL_EXTERNAL_MEMORY_HANDLE_D3D12_HEAP_KHR 0x2065 +CL_EXTERNAL_MEMORY_HANDLE_D3D12_RESOURCE_KHR 0x2066 +---- + +External memory handle type added by `cl_khr_external_memory_opaque_fd`: + +[source] +---- +CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR 0x2060 +---- + +External memory handle types added by `cl_khr_external_memory_win32`: + +[source] +---- +CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR 0x2061 +CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR 0x2062 +---- + +=== Modifications to existing APIs added by this spec + +Following new enums are added to the list of supported _param_names_ by {clGetPlatformInfo}: + +.List of supported param_names by clGetPlatformInfo +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Platform Info | Return Type | Description +| {CL_PLATFORM_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR} + | {cl_external_memory_handle_type_khr_TYPE}[] + | Returns the list of importable external memory handle types supported by all devices in _platform_. +|==== + +{clGetPlatformInfo} when called with _param_name_ {CL_PLATFORM_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR} must return a common list of external memory handle types supported by all devices in the platform. + +Following new enums are added to the list of supported _param_names_ by {clGetDeviceInfo}: + +.List of supported param_names by clGetDeviceInfo +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Device Info | Return Type | Description +| {CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR} + | {cl_external_memory_handle_type_khr_TYPE}[] + | Returns the list of importable external memory handle types supported by _device_. +|==== + +{clGetDeviceInfo} when called with param_name {CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR} must return a non-empty list of external memory handle types for at least one of the devices in the platform. + +Following new properties are added to the list of supported properties by {clCreateBufferWithProperties} and {clCreateImageWithProperties}. + +[[external-memory-properties-table]] +.List of supported buffer and image creation properties +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Property | Property Value | Description +| {CL_DEVICE_HANDLE_LIST_KHR} + | {cl_device_id_TYPE}[] + | Specifies the list of OpenCL devices (terminated with {CL_DEVICE_HANDLE_LIST_END_KHR}) to associate with the external memory handle. +|==== + +If {CL_DEVICE_HANDLE_LIST_KHR} is not specified as part of _properties_, the memory object created by {clCreateBufferWithProperties} or {clCreateImageWithProperties} is by default accessible to all devices in the _context_. + +The properties used to create a buffer or image from an external memory handle are described by related extensions. +When a buffer or image is created from an external memory handle, the _flags_ used to specify usage information for the buffer or image must not include {CL_MEM_USE_HOST_PTR}, {CL_MEM_ALLOC_HOST_PTR}, or {CL_MEM_COPY_HOST_PTR}, and the _host_ptr_ argument must be `NULL`. + +Add to the list of error conditions for {clCreateBufferWithProperties} and {clCreateImageWithProperties}: + +* {CL_INVALID_DEVICE} + ** if a device identified by the property {CL_DEVICE_HANDLE_LIST_KHR} is not a valid device or is not associated with _context_, or + ** if a device identified by property {CL_DEVICE_HANDLE_LIST_KHR} cannot import the requested external memory object type, or + ** if {CL_DEVICE_HANDLE_LIST_KHR} is not specified as part of _properties_ and one or more devices in _context_ cannot import the requested external memory object type. +* {CL_INVALID_VALUE} + ** if _properties_ includes a supported external memory handle and _flags_ includes {CL_MEM_USE_HOST_PTR}, {CL_MEM_ALLOC_HOST_PTR}, or {CL_MEM_COPY_HOST_PTR}. +* {CL_INVALID_HOST_PTR} + ** if _properties_ includes a supported external memory handle and _host_ptr_ is not `NULL`. + +Add images created from an external memory handle to the description of `image_row_pitch` and `image_slice_pitch` for {cl_image_desc_TYPE}: + +* `image_row_pitch` is the scan-line pitch in bytes. +The `image_row_pitch` must be zero if _host_ptr_ is `NULL`, the image is not a +2D image created from a buffer, and the image is not an image created from an +external memory handle. +If `image_row_pitch` is zero and _host_ptr_ is not `NULL` then the image row +pitch is calculated as `image_width` {times} the size of an image element in +bytes. +If `image_row_pitch` is zero and the image is created from an external memory +handle then the image row pitch is implementation-defined. +The image row pitch must be {geq} `image_width` {times} the size of an image +element in bytes and must be a multiple of the size of an image element in +bytes. +For a 2D image created from a buffer the image row pitch must also be a multiple +of the maximum of the {CL_DEVICE_IMAGE_PITCH_ALIGNMENT} value for all devices in +the context that support images. + +* `image_slice_pitch` is the size in bytes of each 2D slice in a 3D image or the +size in bytes of each image in a 1D or 2D image array. +The `image_slice_pitch` must be zero if _host_ptr_ is `NULL` and the image is +not created from an external memory handle. +If `image_slice_pitch` is zero and _host_ptr_ is not `NULL` then the image slice +pitch is calculated as the image row pitch {times} `image_height` for a 2D image +array or a 3D image, and as the image row pitch for a 1D image array. +If `image_slice_pitch` is zero and the image is created from an external memory +handle then the image slice pitch is implementation-defined. +The image slice pitch must be {geq} the image image row pitch {times} +`image_height` for a 2D image array or a 3D image, must be {geq} the image row +pitch for a 1D image array, and must be a multiple of the image row pitch. + +=== Description of new types added by this spec + +The following new APIs are added as part of this spec. The details of each are described below: + +==== Acquiring and Releasing External Memory Objects + +To acquire OpenCL memory objects created from external memory handles, call the function + +include::{generated}/api/protos/clEnqueueAcquireExternalMemObjectsKHR.txt[] + +_command_queue_ specifies a valid command-queue. + +_num_mem_objects_ specifies the number of memory objects to acquire. + +_mem_objects_ points to a list of valid memory objects. + +_num_events_in_wait_list_ specifies the number of events in _event_wait_list_. + +_event_wait_list_ points to the list of events that need to complete before {clEnqueueAcquireExternalMemObjectsKHR} can be executed. +If _event_wait_list_ is `NULL`, then {clEnqueueAcquireExternalMemObjectsKHR} does not explicitly wait on any event to complete. +If _event_wait_list_ is `NULL`, _num_events_in_wait_list_ must be 0. +If _event_wait_list_ is not `NULL`, the list of events pointed to by _event_wait_list_ must be valid and _num_events_in_wait_list_ must be greater than 0. +The events specified in _event_wait_list_ act as synchronization points. +The context associated with events in _event_wait_list_ and that of _command_queue_ must be the same. + +_event_ returns an event object that identifies this particular command and can be used to query or queue a wait for this particular command to complete. +_event_ can be `NULL` in which case it will not be possible for the application to query the status of this command or queue a wait for this command to complete. + +Applications must acquire the memory objects that are created using external handles before they can be used by any OpenCL commands queued to a command-queue. +Behavior is undefined if a memory object created from an external memory handle is used by an OpenCL command queued to a command-queue without being acquired. +This is to guarantee that the state of the memory objects is up-to-date and they are accessible to OpenCL. +See "Example with Acquire / Release" provided in <> for more details on how to use this API. + +If _num_mem_objects_ is 0 and _mem_objects_ is `NULL`, the command will trivially succeed after its event dependencies are satisfied and will update its completion event. + +{clEnqueueAcquireExternalMemObjectsKHR} returns {CL_SUCCESS} if the function is executed successfully. +Otherwise, it returns one of the following errors: + +* {CL_INVALID_VALUE} if _num_mem_objects_ is zero and _mem_objects_ is not a `NULL` value or if _num_mem_objects_ is not 0 and _mem_objects_ is `NULL`. +* {CL_INVALID_MEM_OBJECT} if any of the memory objects in _mem_objects_ is not a valid OpenCL memory object created using an external memory handle. +* {CL_INVALID_COMMAND_QUEUE} +** if _command_queue_ is not a valid command-queue, or +** if device associated with _command_queue_ is not one of the devices specified by {CL_DEVICE_HANDLE_LIST_KHR} at the time of creating one or more of _mem_objects_, or +** if one or more of _mem_objects_ belong to a context that does not contain a device associated _command_queue_, or +** if one or more of _mem_objects_ can not be shared with device associated with _command_queue_. +* {CL_INVALID_EVENT_WAIT_LIST} + ** if _event_wait_list_ is `NULL` and _num_events_in_wait_list_ is not 0, or + ** if _event_wait_list_ is not `NULL` and _num_events_in_wait_list_ is 0, or + ** if event objects in _event_wait_list_ are not valid events. +* {CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST} if the execution status of any of the events in _event_wait_list_ is a negative integer value. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +To release OpenCL memory objects created from external memory handles, call the function + +include::{generated}/api/protos/clEnqueueReleaseExternalMemObjectsKHR.txt[] + +_command_queue_ specifies a valid command-queue. + +_num_mem_objects_ specifies the number of memory objects to release. + +_mem_objects_ points to a list of valid memory objects. + +_num_events_in_wait_list_ specifies the number of events in _event_wait_list_. + +_event_wait_list_ points to the list of events that need to complete before {clEnqueueReleaseExternalMemObjectsKHR} can be executed. +If _event_wait_list_ is `NULL`, then {clEnqueueReleaseExternalMemObjectsKHR} does not wait on any event to complete. +If _event_wait_list_ is `NULL`, _num_events_in_wait_list_ must be 0. +If _event_wait_list_ is not `NULL`, the list of events pointed to by _event_wait_list_ must be valid and _num_events_in_wait_list_ must be greater than 0. +The events specified in _event_wait_list_ act as synchronization points. +The context associated with events in _event_wait_list_ and that of _command_queue_ must be the same. + +_event_ returns an event object that identifies this particular command and can be used to query or queue a wait for this particular command to complete. +_event_ can be `NULL` in which case it will not be possible for the application to query the status of this command or queue a wait for this command to complete. + +Applications must release the memory objects that are acquired using {clEnqueueReleaseExternalMemObjectsKHR} before using them through any commands in the other API. +This is to guarantee that the state of memory objects is up-to-date and they are accessible to the other API. +See "Example with Acquire / Release" provided in <> for more details on how to use this API. + +If _num_mem_objects_ is 0 and _mem_objects_ is `NULL`, the command will trivially succeed after its event dependencies are satisfied and will update its completion event. + +{clEnqueueReleaseExternalMemObjectsKHR} returns {CL_SUCCESS} if the function is executed successfully. +Otherwise, it returns one of the following errors: + +* {CL_INVALID_VALUE} if _num_mem_objects_ is zero and _mem_objects_ is not a `NULL` value or if _num_mem_objects_ is greater than 0 and _mem_objects_ is `NULL`. +* {CL_INVALID_MEM_OBJECT} if any of the memory objects in _mem_objects_ is not a valid OpenCL memory object created using an external memory handle. +* {CL_INVALID_COMMAND_QUEUE} +** if _command_queue_ is not a valid command-queue, or +** if device associated with _command_queue_ is not one of the devices specified by {CL_DEVICE_HANDLE_LIST_KHR} at the time of creating one or more of _mem_objects_, or +** if one or more of _mem_objects_ belong to a cl_context that does not contain device associated _command_queue_, or +** if one or more of _mem_objects_ can not be shared with device associated with _command_queue_. +* {CL_INVALID_EVENT_WAIT_LIST} + ** if _event_wait_list_ is `NULL` and _num_events_in_wait_list_ is not 0, or + ** if _event_wait_list_ is not `NULL` and _num_events_in_wait_list_ is 0, or + ** if event objects in _event_wait_list_ are not valid events. +* {CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST} if the execution status of any of the events in _event_wait_list_ is a negative integer value. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +=== Descriptions of External Memory Handle Types + +This section describes the external memory handle types that are added by related extensions. + +Applications can import the same payload into multiple OpenCL contexts and multiple times into a given OpenCL context. In all cases, each import operation must create a distinct memory object. + +==== File Descriptor Handle Types + +The `cl_khr_external_memory_opaque_fd` extension extends {cl_external_memory_handle_type_khr_TYPE} to support the following new types of handles, and adds as a property that may be specified when creating a buffer or an image memory object from an external handle: + +-- + * {CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR} specifies a POSIX file descriptor handle that has only limited valid usage outside of OpenCL and other compatible APIs. It must be compatible with the POSIX system calls dup, dup2, close, and the non-standard system call dup3. Additionally, it must be transportable over a socket using an SCM_RIGHTS control message. It owns a reference to the underlying memory resource represented by its memory object. +-- + +The `cl_khr_external_memory_dma_buf` extension extends {cl_external_memory_handle_type_khr_TYPE} to support the following types of handles, and adds as a property that may be specified when creating a buffer or an image memory object from an external handle: + +-- + * {CL_EXTERNAL_MEMORY_HANDLE_DMA_BUF_KHR} is a file descriptor for a Linux dma_buf. It owns a reference to the underlying memory resource represented by its memory object. +-- + +For these extensions, importing memory from a file descriptor transfers ownership of the file descriptor from the application to the OpenCL implementation. The application must not perform any operations on the file descriptor after a successful import. The imported memory object holds a reference to its payload. + +==== NT Handle Types + +The `cl_khr_external_memory_dx` extension extends {cl_external_memory_handle_type_khr_TYPE} to support the following new types of handles, and adds as a property that may be specified when creating a buffer or an image memory object from an external handle: + +-- + * {CL_EXTERNAL_MEMORY_HANDLE_D3D11_TEXTURE_KHR} specifies an NT handle returned by IDXGIResource1::CreateSharedHandle referring to a Direct3D 10 or 11 texture resource. It owns a reference to the memory used by the Direct3D resource. + + * {CL_EXTERNAL_MEMORY_HANDLE_D3D11_TEXTURE_KMT_KHR} specifies a global share handle returned by IDXGIResource::GetSharedHandle referring to a Direct3D 10 or 11 texture resource. It does not own a reference to the underlying Direct3D resource, and will therefore become invalid when all memory objects and Direct3D resources associated with it are destroyed. + + * {CL_EXTERNAL_MEMORY_HANDLE_D3D12_HEAP_KHR} specifies an NT handle returned by ID3D12Device::CreateSharedHandle referring to a Direct3D 12 heap resource. It owns a reference to the resources used by the Direct3D heap. + + * {CL_EXTERNAL_MEMORY_HANDLE_D3D12_RESOURCE_KHR} specifies an NT handle returned by ID3D12Device::CreateSharedHandle referring to a Direct3D 12 committed resource. It owns a reference to the memory used by the Direct3D resource. +-- + +The `cl_khr_external_memory_win32` extension extends {cl_external_memory_handle_type_khr_TYPE} to support the following new types of handles, and adds as a property that may be specified when creating a buffer or an image memory object from an external handle: + +-- + * {CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR} specifies an NT handle that has only limited valid usage outside of OpenCL and other compatible APIs. It must be compatible with the functions DuplicateHandle, CloseHandle, CompareObjectHandles, GetHandleInformation, and SetHandleInformation. It owns a reference to the underlying memory resource represented by its memory object. + * {CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR} specifies a global share handle that has only limited valid usage outside of OpenCL and other compatible APIs. It is not compatible with any native APIs. It does not own a reference to the underlying memory resource represented by its memory object, and will therefore become invalid when all memory objects associated with it are destroyed. +-- + +For these extensions, importing memory object payloads from Windows handles does not transfer ownership of the handle to the OpenCL implementation. For handle types defined as NT handles, the application must release handle ownership using the CloseHandle system call when the handle is no longer needed. For handle types defined as NT handles, the imported memory object holds a reference to its payload. + +Note: Non-NT handle import operations do not add a reference to their associated payload. If the original object owning the payload is destroyed, all resources and handles sharing that payload will become invalid. + +[[cl_khr_external_memory-Sample-Code]] +=== Sample Code + +. Example for creating a CL buffer from an exported external buffer in a single device context. ++ +-- +[source] +---- +// Get cl_devices of the platform. +clGetDeviceIDs(..., &devices, &deviceCount); + +// Create cl_context with just first device +clCreateContext(..., 1, devices, ...); + +// Obtain fd/win32 or similar handle for external memory to be imported +// from other API. +int fd = getFdForExternalMemory(); + +// Create extMemBuffer of type cl_mem from fd. +cl_mem_properties_khr extMemProperties[] = +{ + (cl_mem_properties_khr)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR, + (cl_mem_properties_khr)fd, + 0 +}; + +cl_mem extMemBuffer = clCreateBufferWithProperties(/*context*/ clContext, + /*properties*/ extMemProperties, + /*flags*/ 0, + /*size*/ size, + /*host_ptr*/ NULL, + /*errcode_ret*/ &errcode_ret); +---- +-- +. Example for creating a CL Image from an exported external Image for single device usage in a multi-device context ++ +-- +[source] +---- +// Get cl_devices of the platform. +clGetDeviceIDs(..., &devices, &deviceCount); + +// Create cl_context with first two devices +clCreateContext(..., 2, devices, ...); + +// Create img of type cl_mem usable only on devices[0] + +// Create img of type cl_mem. +// Obtain fd/win32 or similar handle for external memory to be imported +// from other API. +int fd = getFdForExternalMemory(); + +// Set cl_image_format based on external image info +cl_image_format clImgFormat = { }; +clImageFormat.image_channel_order = CL_RGBA; +clImageFormat.image_channel_data_type = CL_UNORM_INT8; + +// Set cl_image_desc based on external image info +size_t clImageFormatSize; +cl_image_desc image_desc = { }; +image_desc.image_type = CL_MEM_OBJECT_IMAGE2D_ARRAY; +image_desc.image_width = width; +image_desc.image_height = height; +image_desc.image_depth = depth; +image_desc.image_array_size = num_slices; +image_desc.image_row_pitch = width * 8 * 4; // May need alignment +image_desc.image_slice_pitch = image_desc.image_row_pitch * height; +image_desc.num_mip_levels = 1; +image_desc.num_samples = 0; +image_desc.buffer = NULL; + +cl_mem_properties_khr extMemProperties[] = +{ + (cl_mem_properties_khr)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR, + (cl_mem_properties_khr)fd, + (cl_mem_properties_khr)CL_DEVICE_HANDLE_LIST_KHR, + (cl_mem_properties_khr)devices[0], + CL_DEVICE_HANDLE_LIST_END_KHR, + 0 +}; + +cl_mem img = clCreateImageWithProperties(/*context*/ clContext, + /*properties*/ extMemProperties, + /*flags*/ 0, + /*image_format*/ &clImgFormat, + /*image_desc*/ &image_desc, + /*errcode_ret*/ &errcode_ret); + +// Use clGetImageInfo to get cl_image_format details. +size_t clImageFormatSize; +clGetImageInfo(img, + CL_IMAGE_FORMAT, + sizeof(cl_image_format), + &clImageFormat, + &clImageFormatSize); +---- +-- +. Example for synchronization using Wait and Signal ++ +-- +[source] +---- +// Start the main rendering loop + +// Create extSem of type cl_semaphore_khr using clCreateSemaphoreWithPropertiesKHR + +// Create extMem of type cl_mem using clCreateBufferWithProperties or clCreateImageWithProperties + +while (true) { + // (not shown) Signal the semaphore from the other API + + // Wait for the semaphore in OpenCL, by calling clEnqueueWaitSemaphoresKHR on 'extSem' + clEnqueueWaitSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &extSem, + /*sema_payload_list*/ NULL, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Launch kernel that accesses extMem + clEnqueueNDRangeKernel(command_queue, ...); + + // Signal the semaphore in OpenCL + clEnqueueSignalSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &extSem, + /*sema_payload_list*/ NULL, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // (not shown) Launch work in other API that waits on 'extSem' +} +---- +-- +. Example with memory sharing using acquire/release ++ +-- +[source] +---- +// Create extSem of type cl_semaphore_khr using +// clCreateSemaphoreWithPropertiesKHR with CL_SEMAPHORE_HANDLE_*_KHR. + +// Create extMem1 and extMem2 of type cl_mem using clCreateBufferWithProperties +// or clCreateImageWithProperties + +while (true) { + // (not shown) Signal the semaphore from the other API. Wait for the + // semaphore in OpenCL, by calling clEnqueueWaitForSemaphore on extSem + clEnqueueWaitSemaphoresKHR(/*command_queue*/ cq1, + /*num_sema_objects*/ 1, + /*sema_objects*/ &extSem, + /*sema_payload_list*/ NULL, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Get explicit ownership of extMem1 + clEnqueueAcquireExternalMemObjectsKHR(/*command_queue*/ cq1, + /*num_mem_objects*/ 1, + /*mem_objects*/ extMem1, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Launch kernel that accesses extMem1 on cq1 on cl_device1 + clEnqueueNDRangeKernel(cq1, ..., &event1); + + // Launch kernel that accesses both extMem1 and extMem2 on cq2 on cl_device2 + // Migration of extMem1 and extMem2 handles through regular CL memory + // migration. + clEnqueueNDRangeKernel(cq2, ..., &event1, &event2); + + // Give up ownership of extMem1 before you signal the semaphore. Handle + // memory migration here. + clEnqueueReleaseExternalMemObjectsKHR(/*command_queue*/ cq2 + /*num_mem_objects*/ 1, + /*mem_objects*/ &extMem1, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Signal the semaphore from OpenCL + clEnqueueSignalSemaphoresKHR(/*command_queue*/ cq2, + /*num_sema_objects*/ 1, + /*sema_objects*/ &extSem, + /*sema_payload_list*/ NULL, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // (not shown) Launch work in other API that waits on 'extSem' + // Other API accesses ext1, but not ext2 on device-1 +} +---- +-- diff --git a/ext/cl_khr_external_semaphore.asciidoc b/ext/cl_khr_external_semaphore.asciidoc new file mode 100644 index 00000000..bfeee695 --- /dev/null +++ b/ext/cl_khr_external_semaphore.asciidoc @@ -0,0 +1,654 @@ +// Copyright 2021 The Khronos Group. This work is licensed under a +// Creative Commons Attribution 4.0 International License; see +// http://creativecommons.org/licenses/by/4.0/ + +[[cl_khr_external_semaphore]] +== External Semaphores (Provisional) + +`cl_khr_semaphore` introduced semaphores as a new type along with a set of APIs for create, release, retain, wait and signal operations on it. +This extension defines APIs and mechanisms to share semaphores created in an external API by importing into and exporting from OpenCL. + +This extension defines: + +* New attributes that can be passed as part of {cl_semaphore_properties_khr_TYPE} for specifying properties of external semaphores to be imported or exported. + +* New attributes that can be passed as part of {cl_semaphore_info_khr_TYPE} for specifying properties of external semaphores to be exported. + +* An extension to {clCreateSemaphoreWithPropertiesKHR} to accept external semaphore properties allowing to import or export an external semaphore into or from OpenCL. + +* Semaphore handle types required for importing and exporting semaphores. + +* Modifications to Wait and Signal API behavior when dealing with external semaphores created from different handle types. + +* API query exportable semaphores handles using specified handle type. + +Other related extensions define specific external semaphores that may be imported into or exported from OpenCL. + +=== General Information + +==== Name Strings + +`cl_khr_external_semaphore` + +`cl_khr_external_semaphore_dx_fence` + +`cl_khr_external_semaphore_opaque_fd` + +`cl_khr_external_semaphore_sync_fd` + +`cl_khr_external_semaphore_win32` + +==== Version History + +[cols="1,1,3",options="header",] +|==== +| *Date* | *Version* | *Description* +| 2021-09-10 | 0.9.0 | Initial version (provisional). +|==== + +NOTE: This is a preview of an OpenCL provisional extension specification that has been Ratified under the Khronos Intellectual Property Framework. It is being made publicly available prior to being uploaded to the Khronos registry to enable review and feedback from the community. If you have feedback please create an issue on https://github.com/KhronosGroup/OpenCL-Docs/ + +==== Dependencies + +This extension is written against the OpenCL Specification Version 3.0.8. + +This extension requires OpenCL 1.2. + +The `cl_khr_semaphore` extension is required as it defines semaphore objects as well as for wait and signal operations on semaphores. + +For OpenCL to be able to import external semaphores from other APIs using this extension, the other API is required to provide below mechanisms: + +* Ability to export semaphore handles +* Ability to query semaphore handle in the form of one of the handle type supported by OpenCL. + +The other APIs that want to use semaphore exported by OpenCL using this extension are required to provide below mechanism: + +* Ability to import semaphore handles using handle types exported by OpenCL. + +==== Contributors + +// spell-checker: disable +Ajit Hakke-Patil, NVIDIA + +Amit Rao, NVIDIA + +Balaji Calidas, QUALCOMM + +Ben Ashbaugh, INTEL + +Carsten Rohde, NVIDIA + +Christoph Kubisch, NVIDIA + +Debalina Bhattacharjee, NVIDIA + +James Jones, NVIDIA + +Jason Ekstrand, INTEL + +Jeremy Kemp, IMAGINATION + +Joshua Kelly, QUALCOMM + +Karthik Raghavan Ravi, NVIDIA + +Kedar Patil, NVIDIA + +Kevin Petit, ARM + +Nikhil Joshi, NVIDIA + +Sharan Ashwathnarayan, NVIDIA + +Vivek Kini, NVIDIA + +// spell-checker: enable + +=== New Types + +[source] +---- +typedef cl_uint cl_external_semaphore_handle_type_khr; +---- + +=== New API Functions + +[source] +---- +cl_int clGetSemaphoreHandleForTypeKHR( + cl_semaphore_khr sema_object, + cl_device device, + cl_external_semaphore_handle_type_khr handle_type, + size_t handle_size, + void *handle_ptr, + size_t *handle_size_ret); +---- + +=== New API Enums + +Accepted value for the _param_name_ parameter to {clGetPlatformInfo} to query external semaphore handle types that may be imported or exported by all devices in an OpenCL platform: + +[source] +---- +CL_PLATFORM_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR 0x2037 +CL_PLATFORM_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR 0x2038 +---- + +Accepted value for the _param_name_ parameter to {clGetDeviceInfo} to query external semaphore handle types that may be imported or exported by an OpenCL device: + +[source] +---- +CL_DEVICE_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR 0x204D +CL_DEVICE_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR 0x204E +---- + +Following new attributes can be passed as part of {cl_semaphore_properties_khr_TYPE} and {cl_semaphore_info_khr_TYPE}: + +[source] +---- +CL_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR 0x203F +CL_SEMAPHORE_EXPORT_HANDLE_TYPES_LIST_END_KHR 0 +---- + +External semaphore handle type added by `cl_khr_external_semaphore_dx_fence`: + +[source] +---- +CL_SEMAPHORE_HANDLE_D3D12_FENCE_KHR 0x2059 +---- + +External semaphore handle type added by `cl_khr_external_semaphore_opaque_fd`: + +[source] +---- +CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR 0x2055 +---- + +External semaphore handle type added by `cl_khr_external_semaphore_sync_fd`: + +[source] +---- +CL_SEMAPHORE_HANDLE_SYNC_FD_KHR 0x2058 +---- + +External semaphore handle types added by `cl_khr_external_semaphore_win32`: + +[source] +---- +CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR 0x2056 +CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR 0x2057 +---- + +=== Modifications to existing APIs added by this spec + +Following new enums are added to the list of supported _param_names_ by {clGetPlatformInfo}: + +.List of supported param_names by clGetPlatformInfo +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Platform Info | Return Type | Description +| {CL_PLATFORM_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR} + | {cl_external_semaphore_handle_type_khr_TYPE}[] + | Returns the list of importable external semaphore handle types supported by all devices in _platform_. + This size of this query may be 0 if no importable external semaphore handle types are supported by all devices in _platform_. +| {CL_PLATFORM_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR} + | {cl_external_semaphore_handle_type_khr_TYPE}[] + | Returns the list of exportable external semaphore handle types supported by all devices in the platform. + This size of this query may be 0 if no exportable external semaphore handle types are supported by all devices in _platform_. +|==== + +{clGetPlatformInfo} when called with _param_name_ {CL_PLATFORM_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR} returns a common list of external semaphore handle types supported for importing by all devices in the platform. + +{clGetPlatformInfo} when called with _param_name_ {CL_PLATFORM_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR} returns a common list of external semaphore handle types supported for exporting by all devices in the platform. + +Following new enums are added to the list of supported _param_names_ by {clGetDeviceInfo}: + +.List of supported param_names by clGetDeviceInfo +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Device Info | Return Type | Description +| {CL_DEVICE_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR} + | {cl_external_semaphore_handle_type_khr_TYPE}[] + | Returns the list of importable external semaphore handle types supported by _device_. + This size of this query may be 0 indicating that the device does not support importing semaphores. +| {CL_DEVICE_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR} + | {cl_external_semaphore_handle_type_khr_TYPE}[] + | Returns the list of exportable external semaphore handle types supported by _device_. + This size of this query may be 0 indicating that the device does not support exporting semaphores. +|==== + +{clGetDeviceInfo} when called with _param_name_ {CL_DEVICE_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR} returns a list of external semaphore handle types supported for importing. + +{clGetDeviceInfo} when called with _param_name_ {CL_DEVICE_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR} returns a list of external semaphore handle types supported for exporting. + +One of the above two queries {CL_DEVICE_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR} and {CL_DEVICE_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR} must return a non-empty list indicating support for at least one of the valid semaphore handles types either for import or for export or both. + +Following new properties are added to the list of possible supported properties by {clCreateSemaphoreWithPropertiesKHR}: + +.List of supported semaphore creation properties by clCreateSemaphoreWithPropertiesKHR +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Semaphore Property | Property Value | Description +// This is already described in cl_khr_semaphore so we don't need to describe it again here. +//| {CL_DEVICE_HANDLE_LIST_KHR} +// | {cl_device_id_TYPE}[] +// | Specifies the list of OpenCL devices (terminated with +// {CL_DEVICE_HANDLE_LIST_END_KHR}) to associate with the semaphore. +// This is also already described in cl_khr_semaphore so we don't need to describe it again here. +//| {CL_SEMAPHORE_TYPE_KHR} +// | {cl_semaphore_type_khr_TYPE} +// | Specifies the type of semaphore to create. +| {CL_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR} + | {cl_external_semaphore_handle_type_khr_TYPE}[] + | Specifies the list of semaphore handle type properties terminated with + {CL_SEMAPHORE_EXPORT_HANDLE_TYPES_LIST_END_KHR} that can be used to export + the semaphore being created. +|==== + +Add to the list of error conditions for {clCreateSemaphoreWithPropertiesKHR}: + +// This is in the base spec so we don't need to describe it here. +//* {CL_INVALID_DEVICE} if one or more devices identified by properties +//{CL_DEVICE_HANDLE_LIST_KHR} is/are not part of devices within _context_ to which +//{cl_semaphore_khr_TYPE} being created will belong to. +* {CL_INVALID_DEVICE} if one or more devices identified by properties {CL_DEVICE_HANDLE_LIST_KHR} can not import the requested external semaphore handle type. + +{clCreateSemaphoreWithPropertiesKHR} may return a NULL value on some implementations if _sema_props_ does not contain an external semaphore handle type to import. +Such implementations are required to return a valid semaphore when a supported external memory handle type and valid external semaphore handle is specified. + +Add to the list of supported _param_names_ by {clGetSemaphoreInfoKHR}: + +.List of supported param_names by clGetSemaphoreInfoKHR +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Semaphore Info | Return Type | Description +// These are already in the base cl_khr_semaphore so we don't need to include them again here. +//| *CL_SEMAPHORE_CONTEXT_KHR* +// | cl_context * +// | cl_context +// | Returns the cl_context associated with the {cl_semaphore_khr_TYPE}. +//| *CL_SEMAPHORE_REFERENCE_COUNT_KHR* +// | cl_uint * +// | cl_uint +// | Returns the reference count associated with the {cl_semaphore_khr_TYPE}. +//| *CL_SEMAPHORE_PROPERTIES_KHR* +// | cl_semaphore_properties_khr** +// | cl_semaphore_properties_khr* +// | Returns the array of properties associated with the {cl_semaphore_khr_TYPE}. +//| *CL_SEMAPHORE_TYPE_KHR* +// | cl_semaphore_type_khr * +// | cl_semaphore_type_khr +// | Returns the type of the {cl_semaphore_khr_TYPE}. +//| *CL_SEMAPHORE_PAYLOAD_KHR* +// | cl_semaphore_payload_khr * +// | cl_semaphore_payload_khr +// | Returns the payload value of the {cl_semaphore_khr_TYPE}. For semaphore of +// type CL_SEMAPHORE_TYPE_BINARY_KHR, payload value returned should be 0 if the +// semaphore is in unsignaled state and 1 if it is in signaled +// state. +//| *CL_DEVICE_HANDLE_LIST_KHR* +// | cl_device_id ** +// | cl_device_id * +// | Returns a NULL terminated list of cl_device_id (terminated with +// CL_DEVICE_HANDLE_LIST_END_KHR) for OpenCL devices the semaphore is associated +// with. +| {CL_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR} + | {cl_external_semaphore_handle_type_khr_TYPE}[] + | Returns the list of external semaphore handle types that may be used for + exporting. The size of this query may be 0 indicating that this + semaphore does not support any handle types for exporting. +|==== + +=== Exporting semaphore external handles + +To export an external handle from a semaphore, call the function + +include::{generated}/api/protos/clGetSemaphoreHandleForTypeKHR.txt[] + +_sema_object_ specifies a valid semaphore object with exportable properties. + +_device_ specifies a valid device for which a semaphore handle is being requested. + +_handle_type_ specifies the type of semaphore handle that should be returned for this exportable _sema_object_ and must be one of the values specified when _sema_object_ was created. + +_handle_size_ specifies the size of memory pointed by _handle_ptr_. + +_handle_ptr_ is a pointer to memory where the exported external handle is returned. +If _param_value_ is `NULL`, it is ignored. + +_handle_size_ret_ returns the actual size in bytes for the external handle. +If _handle_size_ret_ is `NULL`, it is ignored. + +{clGetSemaphoreHandleForTypeKHR} returns {CL_SUCCESS} if the semaphore handle is queried successfully. +Otherwise, it returns one of the following errors: + +* {CL_INVALID_SEMAPHORE_KHR} +** if _sema_object_ is not a valid semaphore +// This is redundant with the error below. +** if _sema_object_ is not exportable +* {CL_INVALID_DEVICE} +** if _device_ is not a valid device, or +** if _sema_object_ belongs to a context that is not associated with _device_, or +** if _sema_object_ can not be shared with _device_. +* {CL_INVALID_VALUE} if the requested external semaphore handle type was not specified when _sema_object_ was created. +// I don't think this can happen. This would have been checked when the semaphore was created. +// ** if CL_SEMAPHORE_HANDLE_*_KHR is specified as one of the _sema_props_ and +// the property CL_SEMAPHORE_HANDLE_*_KHR does not identify a valid external +// memory handle poperty reported by +// CL_PLATFORM_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR or +// CL_DEVICE_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR queries. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +=== Importing semaphore external handles + +Applications can import a semaphore payload into an existing semaphore using an +external semaphore handle. The effects of the import operation will be either +temporary or permanent, as specified by the application. If the import is +temporary, the implementation must restore the semaphore to its prior permanent +state after submitting the next semaphore wait operation. Performing a +subsequent temporary import on a semaphore before performing a semaphore wait +has no effect on this requirement; the next wait submitted on the semaphore must +still restore its last permanent state. A permanent payload import behaves as if +the target semaphore was destroyed, and a new semaphore was created with the +same handle but the imported payload. Because importing a semaphore payload +temporarily or permanently detaches the existing payload from a semaphore, +similar usage restrictions to those applied to {clReleaseSemaphoreKHR} are +applied to any command that imports a semaphore payload. Which of these import +types is used is referred to as the import operation's permanence. Each handle +type supports either one or both types of permanence. + +The implementation must perform the import operation by either referencing or +copying the payload referred to by the specified external semaphore handle, +depending on the handle's type. The import method used is referred to as the +handle type's transference. When using handle types with reference transference, +importing a payload to a semaphore adds the semaphore to the set of all +semaphores sharing that payload. This set includes the semaphore from which the +payload was exported. Semaphore signaling and waiting operations performed on +any semaphore in the set must behave as if the set were a single semaphore. +Importing a payload using handle types with copy transference creates a +duplicate copy of the payload at the time of import, but makes no further +reference to it. Semaphore signaling and waiting operations performed on the +target of copy imports must not affect any other semaphore or payload. + +Export operations have the same transference as the specified handle type's +import operations. Additionally, exporting a semaphore payload to a handle with +copy transference has the same side effects on the source semaphore's payload as +executing a semaphore wait operation. If the semaphore was using a temporarily +imported payload, the semaphore's prior permanent payload will be restored. + +Please refer to handle specific specifications for more details on transference and +permanence requirements specific to handle type. + +=== Descriptions of External Semaphore Handle Types + +This section describes the external semaphore handle types that are added by related extensions. + +Applications can import the same semaphore payload into multiple OpenCL contexts, into the same context from which it was exported, and multiple times into a given OpenCL context. +In all cases, each import operation must create a distinct semaphore object. + +==== File Descriptor Handle Types + +The `cl_khr_external_semaphore_opaque_fd` extension extends {cl_external_semaphore_handle_type_khr_TYPE} to support the following new types of handles, and adds as a property that may be specified when creating a semaphore from an external handle: + +-- + * {CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR} specifies a POSIX file descriptor handle that has only limited valid usage outside of OpenCL and other compatible APIs. It must be compatible with the POSIX system calls dup, dup2, close, and the non-standard system call dup3. Additionally, it must be transportable over a socket using an SCM_RIGHTS control message. It owns a reference to the underlying synchronization primitive represented by its semaphore object. +-- + +Transference and permanence properties for handle types added by `cl_khr_external_semaphore_opaque_fd`: + +.Transference and Permanence Properties for `cl_khr_external_semaphore_opaque_fd` handles +[width="100%",cols="60%,<20%,<20%",options="header"] +|==== +| Handle Type | Transference | Permanence +| {CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR} + | Reference + | Temporary, Permanent +|==== + +The `cl_khr_external_semaphore_sync_fd` extension extends {cl_external_semaphore_handle_type_khr_TYPE} to support the following new types of handles, and adds as a property that may be specified when creating a semaphore from an external handle: + +-- + * *CL_SEMAPHORE_HANDLE_SYNC_FD_KHR* specifies a POSIX file descriptor handle to a Linux Sync File or Android Fence object. It can be used with any native API accepting a valid sync file or fence as input. It owns a reference to the underlying synchronization primitive associated with the file descriptor. Implementations which support importing this handle type must accept any type of sync or fence FD supported by the native system they are running on. +-- + +The special value -1 for fd is treated like a valid sync file descriptor referring to an object that has already signaled. The import operation will succeed and the semaphore will have a temporarily imported payload as if a valid file descriptor had been provided. + +Note: This special behavior for importing an invalid sync file descriptor allows easier interoperability with other system APIs which use the convention that an invalid sync file descriptor represents work that has already completed and does not need to be waited for. It is consistent with the option for implementations to return a -1 file descriptor when exporting a {CL_SEMAPHORE_HANDLE_SYNC_FD_KHR} from a {cl_semaphore_khr_TYPE} which is signaled. + +Transference and permanence properties for handle types added by `cl_khr_external_semaphore_sync_fd`: + +.Transference and Permanence Properties for `cl_khr_external_semaphore_sync_fd` handles +[width="100%",cols="60%,<20%,<20%",options="header"] +|==== +| Handle Type | Transference | Permanence +| {CL_SEMAPHORE_HANDLE_SYNC_FD_KHR} + | Copy + | Temporary +|==== + +For these extensions, importing a semaphore payload from a file descriptor transfers ownership of the file descriptor from the application to the OpenCL implementation. The application must not perform any operations on the file descriptor after a successful import. + +==== NT Handle Types + +The `cl_khr_external_semaphore_dx_fence` extension extends {cl_external_semaphore_handle_type_khr_TYPE} to support the following new types of handles, and adds as a property that may be specified when creating a semaphore from an external handle: + +-- + * {CL_SEMAPHORE_HANDLE_D3D12_FENCE_KHR} specifies an NT handle returned by ID3D12Device::CreateSharedHandle referring to a Direct3D 12 fence, or ID3D11Device5::CreateFence referring to a Direct3D 11 fence. It owns a reference to the underlying synchronization primitive associated with the Direct3D fence. +-- + +When waiting on semaphores using {clEnqueueWaitSemaphoresKHR} or signaling semaphores using {clEnqueueSignalSemaphoresKHR}, the semaphore payload must be provided for semaphores created from {CL_SEMAPHORE_HANDLE_D3D12_FENCE_KHR}. + + +* If _sema_objects_ list has a mix of cl_semaphore obtained from *CL_SEMAPHORE_HANDLE_D3D12_FENCE_KHR* and other handle types, +then the _sema_payload_list_ should point to a list of _num_sema_objects_ payload values for each cl_semaphore in _sema_objects_. +However, the payload values corresponding to semaphores with type CL_SEMAPHORE_TYPE_BINARY_KHR can be set to 0 or will be ignored. + +*clEnqueueWaitSemaphoresKHR* and *clEnqueueSignalSemaphoresKHR* may return *CL_INVALID_VALUE* if _sema_objects_ list has one or more cl_semaphore obtained from *CL_SEMAPHORE_HANDLE_D3D12_FENCE_KHR* and _sema_payload_list_ is NULL. + + + +Transference and permanence properties for handle types added by `cl_khr_external_semaphore_dx_fence`: + +-- +.Transference and Permanence Properties for `cl_khr_external_semaphore_dx_fence` handles +[width="100%",cols="60%,<20%,<20%",options="header"] +|==== +| Handle Type | Transference | Permanence +| {CL_SEMAPHORE_HANDLE_D3D12_FENCE_KHR} + | Reference + | Temporary, Permanent +|==== +-- + +The `cl_khr_external_semaphore_win32` extension extends {cl_external_semaphore_handle_type_khr_TYPE} to support the following new types of handles, and adds as a property that may be specified when creating a semaphore from an external handle: + +-- + * {CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR} specifies an NT handle that has only limited valid usage outside of OpenCL and other compatible APIs. It must be compatible with the functions DuplicateHandle, CloseHandle, CompareObjectHandles, GetHandleInformation, and SetHandleInformation. It owns a reference to the underlying synchronization primitive represented by its semaphore object. + * {CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR} specifies a global share handle that has only limited valid usage outside of OpenCL and other compatible APIs. It is not compatible with any native APIs. It does not own a reference to the underlying synchronization primitive represented by its semaphore object, and will therefore become invalid when all semaphore objects associated with it are destroyed. +-- + +Transference and permanence properties for handle types added by `cl_khr_external_semaphore_win32`: + +.Transference and Permanence Properties for `cl_khr_external_semaphore_win32` handles +[width="100%",cols="60%,<20%,<20%",options="header"] +|==== +| Handle Type | Transference | Permanence +| *CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR* + | Reference + | Temporary, Permanent +| *CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR* + | Reference + | Temporary, Permanent +|==== + +For these extensions, importing a semaphore payload from Windows handles does not transfer ownership of the handle to the OpenCL implementation. For handle types defined as NT handles, the application must release ownership using the CloseHandle system call when the handle is no longer needed. + +[[cl_khr_external_semaphore-Sample-Code]] +=== Sample Code + +. Example for importing a semaphore created by another API in OpenCL in a single-device context. ++ +-- +[source] +---- +// Get cl_devices of the platform. +clGetDeviceIDs(..., &devices, &deviceCount); + +// Create cl_context with just first device +clCreateContext(..., 1, devices, ...); + +// Obtain fd/win32 or similar handle for external semaphore to be imported +// from the other API. +int fd = getFdForExternalSemaphore(); + +// Create clSema of type cl_semaphore_khr usable on the only available device +// assuming the semaphore was imported from the same device. + +cl_semaphore_properties_khr sema_props[] = + {(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR, + (cl_semaphore_properties_khr)fd, + 0}; + + +int errcode_ret = 0; +cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context, + sema_props, + &errcode_ret); + +---- +-- +. Example for importing a semaphore created by another API in OpenCL in a multi-device context for single device usage. ++ +-- +[source] +---- +// Get cl_devices of the platform. +clGetDeviceIDs(..., &devices, &deviceCount); + +// Create cl_context with first two devices +clCreateContext(..., 2, devices, ...); + +// Obtain fd/win32 or similar handle for external semaphore to be imported +// from the other API. +int fd = getFdForExternalSemaphore(); + +// Create clSema of type cl_semaphore_khr usable only on devices[1] +// assuming the semaphore was imported from the same device. + +cl_semaphore_properties_khr sema_props[] = + {(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR, + (cl_semaphore_properties_khr)fd, + (cl_semaphore_properties_khr)CL_DEVICE_HANDLE_LIST_KHR, + (cl_semaphore_properties_khr)devices[1], CL_DEVICE_HANDLE_LIST_END_KHR, + 0}; + + +int errcode_ret = 0; +cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context, + sema_props, + &errcode_ret); + +---- +-- +. Example for synchronization using a semaphore created by another API and imported in OpenCL ++ +-- +[source] +---- +// Create clSema using one of the above examples of external semaphore creation. + +int errcode_ret = 0; +cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context, + sema_props, + &errcode_ret); + +// Start the main loop + +while (true) { + // (not shown) Signal the semaphore from the other API + + // Wait for the semaphore in OpenCL + clEnqueueWaitSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &clSema, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Launch kernel + clEnqueueNDRangeKernel(command_queue, ...); + + // Signal the semaphore in OpenCL + clEnqueueSignalSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &clSema, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // (not shown) Launch work in the other API that waits on 'clSema' + +} +---- +-- +. Example for synchronization using semaphore exported by OpenCL ++ +-- +[source] +---- + +// Get cl_devices of the platform. +clGetDeviceIDs(..., &devices, &deviceCount); + +// Create cl_context with first two devices +clCreateContext(..., 2, devices, ...); + +// Create clSema of type cl_semaphore_khr usable only on device 1 +cl_semaphore_properties_khr sema_props[] = + {(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR, + CL_SEMAPHORE_EXPORT_HANDLE_TYPES_LIST_END_KHR, + (cl_semaphore_properties_khr)CL_DEVICE_HANDLE_LIST_KHR, + (cl_semaphore_properties_khr)devices[1], + CL_DEVICE_HANDLE_LIST_END_KHR, + 0}; + +int errcode_ret = 0; +cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context, + sema_props, + &errcode_ret); + +// Application queries handle-type and the exportable handle associated with the semaphore. +clGetSemaphoreInfoKHR(clSema, + CL_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR, + sizeof(cl_external_semaphore_handle_type_khr), + &handle_type, + &handle_type_size); + +// The other API or process can use the exported semaphore handle +// to import +int fd = -1; +if (handle_type == CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR) { + clGetSemaphoreHandleForTypeKHR(clSema, + device, + CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR, + sizeof(int), + &fd, + NULL); +} + +// Start the main rendering loop + +while (true) { + // (not shown) Signal the semaphore from the other API + + // Wait for the semaphore in OpenCL + clEnqueueWaitSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &clSema, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Launch kernel + clEnqueueNDRangeKernel(command_queue, ...); + + // Signal the semaphore in OpenCL + clEnqueueSignalSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &clSema, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // (not shown) Launch work in the other API that waits on 'clSema' + +} +---- +-- diff --git a/ext/cl_khr_semaphore.asciidoc b/ext/cl_khr_semaphore.asciidoc new file mode 100644 index 00000000..53e527fb --- /dev/null +++ b/ext/cl_khr_semaphore.asciidoc @@ -0,0 +1,639 @@ +// Copyright 2021 The Khronos Group. This work is licensed under a +// Creative Commons Attribution 4.0 International License; see +// http://creativecommons.org/licenses/by/4.0/ + +[[cl_khr_semaphore]] +== Semaphores (Provisional) + +OpenCL provides {cl_event_TYPE} as a primary mechanism of synchronization between host and device as well as across devices. +While events can be waited on or can be passed as dependencies across work-submissions, they suffer from following limitations: + +* They are immutable. + +* They are not reusable. + +This extension introduces a new type of synchronization object to represent semaphores that can be reused, waited on, and signaled multiple times by OpenCL work-submissions. + +In particular, this extension defines: + +* A new type called {cl_semaphore_khr_TYPE} to represent the semaphore objects. + +* A new type called {cl_semaphore_properties_khr_TYPE} to specify metadata associated with semaphores. + +* Routines to create, retain, and release semaphores. + +* Routines to wait on and signal semaphore objects. + +* Routine to query the properties of semaphore objects. + +=== General Information + +==== Name Strings + +`cl_khr_semaphore` + +==== Version History + +[cols="1,1,3",options="header",] +|==== +| *Date* | *Version* | *Description* +| 2021-09-10 | 0.9.0 | Initial version (provisional). +|==== + +NOTE: This is a preview of an OpenCL provisional extension specification that has been Ratified under the Khronos Intellectual Property Framework. It is being made publicly available prior to being uploaded to the Khronos registry to enable review and feedback from the community. If you have feedback please create an issue on https://github.com/KhronosGroup/OpenCL-Docs/ + +==== Dependencies + +This extension is written against the OpenCL Specification Version 3.0.8. + +This extension requires OpenCL 1.2. + +==== Contributors + +// spell-checker: disable +Ajit Hakke-Patil, NVIDIA + +Amit Rao, NVIDIA + +Balaji Calidas, QUALCOMM + +Ben Ashbaugh, INTEL + +Carsten Rohde, NVIDIA + +Christoph Kubisch, NVIDIA + +Debalina Bhattacharjee, NVIDIA + +James Jones, NVIDIA + +Jason Ekstrand, INTEL + +Jeremy Kemp, IMAGINATION + +Joshua Kelly, QUALCOMM + +Karthik Raghavan Ravi, NVIDIA + +Kedar Patil, NVIDIA + +Kevin Petit, ARM + +Nikhil Joshi, NVIDIA + +Sharan Ashwathnarayan, NVIDIA + +Vivek Kini, NVIDIA + +// spell-checker: enable + +=== New Types + +[source] +---- +typedef struct _cl_semaphore_khr* cl_semaphore_khr; + +typedef cl_properties cl_semaphore_properties_khr; +typedef cl_uint cl_semaphore_info_khr; +typedef cl_uint cl_semaphore_type_khr; +typedef cl_ulong cl_semaphore_payload_khr; +---- + +=== New API Functions + +[source] +---- +cl_semaphore_khr clCreateSemaphoreWithPropertiesKHR( + cl_context context, + const cl_semaphore_properties_khr *sema_props, + cl_int *errcode_ret); + +cl_int clEnqueueWaitSemaphoresKHR( + cl_command_queue command_queue, + cl_uint num_sema_objects, + const cl_semaphore_khr *sema_objects, + const cl_semaphore_payload_khr *sema_payload_list, + cl_uint num_events_in_wait_list, + const cl_event *event_wait_list, + cl_event *event); + +cl_int clEnqueueSignalSemaphoresKHR( + cl_command_queue command_queue, + cl_uint num_sema_objects, + const cl_semaphore_khr *sema_objects, + const cl_semaphore_payload_khr *sema_payload_list, + cl_uint num_events_in_wait_list, + const cl_event *event_wait_list, + cl_event *event); + +cl_int clGetSemaphoreInfoKHR( + cl_semaphore_khr sema_object, + cl_semaphore_info_khr param_name, + size_t param_value_size, + void *param_value, + size_t *param_value_size_ret); + +cl_int clReleaseSemaphoreKHR(cl_semaphore_khr sema_object); + +cl_int clRetainSemaphoreKHR(cl_semaphore_khr sema_object); +---- + +=== New API Enums + +Accepted value for the _param_name_ parameter to {clGetPlatformInfo} to query the semaphore types supported by an OpenCL platform: + +[source] +---- +CL_PLATFORM_SEMAPHORE_TYPES_KHR 0x2036 +---- + +Accepted value for the _param_name_ parameter to {clGetDeviceInfo} to query the semaphore types supported by an OpenCL device: + +[source] +---- +CL_DEVICE_SEMAPHORE_TYPES_KHR 0x204C +---- + +Semaphore types: + +[source] +---- +CL_SEMAPHORE_TYPE_BINARY_KHR 1 +---- + +New attributes that can be passed as part of {cl_semaphore_info_khr_TYPE}: + +[source] +---- +CL_SEMAPHORE_CONTEXT_KHR 0x2039 +CL_SEMAPHORE_REFERENCE_COUNT_KHR 0x203A +CL_SEMAPHORE_PROPERTIES_KHR 0x203B +CL_SEMAPHORE_PAYLOAD_KHR 0x203C +---- + +// TODO: We don't need an enum assigned for CL_DEVICE_HANDLE_LIST_END_KHR and should just use 0. +// TODO: Do we need to define CL_DEVICE_HANDLE_LIST here or should it be in the external semaphore spec instead? + +New attributes that can be passed as part of {cl_semaphore_info_khr_TYPE} or {cl_semaphore_properties_khr_TYPE}: + +[source] +---- +CL_SEMAPHORE_TYPE_KHR 0x203D +CL_DEVICE_HANDLE_LIST_KHR 0x2051 +CL_DEVICE_HANDLE_LIST_END_KHR 0 +---- + +New return values from {clGetEventInfo} when _param_name_ is {CL_EVENT_COMMAND_TYPE}: + +[source] +---- +CL_COMMAND_SEMAPHORE_WAIT_KHR 0x2042 +CL_COMMAND_SEMAPHORE_SIGNAL_KHR 0x2043 +---- + +// TODO: This error code is already allocated. Use -1141 instead? + +The following error codes can be returned by APIs introduced as part of this specification or the specifications that depend on this: +[source] +---- +CL_INVALID_SEMAPHORE_KHR -1142 +---- + +=== Modifications to existing APIs added by this spec + +Following new enums are added to the list of supported _param_names_ by {clGetPlatformInfo}: + +.List of supported param_names by clGetPlatformInfo +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Platform Info | Return Type | Description +| {CL_PLATFORM_SEMAPHORE_TYPES_KHR} + | {cl_semaphore_type_khr_TYPE}[] + | Returns the list of the semaphore types supported all devices in _platform_. +|==== + +{clGetPlatformInfo} when called with _param_name_ {CL_PLATFORM_SEMAPHORE_TYPES_KHR} must return common list of semaphore types supported by all devices in the platform. + +Following new enums are added to the list of supported _param_names_ by {clGetDeviceInfo}: + +.List of supported param_names by clGetDeviceInfo +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Device Info | Return Type | Description +| {CL_DEVICE_SEMAPHORE_TYPES_KHR} + | {cl_semaphore_type_khr_TYPE}[] + | Returns the list of the semaphore types supported by _device_. +|==== + +{clGetDeviceInfo} when called with param_name {CL_DEVICE_SEMAPHORE_TYPES_KHR} must return a non-empty list of semaphore types for at least one of the devices in the platform. +The results of this query should meet minimum requirements for {cl_semaphore_type_khr_TYPE} as described by <>. + +=== Description of new types added by this spec + +Following new types are added: + +* {cl_semaphore_type_khr_TYPE} to represent the different types of semaphores. + ** It is mandatory to support {CL_SEMAPHORE_TYPE_BINARY_KHR}. + +* {cl_semaphore_properties_khr_TYPE} to represent properties associated with semaphores. + ** {CL_SEMAPHORE_TYPE_KHR} must be supported. + +* {cl_semaphore_info_khr_TYPE} to represent queries to get additional information about semaphores. + ** All enums described in New API Enums for {cl_semaphore_info_khr_TYPE} must be supported. + +* {cl_semaphore_payload_khr_TYPE} to represent payload values of semaphores. + +* {cl_semaphore_khr_TYPE} to represent semaphore objects. + +Note that above types can be extended in future based on the need for additional types of semaphore and properties required by them. +The specifics of the same can be added as a newer version of this specification or by a separate specification that depends on this for basic semaphore support. + +=== Description of new APIs added by this spec + +The following new APIs are added as part of this spec. The details of each are described below: + +==== Creating semaphores + +A *semaphore object* may be created using the function + +include::{generated}/api/protos/clCreateSemaphoreWithPropertiesKHR.txt[] + +_context_ identifies a valid OpenCL context that the created {cl_semaphore_khr_TYPE} will belong to. + +// TODO: Do we want the same "all devices in the context" behavior if CL_DEVICE_HANDLE_LIST_KHR is not specified? + +_sema_props_ specifies additional semaphore properties in the form list of pairs terminated with 0. +{CL_SEMAPHORE_TYPE_KHR} must be part of the list of properties specified by _sema_props_. + +Following new properties are added to the list of possible supported properties by cl_semaphore_properties_khr that can be passed to {clCreateSemaphoreWithPropertiesKHR}: + +.List of supported semaphore creation properties by clCreateSemaphoreWithPropertiesKHR +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Semaphore Property | Property Value | Description +| {CL_SEMAPHORE_TYPE_KHR} + | {cl_semaphore_type_khr_TYPE} + | Specifies the type of semaphore to create. +| {CL_DEVICE_HANDLE_LIST_KHR} + | {cl_device_id_TYPE}[] + | Specifies the list of OpenCL devices (terminated with {CL_DEVICE_HANDLE_LIST_END_KHR}) to associate with the semaphore. +|==== + +If {CL_DEVICE_HANDLE_LIST_KHR} is not specified as part of _sema_props_, the semaphore object created by {clCreateSemaphoreWithPropertiesKHR} is by default accessible to all devices in the _context_. + +_errcode_ret_ returns an appropriate error code. If _errcode_ret_ is `NULL`, no error code is returned. + +{clCreateSemaphoreWithPropertiesKHR} returns a valid semaphore object in an un-signaled state and and _errcode_ret_ is set to {CL_SUCCESS} if the function is executed successfully. +Otherwise, it returns a `NULL` value with one of the following error values returned in _errcode_ret_: + +* {CL_INVALID_CONTEXT} if _context_ is not a valid context. +* {CL_INVALID_PROPERTY} if a property name in _sema_props_ is not a supported property name, if the value specified for a supported property name is not valid, or if the same property name is specified more than once. +* {CL_INVALID_DEVICE} if {CL_DEVICE_HANDLE_LIST_KHR} is specified as part of _sema_props_, but it does not identify a valid device or if a device identified by {CL_DEVICE_HANDLE_LIST_KHR} is not one of the devices within _context_. +* {CL_INVALID_VALUE} +** if _sema_props_ is `NULL`, or +** if _sema_props_ do not specify pairs for minimum set of properties (i.e. {CL_SEMAPHORE_TYPE_KHR}) required for successful creation of a {cl_semaphore_khr_TYPE}, or +** if one or more properties and/or their values specified by _sema_props_ are not valid. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +==== Waiting on and signaling semaphores + +To wait on a set of semaphores, call the function + +include::{generated}/api/protos/clEnqueueWaitSemaphoresKHR.txt[] + +_command_queue_ specifies a valid command-queue. + +_num_sema_objects_ specifies the number of semaphore objects to wait on. + +_sema_objects_ points to the list of semaphore objects to wait on. +The length of the list must be at least _num_sema_objects_. + +_sema_payload_list_ points to the list of values of type {cl_semaphore_payload_khr_TYPE} containing valid semaphore payload values to wait on. +This can be set to `NULL` or will be ignored when all semaphores in the list of _sema_objects_ are of type {CL_SEMAPHORE_TYPE_BINARY_KHR}. + +_num_events_in_wait_list_ specifies the number of events in _event_wait_list_. + +_event_wait_list_ specifies list of events that need to complete before {clEnqueueWaitSemaphoresKHR} can be executed. +If _event_wait_list_ is `NULL`, then {clEnqueueWaitSemaphoresKHR} does not wait on any event to complete. +If _event_wait_list_ is `NULL`, _num_events_in_wait_list_ must be 0. +If _event_wait_list_ is not `NULL`, the list of events pointed to by _event_wait_list_ must be valid and _num_events_in_wait_list_ must be greater than 0. +The events specified in _event_wait_list_ act as synchronization points. +The context associated with events in _event_wait_list_ and that associated with _command_queue_ must be the same. + +_event_ returns an event object that identifies this particular command and can be used to query or queue a wait for this particular command to complete. +_event_ can be `NULL` in which case it will not be possible for the application to query the status of this command or queue a wait for this command to complete. + +The successful completion of the event generated by {clEnqueueWaitSemaphoresKHR} called on one or more semaphore objects of type {CL_SEMAPHORE_TYPE_BINARY_KHR} leads to un-signaling the corresponding semaphore objects and the state of these semaphore objects will be reset. Any subsequent {clEnqueueWaitSemaphoresKHR} operation on these semaphores without a pending {clEnqueueSignalSemaphoresKHR} may lead to implementation-defined behavior. + +{clEnqueueWaitSemaphoresKHR} returns {CL_SUCCESS} if the function is executed successfully. +Otherwise, it returns one of the following errors: + +* {CL_INVALID_COMMAND_QUEUE} +** if _command_queue_ is not a valid command-queue, or +** if the device associated with _command_queue_ is not same as one of the devices specified by {CL_DEVICE_HANDLE_LIST_KHR} at the time of creating one or more of _sema_objects_, or +** if one or more of _sema_objects_ belong to a context that does not contain a device associated with_command_queue_, or +** if one or more of _sema_objects_ can not be shared with the device associated with _command_queue_. +* {CL_INVALID_VALUE} if _num_sema_objects_ is 0. +* {CL_INVALID_SEMAPHORE_KHR} if any of the semaphore objects specified by _sema_objects_ is not valid. +* {CL_INVALID_CONTEXT} if the context associated with _command_queue_ and any of the semaphore objects in _sema_objects_ are not the same or if the context associated with _command_queue_ and that associated with events in _event_wait_list_ are not the same. +* {CL_INVALID_VALUE} if any of the semaphore objects specified by _sema_objects_ requires a semaphore payload and _sema_payload_list_ is `NULL`. +* {CL_INVALID_EVENT_WAIT_LIST} + ** if _event_wait_list_ is `NULL` and _num_events_in_wait_list_ is not 0, or + ** if _event_wait_list_ is not `NULL` and _num_events_in_wait_list_ is 0, or + ** if event objects in _event_wait_list_ are not valid events. +* {CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST} if the execution status of any of the events in _event_wait_list_ is a negative integer value. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +To signal a set of semaphores, call the function + +include::{generated}/api/protos/clEnqueueSignalSemaphoresKHR.txt[] + +_command_queue_ specifies a valid command-queue. + +_num_sema_objects_ specifies the number of semaphore objects to signal. + +_sema_objects_ points to the list of semaphore objects to signal. +The length of the list must be at least _num_sema_objects_. + +_sema_payload_list_ points to the list of values of type {cl_semaphore_payload_khr_TYPE} containing semaphore payload values to signal. +This can be set to `NULL` or will be ignored when all semaphores in the list of _sema_objects_ are of type {CL_SEMAPHORE_TYPE_BINARY_KHR}. + +_num_events_in_wait_list_ specifies the number of events in event_wait_list. + +_event_wait_list_ points to the list of events that need to complete before {clEnqueueSignalSemaphoresKHR} can be executed. +If _event_wait_list_ is `NULL`, then {clEnqueueSignalSemaphoresKHR} does not wait on any event to complete. +If _event_wait_list_ is `NULL`, _num_events_in_wait_list_ must be 0. +If _event_wait_list_ is not `NULL`, the list of events pointed to by _event_wait_list_ must be valid and +_num_events_in_wait_list_ must be greater than 0. +The events specified in _event_wait_list_ act as synchronization points. +The context associated with events in _event_wait_list_ and that associated with _command_queue_ must be the same. + +_event_ returns an event object that identifies this particular command and can be used to query or queue a wait for this particular command to complete. +_event_ can be `NULL` in which case it will not be possible for the application to query the status of this command +or queue a wait for this command to complete. + +The successful completion of event generated by {clEnqueueSignalSemaphoresKHR} called on one or more semaphore objects of type {CL_SEMAPHORE_TYPE_BINARY_KHR} changes the state of the corresponding semaphore objects to signaled. Any subsequent {clEnqueueSignalSemaphoresKHR} operation on these semaphores without a pending {clEnqueueWaitSemaphoresKHR} may lead to implementation-defined behavior. + +{clEnqueueSignalSemaphoresKHR} returns {CL_SUCCESS} if the function is executed successfully. +Otherwise, it returns one of the following errors: + +* {CL_INVALID_COMMAND_QUEUE} +** if _command_queue_ is not a valid command-queue, or +** if device associated with _command_queue_ is not same as one of devices specified by {CL_DEVICE_HANDLE_LIST_KHR} at the time of creating one or more of _sema_objects_, or +** if one or more of _sema_objects_ belong to a context that does not contain a device associated _command_queue_, or +** if one or more of _sema_objects_ can not be shared with device associated with _command_queue_. +* {CL_INVALID_VALUE} if _num_sema_objects_ is 0 +* {CL_INVALID_SEMAPHORE_KHR} if any of the semaphore objects specified by _sema_objects_ is not valid. +* {CL_INVALID_CONTEXT} if the context associated with _command_queue_ and any of the semaphore objects in _sema_objects_ are not the same or if the context associated with _command_queue_ and that associated with events in _event_wait_list_ are not the same. +* {CL_INVALID_VALUE} if any of the semaphore objects specified by _sema_objects_ requires a semaphore payload and _sema_payload_list_ is `NULL`. +* {CL_INVALID_EVENT_WAIT_LIST} + ** if _event_wait_list_ is `NULL` and _num_events_in_wait_list_ is not 0, or + ** if _event_wait_list_ is not `NULL` and _num_events_in_wait_list_ is 0, or + ** if event objects in _event_wait_list_ are not valid events. +* {CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST} if the execution status of any of the events in _event_wait_list_ is a negative integer value. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +==== Semaphore Queries + +To query information about a semaphore object, call the function + +include::{generated}/api/protos/clGetSemaphoreInfoKHR.txt[] + +_sema_object_ specifies the semaphore object being queried. + +_param_name_ is a constant that specifies the semaphore information to query, and must be one of the values shown in <>. + +_param_value_ is a pointer to memory where the result of the query is returned as described in <>. If _param_value_ is `NULL`, it is ignored. + +_param_value_size_ specifies the size in bytes of memory pointed to _param_value_. This size must be greater than or equal to the size of the return type described in table <>. + +_param_value_size_ret_ returns the actual size in bytes of data +being queried by _param_value_. If _param_value_size_ret_ is `NULL`, it is ignored. + +[[cl_khr_semaphore_info-table]] +.List of supported param_names by clGetSemaphoreInfoKHR +[width="100%",cols="<33%,<17%,<50%",options="header"] +|==== +| Semaphore Info | Return Type | Description +| {CL_SEMAPHORE_CONTEXT_KHR} + | {cl_context_TYPE} + | Returns the context specified when the semaphore is created. + +| {CL_SEMAPHORE_REFERENCE_COUNT_KHR} + | {cl_uint_TYPE} + | Returns the semaphore reference count. + +| {CL_SEMAPHORE_PROPERTIES_KHR} + | {cl_semaphore_properties_khr_TYPE}[] + | Return the properties argument specified in + {clCreateSemaphoreWithPropertiesKHR}. + + The implementation must return the values specified in the properties + argument in the same order and without including additional properties. + +| {CL_SEMAPHORE_TYPE_KHR} + | {cl_semaphore_type_khr_TYPE} + | Returns the semaphore type. + +| {CL_SEMAPHORE_PAYLOAD_KHR} + | {cl_semaphore_payload_khr_TYPE} + | Returns the semaphore payload value. For semaphores of type + {CL_SEMAPHORE_TYPE_BINARY_KHR}, the payload value returned will be `0` + if the semaphore is in an un-signaled state and `1` if it is in a + signaled state. + +| {CL_DEVICE_HANDLE_LIST_KHR} + | {cl_device_id_TYPE}[] + | Returns the list of OpenCL devices the semaphore is associated with. +|==== + +{clGetSemaphoreInfoKHR} returns {CL_SUCCESS} if the information is queried successfully. +Otherwise, it returns one of the following errors: + +* {CL_INVALID_SEMAPHORE_KHR} +** if _sema_object_ is not a valid semaphore +* {CL_INVALID_VALUE} +** if _param_name_ is not one of the attribute defined in table <> or +** if _param_value_size is less than the size of Return Type of the corresponding _param_name_ attribute as defined in table <>. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +==== Retaining and Releasing Semaphores + +To release a semaphore object, call the function + +include::{generated}/api/protos/clReleaseSemaphoreKHR.txt[] + +_sema_object_ specifies the semaphore object to be released. + +The _sema_object_ reference count is decremented. + +{clReleaseSemaphoreKHR} returns {CL_SUCCESS} if the function is executed successfully. +Otherwise, it returns one of the following errors: + +* {CL_INVALID_SEMAPHORE_KHR} if _sema_object_ is not a valid semaphore object. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +After the reference count becomes zero and commands queued for execution on a +command-queue(s) that use _sema_object_ have finished, the semaphore object is +deleted. +Using this function to release a reference that was not obtained by creating the +object via {clCreateSemaphoreWithPropertiesKHR} or by calling +{clRetainSemaphoreKHR} causes undefined behavior. + +To retain a semaphore object, call the function + +include::{generated}/api/protos/clRetainSemaphoreKHR.txt[] + +_sema_object_ specifies the semaphore object to be retained. + +increments the reference count of _sema_object_. + +{clRetainSemaphoreKHR} returns {CL_SUCCESS} if the function is executed successfully. +Otherwise, it returns one of the following errors: + +* {CL_INVALID_SEMAPHORE_KHR} if _sema_object_ is not a valid semaphore object. +* {CL_OUT_OF_RESOURCES} if there is a failure to allocate resources required by the OpenCL implementation on the device. +* {CL_OUT_OF_HOST_MEMORY} if there is a failure to allocate resources required by the OpenCL implementation on the host. + +[[cl_khr_semaphore-Sample-Code]] +=== Sample Code + +. Example for semaphore creation in a single device context ++ +-- +[source] +---- +// Get cl_devices of the platform. +clGetDeviceIDs(..., &devices, &deviceCount); + +// Create cl_context with just first device +context = clCreateContext(..., 1, devices, ...); + +// Create clSema of type cl_semaphore_khr usable on single device in the context + +cl_semaphore_properties_khr sema_props[] = + {(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR, + 0}; + +int errcode_ret = 0; + +cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context, + sema_props, + &errcode_ret); +---- +-- + +. Example for semaphore creation for a single device in a multi-device context ++ +-- +[source] +---- +// Get cl_devices of the platform. +clGetDeviceIDs(..., &devices, &deviceCount); + +// Create cl_context with first two devices +clCreateContext(..., 2, devices, ...); + +// Create clSema of type cl_semaphore_khr usable only on devices[0] + +cl_semaphore_properties_khr sema_props[] = + {(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR, + (cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR, + (cl_semaphore_properties_khr)CL_DEVICE_HANDLE_LIST_KHR, + (cl_semaphore_properties_khr)devices[0], CL_DEVICE_HANDLE_LIST_END_KHR, + 0}; + +int errcode_ret = 0; + +cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context, + sema_props, + &errcode_ret); +---- +-- +. Example for synchronization using Wait and Signal ++ +-- +[source] +---- +// clSema is created using clCreateSemaphoreWithPropertiesKHR +// using one of the examples for semaphore creation. + +cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context, + sema_props, + &errcode_ret); + +// Start the main loop + +while (true) { + // (not shown) Signal the semaphore from other work + + // Wait for the semaphore in OpenCL + // by calling clEnqueueWaitSemaphoresKHR on 'clSema' + clEnqueueWaitSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &clSema, + /*sema_payload_list*/ NULL, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Launch kernel that accesses extMem + clEnqueueNDRangeKernel(command_queue, ...); + + // Signal the semaphore in OpenCL + clEnqueueSignalSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &clSema, + /*sema_payload_list*/ NULL, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // (not shown) Launch other work that waits on 'clSema' +} +---- +-- +. Example for {clGetSemaphoreInfoKHR} ++ +-- +[source] +---- +// clSema is created using clCreateSemaphoreWithPropertiesKHR +// using one of the examples for semaphore creation. + +cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context, + sema_props, + &errcode_ret); + +// Start the main rendering loop + +while (true) { + // (not shown) Signal the semaphore from other work + + // Wait for the semaphore in OpenCL, by calling clEnqueueWaitSemaphoresKHR on 'clSema' + clEnqueueWaitSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &clSema, + /*sema_payload_list*/ NULL, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Launch kernel in OpenCL + clEnqueueNDRangeKernel(command_queue, ...); + + // Signal the semaphore in OpenCL + clEnqueueSignalSemaphoresKHR(/*command_queue*/ command_queue, + /*num_sema_objects*/ 1, + /*sema_objects*/ &clSema, + /*sema_payload_list*/ NULL, + /*num_events_in_wait_list*/ 0, + /*event_wait_list*/ NULL, + /*event*/ NULL); + + // Query type of clSema + clGetSemaphoreInfoKHR(/*sema_object*/ clSema, + /*param_name*/ CL_SEMAPHORE_TYPE_KHR, + /*param_value_size*/ sizeof(cl_semaphore_type_khr), + /*param_value*/ &clSemaType, + /*param_value_ret_size*/ &clSemaTypeSize); + + if (clSemaType == CL_SEMAPHORE_TYPE_BINARY_KHR) { + // Do something + } + else { + // Do something else + } + // (not shown) Launch other work that waits on 'clSema' +} +---- +-- diff --git a/xml/cl.xml b/xml/cl.xml index a945b5c0..5dc4fd8c 100644 --- a/xml/cl.xml +++ b/xml/cl.xml @@ -155,6 +155,7 @@ server's OpenCL/api-docs repository. typedef struct _cl_kernel * cl_kernel; typedef struct _cl_event * cl_event; typedef struct _cl_sampler * cl_sampler; + typedef struct _cl_semaphore_khr * cl_semaphore_khr; typedef cl_uint cl_bool; typedef cl_ulong cl_bitfield; typedef cl_ulong cl_properties; @@ -227,6 +228,12 @@ server's OpenCL/api-docs repository. typedef cl_bitfield cl_command_queue_capabilities_intel; typedef cl_bitfield cl_device_feature_capabilities_intel; typedef cl_bitfield cl_device_integer_dot_product_capabilities_khr; + typedef cl_properties cl_semaphore_properties_khr; + typedef cl_uint cl_semaphore_info_khr; + typedef cl_uint cl_semaphore_type_khr; + typedef cl_ulong cl_semaphore_payload_khr; + typedef cl_uint cl_external_semaphore_handle_type_khr; + typedef cl_uint cl_external_memory_handle_type_khr; Structure types @@ -621,7 +628,11 @@ server's OpenCL/api-docs repository. - + + + + + @@ -639,6 +650,8 @@ server's OpenCL/api-docs repository. + + @@ -1207,6 +1220,12 @@ server's OpenCL/api-docs repository. + + + + + + In order to synchronize vendor IDs across Khronos APIs, Vulkan's vk.xml @@ -1635,7 +1654,40 @@ server's OpenCL/api-docs repository. - + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + @@ -2380,6 +2432,75 @@ server's OpenCL/api-docs repository. const size_t* global_work_size size_t* suggested_local_work_size + + cl_semaphore_khr clCreateSemaphoreWithPropertiesKHR + cl_context context + const cl_semaphore_properties_khr* sema_props + cl_int* errcode_ret + + + cl_int clEnqueueWaitSemaphoresKHR + cl_command_queue command_queue + cl_uint num_sema_objects + const cl_semaphore_khr* sema_objects + const cl_semaphore_payload_khr* sema_payload_list + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSignalSemaphoresKHR + cl_command_queue command_queue + cl_uint num_sema_objects + const cl_semaphore_khr* sema_objects + const cl_semaphore_payload_khr* sema_payload_list + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetSemaphoreInfoKHR + cl_semaphore_khr sema_object + cl_semaphore_info_khr param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clReleaseSemaphoreKHR + cl_semaphore_khr sema_object + + + cl_int clRetainSemaphoreKHR + cl_semaphore_khr sema_object + + + cl_int clGetSemaphoreHandleForTypeKHR + cl_semaphore_khr sema_object + cl_device_id device + cl_external_semaphore_handle_type_khr handle_type + size_t handle_size + void* handle_ptr + size_t* handle_size_ret + + + cl_int clEnqueueAcquireExternalMemObjectsKHR + cl_command_queue command_queue + cl_uint num_mem_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseExternalMemObjectsKHR + cl_command_queue command_queue + cl_uint num_mem_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + cl_mem clImportMemoryARM cl_context context @@ -6273,6 +6394,172 @@ server's OpenCL/api-docs repository. + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + +