### OpenCL Program Setup and Kernel Argument Setting

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_command_buffer.html

Demonstrates the basic setup for an OpenCL program, including context creation, program compilation, kernel creation, and setting kernel arguments. This is a foundational example for executing OpenCL kernels.

```c++
int main() {
  cl_platform_id platform;
  CL_CHECK(clGetPlatformIDs(1, &platform, nullptr));
  cl_device_id device;
  CL_CHECK(clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 1, &device, nullptr));

  cl_int error;
  cl_context context =
      clCreateContext(nullptr, 1, &device, nullptr, nullptr, &error);
  CL_CHECK(error);

  const char* code = R"OpenCLC(
      kernel void vector_addition(global int* tile1, global int* tile2,
                                  global int* res) {
        size_t index = get_global_id(0);
        res[index] = tile1[index] + tile2[index];
      }
      )OpenCLC";
  const size_t length = std::strlen(code);

  cl_program program =
      clCreateProgramWithSource(context, 1, &code, &length, &error);
  CL_CHECK(error);

  CL_CHECK(clBuildProgram(program, 1, &device, nullptr, nullptr, nullptr));

  cl_kernel kernel = clCreateKernel(program, "vector_addition", &error);
  CL_CHECK(error);

  constexpr size_t frame_count = 60;
  constexpr size_t frame_elements = 1024;
  constexpr size_t frame_size = frame_elements * sizeof(cl_int);

  constexpr size_t tile_count = 16;
  constexpr size_t tile_elements = frame_elements / tile_count;
  constexpr size_t tile_size = tile_elements * sizeof(cl_int);

  cl_mem buffer_tile1 =
      clCreateBuffer(context, CL_MEM_READ_ONLY, tile_size, nullptr, &error);
  CL_CHECK(error);
  cl_mem buffer_tile2 =
      clCreateBuffer(context, CL_MEM_READ_ONLY, tile_size, nullptr, &error);
  CL_CHECK(error);
  cl_mem buffer_res =
      clCreateBuffer(context, CL_MEM_WRITE_ONLY, tile_size, nullptr, &error);
  CL_CHECK(error);

  CL_CHECK(clSetKernelArg(kernel, 0, sizeof(buffer_tile1), &buffer_tile1));
  CL_CHECK(clSetKernelArg(kernel, 1, sizeof(buffer_tile2), &buffer_tile2));
  CL_CHECK(clSetKernelArg(kernel, 2, sizeof(buffer_res), &buffer_res));

```

--------------------------------

### CL_DEVICE_PARTITION_EQUALLY_EXT Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/ext/cl_ext_device_fission.txt

Example of partitioning a device equally into sub-devices, each with a specified number of compute units.

```c
{ CL_DEVICE_PARTITION_EQUALLY_EXT, 8,
  CL_PROPERTIES_LIST_END_EXT }
```

--------------------------------

### CL_DEVICE_PARTITION_BY_COUNTS_EXT Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/ext/cl_ext_device_fission.txt

Example of partitioning a device into sub-devices with specific compute unit counts.

```c
{ CL_DEVICE_PARTITION_BY_COUNTS_EXT,
  2, 2, CL_PARTITION_BY_COUNTS_LIST_END_EXT,
  CL_PROPERTIES_LIST_END_EXT }
```

--------------------------------

### OpenCL Basic Setup and Kernel Compilation

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

Demonstrates basic OpenCL setup including platform and device identification, context creation, and program compilation from source. Requires OpenCL headers and a C++ compiler.

```c++
int main() {
  cl_platform_id platform;
  CL_CHECK(clGetPlatformIDs(1, &platform, nullptr));
  cl_device_id device;
  CL_CHECK(clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 1, &device, nullptr));

  cl_int error;
  cl_context context =
      clCreateContext(nullptr, 1, &device, nullptr, nullptr, &error);
  CL_CHECK(error);

  const char* code = R"OpenCLC(\n
      kernel void vector_addition(global int* tile1, global int* tile2,
                                  global int* res) {
        size_t index = get_global_id(0);
        res[index] = tile1[index] + tile2[index];
      }
      )OpenCLC";
  const size_t length = std::strlen(code);

  cl_program program =
      clCreateProgramWithSource(context, 1, &code, &length, &error);
  CL_CHECK(error);

```

--------------------------------

### clGetSemaphoreInfoKHR Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_semaphore.html

This example demonstrates how to query information about a semaphore object, specifically its type.

```APIDOC
## clGetSemaphoreInfoKHR

### Description
Queries information about a semaphore object. This function can be used to retrieve various properties of a semaphore, such as its type.

### Method
`clGetSemaphoreInfoKHR`

### Parameters
#### Semaphore Object
- **sema_object** (cl_semaphore_khr) - The semaphore object to query.
- **param_name** (cl_semaphore_info) - The information to retrieve. For example, `CL_SEMAPHORE_TYPE_KHR` to get the semaphore type.
- **param_value_size** (size_t) - The size of the buffer pointed to by `param_value`.
- **param_value** (void *) - A pointer to a buffer where the retrieved information will be stored.
- **param_value_ret_size** (size_t *) - A pointer to a `size_t` that stores the actual size of the data returned in `param_value`. Can be NULL.

### Request Example
```c
cl_semaphore_type_khr clSemaType;
size_t clSemaTypeSize;

clGetSemaphoreInfoKHR(clSema, CL_SEMAPHORE_TYPE_KHR, sizeof(cl_semaphore_type_khr), &clSemaType, &clSemaTypeSize);
```

### Response
#### Success Response
- **param_value** (cl_semaphore_type_khr) - The type of the semaphore (e.g., `CL_SEMAPHORE_TYPE_BINARY_KHR`).
- **param_value_ret_size** - The size of the returned semaphore type.
```

--------------------------------

### cl_khr_create_command_queue Extension Header Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

Example header definitions for the cl_khr_create_command_queue extension, demonstrating the application of the generic conventions. Includes type definitions and function prototypes.

```c
#define cl_khr_create_command_queue 1
#define CL_KHR_CREATE_COMMAND_QUEUE_EXTENSION_NAME \
    "cl_khr_create_command_queue"

#define CL_KHR_CREATE_COMMAND_QUEUE_EXTENSION_VERSION CL_MAKE_VERSION(1, 0, 0)

typedef cl_properties       cl_queue_properties_khr;

typedef cl_command_queue CL_API_CALL
clCreateCommandQueueWithPropertiesKHR_t(
    cl_context context,
    cl_device_id device,
    const cl_queue_properties_khr* properties,
    cl_int* errcode_ret);

typedef clCreateCommandQueueWithPropertiesKHR_t *
clCreateCommandQueueWithPropertiesKHR_fn CL_API_SUFFIX__VERSION_1_2;

#if !defined(CL_NO_NON_ICD_DISPATCH_EXTENSION_PROTOTYPES) 

extern CL_API_ENTRY cl_command_queue CL_API_CALL
clCreateCommandQueueWithPropertiesKHR(
    cl_context context,
    cl_device_id device,
    const cl_queue_properties_khr* properties,
    cl_int* errcode_ret) CL_API_SUFFIX__VERSION_1_2;

#endif /* !defined(CL_NO_NON_ICD_DISPATCH_EXTENSION_PROTOTYPES) */
```

--------------------------------

### Device Information Queries (Examples)

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/clGetDeviceInfo.html

Examples of common device information parameters that can be queried using clGetDeviceInfo.

```APIDOC
## Supported Device Info Parameters (Examples)

### `CL_DEVICE_TYPE`
*   **Description**: The type of the OpenCL device.
*   **Return Type**: `cl_device_type`

### `CL_DEVICE_VENDOR_ID`
*   **Description**: A unique device vendor identifier.
*   **Return Type**: `cl_uint`

### `CL_DEVICE_MAX_COMPUTE_UNITS`
*   **Description**: The number of parallel compute units on the OpenCL device.
*   **Return Type**: `cl_uint`

### `CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS`
*   **Description**: Maximum dimensions that specify the global and local work-item IDs.
*   **Return Type**: `cl_uint`

### `CL_DEVICE_MAX_WORK_GROUP_SIZES`
*   **Description**: The maximum number of work-items that can be specified in each dimension of a work-group.
*   **Return Type**: `size_t[]`

### `CL_DEVICE_MAX_WORK_GROUP_SIZE`
*   **Description**: The maximum total number of work-items that can be specified for a work-group.
*   **Return Type**: `size_t`

### `CL_DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT`
*   **Description**: Preferred vector width for `float` data types.
*   **Return Type**: `cl_uint`
```

--------------------------------

### Example Partitioning Properties

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/intel/cl_intel_device_partition_by_names.txt

Demonstrates the properties array used to define a sub-device with specific compute units.

```c
{
    CL_DEVICE_PARTITION_BY_NAMES_INTEL, 0, 1, 3, CL_PARTITION_BY_NAMES_LIST_END_INTEL, CL_PROPERTIES_LIST_END
}
```

--------------------------------

### clEnqueueSignalSemaphoresKHR Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_semaphore.html

This example demonstrates how to enqueue a signal operation on a semaphore in OpenCL.

```APIDOC
## clEnqueueSignalSemaphoresKHR

### Description
Enqueues a signal operation on one or more semaphores. This command will signal the specified semaphores after it completes.

### Method
`clEnqueueSignalSemaphoresKHR`

### Parameters
#### Command Queue
- **command_queue** (cl_command_queue) - The command queue to enqueue the signal command into.

#### Semaphore Objects
- **num_sema_objects** (cl_uint) - The number of semaphore objects in `sema_objects`.
- **sema_objects** (const cl_semaphore_khr *) - A list of semaphore objects to signal.
- **sema_payload_list** (const cl_semaphore_payload_khr *) - A list of payloads corresponding to the semaphores. Can be NULL if not needed.

#### Event Wait List
- **num_events_in_wait_list** (cl_uint) - The number of events in `event_wait_list`.
- **event_wait_list** (const cl_event *) - A list of events that must complete before this command can execute.

#### Event
- **event** (cl_event *) - Returns an event object that identifies this particular command in the command queue. Can be NULL.

### Request Example
```c
cl_semaphore_khr clSema;
cl_command_queue command_queue;

clEnqueueSignalSemaphoresKHR(command_queue, 1, &clSema, NULL, 0, NULL, NULL);
```

### Response
#### Success Response
- **event** (cl_event) - An event object representing the enqueued signal operation.
```

--------------------------------

### Synchronization Example Using an OpenCL Exported Semaphore

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_external_semaphore.html

This example demonstrates how to create, export, import, and use an external semaphore for synchronization between OpenCL operations and potentially other APIs or processes.

```APIDOC
## Example for Synchronization Using a Semaphore Exported by OpenCL

This example illustrates the workflow for using external semaphores with OpenCL.

### Steps:

1.  **Get OpenCL Devices and Create Context**: Obtain device IDs and create an OpenCL context.
    ```c
    // Get cl_devices of the platform.
    clGetDeviceIDs(..., &devices, &deviceCount);
    
    // Create cl_context with first two devices
    clCreateContext(..., 2, devices, ...);
    ```

2.  **Create External Semaphore**: Create a semaphore with specific properties, including its type and export handle types.
    ```c
    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_SEMAPHORE_DEVICE_HANDLE_LIST_KHR,
        (cl_semaphore_properties_khr)devices[1],
        CL_SEMAPHORE_DEVICE_HANDLE_LIST_END_KHR,
        0
    };
    
    int errcode_ret = 0;
    cl_semaphore_khr clSema = clCreateSemaphoreWithPropertiesKHR(context,
                                                                 sema_props,
                                                                 &errcode_ret);
    ```

3.  **Query Semaphore Handle Type**: Determine the type of handle associated with the semaphore.
    ```c
    cl_external_semaphore_handle_type_khr handle_type;
    size_t handle_type_size;
    clGetSemaphoreInfoKHR(clSema,
                          CL_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR,
                          sizeof(cl_external_semaphore_handle_type_khr),
                          &handle_type,
                          &handle_type_size);
    ```

4.  **Get Exported Semaphore Handle**: Retrieve the actual handle (e.g., a file descriptor) for the semaphore.
    ```c
    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);
    }
    ```

5.  **Synchronization Loop**: Use the semaphore for synchronization within a rendering loop.
    ```c
    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'
    }
    ```

### Notes:

*   This example assumes the existence of `context`, `devices`, `command_queue`, and `device` variables.
*   The `cl_semaphore_properties_khr` array defines the properties of the semaphore, including its type and the types of handles that can be exported.
*   `CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR` indicates that the semaphore handle is an opaque file descriptor.
*   The `clEnqueueWaitSemaphoresKHR` and `clEnqueueSignalSemaphoresKHR` functions are used to wait for and signal the semaphore, respectively.
```

--------------------------------

### Partitioning by Equal Compute Units

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/clCreateSubDevices.html

Example of partitioning a device into two sub-devices, each with an equal number of compute units.

```APIDOC
{
    CL_DEVICE_PARTITION_EQUALLY, 8,
    0
}
```

--------------------------------

### Get MCE Packed BMVs Result

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/intel/cl_intel_device_side_avc_motion_estimation.txt

Retrieves the Motion Compensation Estimation (MCE) packed Block Motion Vectors (BMVs) result. Up to 16 packed BMVs are returned, one per work-item. The validity of forward and backward packed MVs depends on the search operation's payload setup.

```APIDOC
## intel_sub_group_avc_mce_get_motion_vectors

### Description
Get the MCE packed BMVs result. Up to 16 packed BMVs are returned, one per work-item. If the MCE search operation's payload was setup for unidirectional search then only the forward packed MV will be valid in each BMV, otherwise both packed MVs will be valid. The BMVs have to be selected by their respective work-items based on the result block major and minor shapes.

### Parameters
#### Path Parameters
- **result** (intel_sub_group_avc_mce_result_t) - Description: The MCE result structure containing the packed BMVs.

### Returns
- **ulong** - The packed BMVs. The structure of the returned BMVs depends on the major shape:
  - 16x16: One BMV returned by work-item 0.
  - 16x8 or 8x16: Two BMVs returned by work-items 0 and 8.
  - 8x8: Four sets of BMVs returned.

### Method
(Implicitly a function call within OpenCL kernel)
```

--------------------------------

### clEnqueueWaitSemaphoresKHR Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_semaphore.html

This example shows how to enqueue a wait operation on a semaphore in OpenCL.

```APIDOC
## clEnqueueWaitSemaphoresKHR

### Description
Enqueues a wait operation on one or more semaphores. This command will block until the specified semaphores are signaled.

### Method
`clEnqueueWaitSemaphoresKHR`

### Parameters
#### Command Queue
- **command_queue** (cl_command_queue) - The command queue to enqueue the wait command into.

#### Semaphore Objects
- **num_sema_objects** (cl_uint) - The number of semaphore objects in `sema_objects`.
- **sema_objects** (const cl_semaphore_khr *) - A list of semaphore objects to wait on.
- **sema_payload_list** (const cl_semaphore_payload_khr *) - A list of payloads corresponding to the semaphores. Can be NULL if not needed.

#### Event Wait List
- **num_events_in_wait_list** (cl_uint) - The number of events in `event_wait_list`.
- **event_wait_list** (const cl_event *) - A list of events that must complete before this command can execute.

#### Event
- **event** (cl_event *) - Returns an event object that identifies this particular command in the command queue. Can be NULL.

### Request Example
```c
cl_semaphore_khr clSema;
cl_command_queue command_queue;

clEnqueueWaitSemaphoresKHR(command_queue, 1, &clSema, NULL, 0, NULL, NULL);
```

### Response
#### Success Response
- **event** (cl_event) - An event object representing the enqueued wait operation.
```

--------------------------------

### Creating an OpenCL Program from Binary

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/clCreateProgramWithBinary.html

Demonstrates how to create an OpenCL program object from binary data. This involves specifying the devices, the binary data, and the format of the binary.

```APIDOC
## clCreateProgramWithBinary

### Description
Creates an OpenCL program object from a binary representation.

### Method Signature
```c
cl_program clCreateProgramWithBinary(
    cl_context context,
    cl_uint num_devices,
    const cl_device_id *device_list,
    const unsigned char **binaries,
    cl_int *binary_status,
    cl_int *errcode_ret
);
```

### Parameters
*   **context**: A valid OpenCL context.
*   **num_devices**: The number of devices in `device_list` and `binaries`.
*   **device_list**: A list of OpenCL devices associated with the program.
*   **binaries**: A list of pointers to the binary program data for each device.
*   **binary_status**: An array that receives the status of loading the binary for each device. Can be NULL.
*   **errcode_ret**: A pointer to a location that receives the return error code. Can be NULL.

### Return Value
On success, `clCreateProgramWithBinary` returns a valid program object. Otherwise, it returns NULL.

### Error Codes
*   `CL_SUCCESS`
*   `CL_INVALID_CONTEXT`
*   `CL_INVALID_VALUE`
*   `CL_INVALID_DEVICE`
*   `CL_INVALID_BINARY`
*   `CL_OUT_OF_HOST_MEMORY`
*   `CL_OUT_OF_RESOURCES`

### Example
```c
// Assume context, num_devices, device_list are already initialized
unsigned char *binary_data[1]; // Pointer to your binary data
cl_int binary_status[1];
cl_int errcode;

cl_program program = clCreateProgramWithBinary(context, num_devices, device_list, (const unsigned char **)binary_data, binary_status, &errcode);

if (errcode != CL_SUCCESS) {
    // Handle error
}
```
```

--------------------------------

### OpenCL Command-Buffer Capabilities Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

Demonstrates querying platform command-buffer capabilities and setting up OpenCL resources for command-buffer operations. It checks for automatic remapping support and creates context, command queues, and a program.

```C++
int main() {
  cl_platform_id platform;
  CL_CHECK(clGetPlatformIDs(1, &platform, nullptr));
  cl_platform_command_buffer_capabilities_khr platform_caps;
  CL_CHECK(clGetPlatformInfo(platform,
                             CL_PLATFORM_COMMAND_BUFFER_CAPABILITIES_KHR,
                             sizeof(platform_caps), &platform_caps, NULL));
  if (!(platform_caps & CL_COMMAND_BUFFER_PLATFORM_AUTOMATIC_REMAP_KHR)) {
    std::cerr << "Command-buffer remapping not supported but used in example, "
                 "skipping\n";
    return 0;
  }

  cl_uint num_devices = 0;
  CL_CHECK(clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 0, NULL, &num_devices));
  std::vector<cl_device_id> devices(num_devices);
  CL_CHECK(
      clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 1, devices.data(), nullptr));

  // Checks omitted for brevity that either a) the platform supports
  // CL_COMMAND_BUFFER_PLATFORM_UNIVERSAL_SYNC_KHR or b) each device is listed
  // in the others CL_DEVICE_COMMAND_BUFFER_SYNC_DEVICES_KHR

  cl_int error;
  cl_context context =
      clCreateContext(NULL, num_devices, devices.data(), NULL, NULL, &error);
  CL_CHECK(error);

  std::vector<cl_command_queue> queues(num_devices);
  for (cl_uint i = 0; i < num_devices; i++) {
    queues[i] = clCreateCommandQueue(context, devices[i], 0, &error);
    CL_CHECK(error);
  }

  const char *code = R"OpenCLC(
      kernel void vector_addition(global int* tile1, global int* tile2,
                                  global int* res) {
        size_t index = get_global_id(0);
        res[index] = tile1[index] + tile2[index];
      }
      )OpenCLC";
  const size_t length = std::strlen(code);

  cl_program program =
      clCreateProgramWithSource(context, 1, &code, &length, &error);
  CL_CHECK(error);

  CL_CHECK(
      clBuildProgram(program, num_devices, devices.data(), NULL, NULL, NULL));

  cl_kernel kernel = clCreateKernel(program, "vector_addition", &error);
  CL_CHECK(error);

  constexpr size_t frame_count = 60;
  constexpr size_t frame_elements = 1024;
  constexpr size_t frame_size = frame_elements * sizeof(cl_int);

  constexpr size_t tile_count = 16;
  constexpr size_t tile_elements = frame_elements / tile_count;
  constexpr size_t tile_size = tile_elements * sizeof(cl_int);

  cl_mem buffer_tile1 =
      clCreateBuffer(context, CL_MEM_READ_ONLY, tile_size, NULL, &error);
  CL_CHECK(error);

  cl_mem buffer_tile2 =
      clCreateBuffer(context, CL_MEM_READ_ONLY, tile_size, NULL, &error);
  CL_CHECK(error);

  cl_mem buffer_res =
      clCreateBuffer(context, CL_MEM_WRITE_ONLY, tile_size, NULL, &error);
  CL_CHECK(error);

  CL_CHECK(clSetKernelArg(kernel, 0, sizeof(buffer_tile1), &buffer_tile1));
```

--------------------------------

### Creating a Sub-device by Name

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/intel/cl_intel_device_partition_by_names.txt

Shows how to use clCreateSubDevices with CL_DEVICE_PARTITION_BY_NAMES_INTEL to create a sub-device.

```c
cl_device_id                     subdevice_id;
cl_uint                          num_entries  = 1;
cl_uint                          num_devices  = 1;
cl_device_partition_property properties[] = {CL_DEVICE_PARTITION_BY_NAMES_INTEL, 0, CL_PARTITION_BY_NAMES_LIST_END_INTEL, 0};

err = clCreateSubDevices(device, properties, num_entries, &subdevice_id, &num_devices);
```

--------------------------------

### OpenCL half Kernel Function Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/halfDataType.html

An example of a kernel function using the half data type for global and local memory pointers.

```c
__kernel void
    foo (__global half *pg, __local half *pl)
    {
        __global half *ptr;
        int offset;

        ptr = pg + offset;
        bar(ptr);
    }
```

--------------------------------

### Synchronization Using Wait and Signal

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_semaphore.html

This example demonstrates how to use semaphores for synchronization. It shows the creation of a semaphore, waiting for it using clEnqueueWaitSemaphoresKHR, launching a kernel, and signaling the semaphore using clEnqueueSignalSemaphoresKHR.

```APIDOC
## clEnqueueWaitSemaphoresKHR

### Description
Enqueues a wait operation on one or more semaphore objects.

### Method
`clEnqueueWaitSemaphoresKHR`

### Parameters
- `command_queue` (cl_command_queue) - The command queue.
- `num_sema_objects` (cl_uint) - The number of semaphore objects to wait on.
- `sema_objects` (const cl_semaphore_khr *) - A list of semaphore objects to wait on.
- `sema_payload_list` (const cl_semaphore_payload_khr *) - A list of payloads corresponding to the semaphore objects. Can be NULL if not needed.
- `num_events_in_wait_list` (cl_uint) - The number of events in the `event_wait_list`.
- `event_wait_list` (const cl_event *) - A list of events that must be completed before this command can be executed.
- `event` (cl_event *) - Returns an event object that identifies this particular command in the command queue. Can be NULL.

## clEnqueueSignalSemaphoresKHR

### Description
Enqueues a signal operation on one or more semaphore objects.

### Method
`clEnqueueSignalSemaphoresKHR`

### Parameters
- `command_queue` (cl_command_queue) - The command queue.
- `num_sema_objects` (cl_uint) - The number of semaphore objects to signal.
- `sema_objects` (const cl_semaphore_khr *) - A list of semaphore objects to signal.
- `sema_payload_list` (const cl_semaphore_payload_khr *) - A list of payloads corresponding to the semaphore objects. Can be NULL if not needed.
- `num_events_in_wait_list` (cl_uint) - The number of events in the `event_wait_list`.
- `event_wait_list` (const cl_event *) - A list of events that must be completed before this command can be executed.
- `event` (cl_event *) - Returns an event object that identifies this particular command in the command queue. Can be NULL.

## Example Usage

```c
// clSema is created using clCreateSemaphoreWithPropertiesKHR
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
    clEnqueueWaitSemaphoresKHR(command_queue, 1, &clSema, NULL, 0, NULL, NULL);

    // Launch kernel that accesses extMem
    clEnqueueNDRangeKernel(command_queue, kernel, ...);

    // Signal the semaphore in OpenCL
    clEnqueueSignalSemaphoresKHR(command_queue, 1, &clSema, NULL, 0, NULL, NULL);

    // (not shown) Launch other work that waits on 'clSema'
}
```
```

--------------------------------

### OpenCL Kernel Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

This is an example of an OpenCL kernel function that performs image filtering. It includes standard C-like syntax and OpenCL image types.

```c
#include <foo.h>
    #include <mydir/myinc.h>
    __kernel void
    image_filter (int n, int m,
                  __constant float *filter_weights,
                  __read_only image2d_t src_image,
                  __write_only image2d_t dst_image)
    {
    ...
    }
```

--------------------------------

### Creating and Building an OpenCL Program

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_command_buffer_multi_device.html

Demonstrates the process of creating an OpenCL program from source code and building it for the specified devices. Ensure the CL_CHECK macro is defined for error handling.

```c
const size_t length = std::strlen(code);

      cl_program program = 
          clCreateProgramWithSource(context, 1, &code, &length, &error);
      CL_CHECK(error);

      CL_CHECK(
          clBuildProgram(program, num_devices, devices.data(), NULL, NULL, NULL));

      cl_kernel kernel = clCreateKernel(program, "vector_addition", &error);
      CL_CHECK(error);
```

--------------------------------

### Example Linux ICD File Content

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

An example of the content within a file in the '/etc/OpenCL/vendors' directory on Linux. This line specifies the shared object library to be loaded.

```text
libVendorAOpenCL.so
```

--------------------------------

### Example Windows Registry Value for OpenCLDriverName

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

An example of a registry value entry under a Class key, specifying the path to a vendor's OpenCL ICD library.

```ini
[HKLM\SYSTEM\CurrentControlSet\Control\Class\{GUID}\\{Instance}]
"OpenCLDriverName"="c:\\vendor a\\vndra_ocl.dll"
```

--------------------------------

### OpenCL C Variable Initialization Examples

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_C.html

Demonstrates allowed and disallowed initializations for variables in different address spaces (__global, __local, __constant, private).

```c
global int a = 12;      // Initialization is allowed.
global int b;
constant int c = 12;    // Initializer is a compile time constant.
constant int d;         // Error. No initializer provided.
kernel void my_func(...)
{
    local float e = 1;  // Error. Initializer is not allowed. 

    local float f;
    f = 1;              // Allowed
    private int g;      // Uninitialized.
    constant int h = g; // Error. Initializer is not a constant expression.
}
```

--------------------------------

### Create and Populate Command Buffer

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_command_buffer.html

Demonstrates the creation of a command buffer and populating it with copy and kernel execution commands. Ensures proper synchronization between operations using sync points.

```c++
cl_command_queue command_queue =
          clCreateCommandQueue(context, device,
                               CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &error);
        CL_CHECK(error);
    
        cl_command_buffer_khr command_buffer =
            clCreateCommandBufferKHR(1, &command_queue, nullptr, &error);
        CL_CHECK(error);
    
        cl_mem buffer_src1 =
            clCreateBuffer(context, CL_MEM_READ_ONLY, frame_size, nullptr, &error);
        CL_CHECK(error);
        cl_mem buffer_src2 =
            clCreateBuffer(context, CL_MEM_READ_ONLY, frame_size, nullptr, &error);
        CL_CHECK(error);
        cl_mem buffer_dst =
            clCreateBuffer(context, CL_MEM_WRITE_ONLY, frame_size, nullptr, &error);
        CL_CHECK(error);
    
        cl_sync_point_khr tile_sync_point = 0;
        for (size_t tile_index = 0; tile_index < tile_count; tile_index++) {
          std::array<cl_sync_point_khr, 2> copy_sync_points;
          CL_CHECK(clCommandCopyBufferKHR(command_buffer,
              command_queue, buffer_src1, buffer_tile1, tile_index * tile_size, 0,
              tile_size, tile_sync_point ? 1 : 0,
              tile_sync_point ? &tile_sync_point : nullptr, &copy_sync_points[0]),
              nullptr);
          CL_CHECK(clCommandCopyBufferKHR(command_buffer,
              command_queue, buffer_src2, buffer_tile2, tile_index * tile_size, 0,
              tile_size, tile_sync_point ? 1 : 0,
              tile_sync_point ? &tile_sync_point : nullptr, &copy_sync_points[1]),
              nullptr);
    
          cl_sync_point_khr nd_sync_point;
          CL_CHECK(clCommandNDRangeKernelKHR(command_buffer,
              command_queue, nullptr, kernel, 1, nullptr, &tile_elements, nullptr,
              copy_sync_points.size(), copy_sync_points.data(), &nd_sync_point,
              nullptr));
    
          CL_CHECK(clCommandCopyBufferKHR(command_buffer,
              command_queue, buffer_res, buffer_dst, 0, tile_index * tile_size,
              tile_size, 1, &nd_sync_point, &tile_sync_point, nullptr));
        }
    
        CL_CHECK(clFinalizeCommandBufferKHR(command_buffer));
```

--------------------------------

### Creating a program with built-in kernels

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/clCreateProgramWithBuiltInKernels.html

Demonstrates how to create an OpenCL program object using a list of built-in kernel names.

```APIDOC
## clCreateProgramWithBuiltInKernels

### Description
Creates an OpenCL program object from a set of built-in kernels.

### Parameters
#### Path Parameters
- **context** (cl_context) - Required - The OpenCL context in which to create the program object.
- **device_list** (const cl_device_id *) - Required - A list of devices for which the built-in kernels will be available. The list must be NULL-terminated.
- **kernel_names** (const char *) - Required - A null-terminated string containing the names of the built-in kernels to be included in the program object. The names in the string must be separated by semicolons. If an empty string is provided, all built-in kernels supported by the devices in `device_list` are included.

### Returns
- **cl_program** - A handle to the created OpenCL program object. Returns NULL on failure.

### Errors
- **CL_INVALID_CONTEXT**: If `context` is not a valid OpenCL context.
- **CL_INVALID_VALUE**: If `device_list` or `kernel_names` are NULL, or if `kernel_names` contains invalid kernel names or is malformed.
- **CL_INVALID_DEVICE**: If `device_list` contains devices that are not associated with `context`.
- **CL_OUT_OF_HOST_MEMORY**: If the program could not be allocated due to host memory limitations.

### Example
```c
cl_int err;
cl_program program;
const char* kernel_names = "my_kernel1;my_kernel2";

program = clCreateProgramWithBuiltInKernels(context, 1, &device, kernel_names, &err);
if (err != CL_SUCCESS) {
    // Handle error
}
```
```

--------------------------------

### Example: Inclusive Scan Add Operation

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/workGroupFunctions.html

Demonstrates the usage of `work_group_scan_inclusive_add` for calculating prefix sums within a work-group. The example shows how values are accumulated across work-items.

```c
void foo(int *p)
{
    ...
    int prefix_sum_val = work_group_scan_inclusive_add(
                                p[get_local_id(0)]);
}

```

--------------------------------

### Example Usage of atomic_init

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/atomic_init.html

This example demonstrates how to use atomic_init to initialize a local atomic integer variable within a work-group. It ensures initialization happens only once per work-group.

```c
local atomic_int guide;
if (get_local_id(0) == 0)
    atomic_init(&guide, 42);
work_group_barrier(CLK_LOCAL_MEM_FENCE);
```

--------------------------------

### Example Windows Registry Value for Vendors Key

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

An example of a registry value entry under the 'Vendors' key, where the value name is the library path and the data is a DWORD set to 0.

```ini
[HKLM\SOFTWARE\Khronos\OpenCL\Vendors]
"c:\\vendor a\\vndra_ocl.dll"=dword:00000000
```

--------------------------------

### Sample Code for ARM printf Extension

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/arm/cl_arm_printf.html

Example demonstrating how to use the ARM printf extension in both host C code and device OpenCL C code.

```APIDOC
## Host C Code Example

```c
/* Define a printf callback function. */
void printf_callback( const char *buffer, size_t len, size_t complete, void *user_data )
{
    printf( "%.*s", len, buffer );
}

/* Create a cl_context with a printf_callback and user specified buffer size. */
cl_context_properties properties[] =
{
    /* Enable a printf callback function for this context. */
    CL_PRINTF_CALLBACK_ARM,   (cl_context_properties) printf_callback,

    /* Request a minimum printf buffer size of 4MiB for devices in the
       context that support this extension. */
    CL_PRINTF_BUFFERSIZE_ARM, (cl_context_properties) 0x100000,

    CL_CONTEXT_PLATFORM,      (cl_context_properties) platform,
    0
};
cl_context context = clCreateContext( properties, 1, &device, NULL, NULL, NULL );
```

## Device OpenCL C Code Example

```c
// Only required by version 1.0.0 of the extension, version 2.0.0 does not
// require the following pragma.
#pragma OPENCL EXTENSION cl_arm_printf : enable

kernel void hello_world( void )
{
    printf( "Hello from work item %lu!\n", (ulong) get_global_id(0) );
}
```
```

--------------------------------

### Create Command Queue with Thread Local Execution

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/intel/cl_intel_exec_by_local_thread.txt

Example of creating an in-order command queue that enables thread-local execution using the CL_QUEUE_THREAD_LOCAL_EXEC_ENABLE_INTEL property.

```c
cl_command_queue queue1 = clCreateCommandQueue (context, device_id, CL_QUEUE_THREAD_LOCAL_EXEC_ENABLE_INTEL, &err);
```

--------------------------------

### Get Suggested Local Work Size (KHR Extension)

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

Use clGetKernelSuggestedLocalWorkSizeKHR, provided by the cl_khr_suggested_local_work_size extension, to get a suggested local work size. This function is an alternative to the standard OpenCL 3.1 version.

```c
// Provided by cl_khr_suggested_local_work_size
cl_int clGetKernelSuggestedLocalWorkSizeKHR(
    cl_command_queue command_queue,
    cl_kernel kernel,
    cl_uint work_dim,
    const size_t* global_work_offset,
    const size_t* global_work_size,
    size_t* suggested_local_work_size);
```

--------------------------------

### Execute and Remap Command Buffer

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/cl_khr_command_buffer_multi_device.html

Demonstrates the execution of a command buffer on multiple devices and its subsequent remapping to a single queue for re-execution. Includes resource cleanup.

```c++
CL_CHECK(clWaitForEvents(1, enqueue_events[2]));
    
          for (auto e : enqueue_events) {
            CL_CHECK(clReleaseEvent(e));
          }
        }
        return 0;
      };
    
      error = enqueue_frame(original_cmdbuf);
      CL_CHECK(error);
    
      // Remap from N queues to 1 queue and run again
      cl_command_buffer_khr remapped_cmdbuf = clRemapCommandBufferKHR(
          original_cmdbuf, CL_TRUE, 1, queues.data(), 0, NULL, NULL, &error);
      CL_CHECK(error);
    
      error = enqueue_frame(remapped_cmdbuf);
      CL_CHECK(error);
    
      for (unsigned i = 0; i < num_devices; ++i) {
        CL_CHECK(clReleaseCommandQueue(queues[i]));
      }
      CL_CHECK(clReleaseMemObject(buffer_src1));
      CL_CHECK(clReleaseMemObject(buffer_src2));
      CL_CHECK(clReleaseMemObject(buffer_dst));
    
      CL_CHECK(clReleaseMemObject(buffer_tile1));
      CL_CHECK(clReleaseMemObject(buffer_tile2));
      CL_CHECK(clReleaseMemObject(buffer_res));
    
      CL_CHECK(clReleaseCommandBufferKHR(original_cmdbuf));
      CL_CHECK(clReleaseCommandBufferKHR(remapped_cmdbuf));
    
      CL_CHECK(clReleaseKernel(kernel));
      CL_CHECK(clReleaseProgram(program));
      CL_CHECK(clReleaseContext(context));
    
      return 0;
    }
```

--------------------------------

### Sample Kernel for NV12 Image Operations (CL_INTEL_PLANAR_YUV)

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/intel/cl_intel_planar_yuv.html

Demonstrates reading from and writing to a whole NV12 image. Ensure the CL_NV12_INTEL format is supported with appropriate access flags (CL_MEM_READ_ONLY or CL_MEM_READ_WRITE) as determined by clGetSupportedImageFormats.

```c
// do something with a whole NV12 image
    kernel void DoSomethingWithNV12
    (
        ...
        read_write image2d_t nv12Img,
        ...
    )
    {
        ...
        // sample the CL_NV12_INTEL image - supported if CL_NV12_INTEL format is
        // available with CL_MEM_READ_ONLY or CL_MEM_READ_WRITE access flags
        // based on clGetSupportedImageFormats query.
        float4 p = read_imagef(nv12Img, sampler, coord);
        ...
        // write to the CL_NV12_INTEL image - supported if CL_NV12_INTEL format is
        // available with CL_MEM_WRITE_ONLY or CL_MEM_READ_WRITE access flags
        // based on clGetSupportedImageFormats query.
        write_imagef(nv12Img, coord, p);
        ...
    }
```

--------------------------------

### Get Vector Element Count with vec_step

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/miscVectorFunctions.html

Use vec_step to get the number of elements in a scalar or vector type. For all scalar types, it returns 1. For 3-component vectors, it returns 4. It can also take a type name as an argument.

```c
vec_step(float2)
```

--------------------------------

### OpenCL Semaphore Synchronization Example

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_API.html

Demonstrates creating, exporting, and synchronizing with an OpenCL semaphore across different APIs or processes. Requires OpenCL 2.0 or later and the cl_khr_semaphore extension.

```c
// 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_SEMAPHORE_DEVICE_HANDLE_LIST_KHR,
        (cl_semaphore_properties_khr)devices[1],
        CL_SEMAPHORE_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'
    }
```

--------------------------------

### Partitioning by Affinity Domain

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/refpages/man/html/clCreateSubDevices.html

Example of partitioning a device along its outermost cache line.

```APIDOC
{
    CL_DEVICE_PARTITION_BY_AFFINITY_DOMAIN,
    CL_DEVICE_AFFINITY_DOMAIN_NEXT_PARTITIONABLE,
    0
}
```

--------------------------------

### Get Image Depth (OpenCL C)

Source: https://github.com/khronosgroup/opencl-registry/blob/main/specs/unified/html/OpenCL_C.html

Returns the depth of a 3D image in pixels.

```c
int **get_image_depth**(image3d_t _image_)

```

--------------------------------

### Enqueuing Kernel with Non-Uniform Work Group Size

Source: https://github.com/khronosgroup/opencl-registry/blob/main/extensions/arm/cl_arm_non_uniform_work_group_size.txt

This example demonstrates how to build a program with the non-uniform work group size extension enabled and then enqueue a kernel with a local work-group size that is not a factor of the global work-group size. Ensure the OpenCL context and queue are properly initialized.

```c
const char *source[] = { "kernel void example( void ){}\n" };

cl_program program = clCreateProgramWithSource( context, 1, source,
                                                NULL, NULL );

clBuildProgram( program, 0, NULL, "-cl-arm-non-uniform-work-group-size",
                NULL, NULL );

cl_kernel kernel = clCreateKernel( program, "example", NULL );

size_t global_work_size = 10;
size_t local_work_size = 9;
err = clEnqueueNDRangeKernel( queue, kernel, 1, NULL,
                              &global_work_size, &local_work_size,
                              0, NULL, NULL );
assert( CL_SUCCESS == err );
```