/**
* @internal
*
* @brief Implementation of ::ccl_dev_container_get_cldevices() for the
* platform wrapper.
*
* @private @memberof ccl_platform
*
* @param[in] devcon A ::CCLPlatform wrapper, passed as a ::CCLDevContainer.
* @param[out] err Return location for a ::CCLErr object, or `NULL` if error
* reporting is to be ignored.
* @return A list of `cl_device_id` objects inside a ::CCLWrapperInfo object.
* */
static CCLWrapperInfo * ccl_platform_get_cldevices(
CCLDevContainer * devcon, CCLErr ** err) {
/* Make sure err is NULL or it is not set. */
g_return_val_if_fail(err == NULL || *err == NULL, NULL);
/* Make sure devcon is not NULL. */
g_return_val_if_fail(devcon != NULL, NULL);
CCLWrapperInfo * info = NULL;
cl_int ocl_status;
/* Determine number of devices. */
ocl_status = clGetDeviceIDs(devcon->base.cl_object,
CL_DEVICE_TYPE_ALL, 0, NULL, &devcon->num_devices);
g_if_err_create_goto(*err, CCL_OCL_ERROR,
CL_SUCCESS != ocl_status, ocl_status, error_handler,
"%s: get number of devices (OpenCL error %d: %s).",
CCL_STRD, ocl_status, ccl_err(ocl_status));
/* Create info object with size in bytes of array of device IDs. */
info = ccl_wrapper_info_new(
sizeof(cl_device_id) * devcon->num_devices);
/* Get existing device IDs. */
ocl_status = clGetDeviceIDs(devcon->base.cl_object,
CL_DEVICE_TYPE_ALL, devcon->num_devices, info->value, NULL);
g_if_err_create_goto(*err, CCL_OCL_ERROR,
CL_SUCCESS != ocl_status, ocl_status, error_handler,
"%s: get device IDs (OpenCL error %d: %s).",
CCL_STRD, ocl_status, ccl_err(ocl_status));
/* Add device list to info table, so that it will be
* automatically released. Because the cl_platform_id object
* doesn't have a CL_PLATFORM_DEVICES parameter, we keep this info
* referenced has CL_CONTEXT_DEVICES. */
ccl_wrapper_add_info((CCLWrapper *) devcon, CL_CONTEXT_DEVICES, info);
/* If we got here, everything is OK. */
g_assert(err == NULL || *err == NULL);
goto finish;
error_handler:
/* If we got here there was an error, verify that it is so. */
g_assert(err == NULL || *err != NULL);
/* Free info if it was created. */
if (info != NULL) ccl_wrapper_info_destroy(info);
finish:
/* Terminate function. */
return info;
}
/**
* Create a new kernel wrapper object.
*
* @public @memberof ccl_kernel
*
* @param[in] prg A program wrapper object.
* @param[in] kernel_name The kernel name.
* @param[out] err Return location for a ::CCLErr object, or `NULL` if error
* reporting is to be ignored.
* @return A new kernel wrapper object.
* */
CCL_EXPORT
CCLKernel* ccl_kernel_new(
CCLProgram* prg, const char* kernel_name, CCLErr** err) {
/* Make sure err is NULL or it is not set. */
g_return_val_if_fail((err) == NULL || *(err) == NULL, NULL);
/* Make sure prg is not NULL. */
g_return_val_if_fail(prg != NULL, NULL);
/* Make sure kernel_name is not NULL. */
g_return_val_if_fail(kernel_name != NULL, NULL);
/* Kernel wrapper object. */
CCLKernel* krnl = NULL;
/* OpenCL return status. */
cl_int ocl_status;
/* The OpenCL kernel object. */
cl_kernel kernel = NULL;
/* Create kernel. */
kernel = clCreateKernel(ccl_program_unwrap(prg),
kernel_name, &ocl_status);
ccl_if_err_create_goto(*err, CCL_OCL_ERROR,
CL_SUCCESS != ocl_status, ocl_status, error_handler,
"%s: unable to create kernel (OpenCL error %d: %s).",
CCL_STRD, ocl_status, ccl_err(ocl_status));
/* Create kernel wrapper. */
krnl = ccl_kernel_new_wrap(kernel);
/* If we got here, everything is OK. */
g_assert(err == NULL || *err == NULL);
goto finish;
error_handler:
/* If we got here there was an error, verify that it is so. */
g_assert(err == NULL || *err != NULL);
krnl = NULL;
finish:
/* Return kernel wrapper. */
return krnl;
}
/**
* Enqueues a command to execute a native C/C++ function not compiled
* using the OpenCL compiler. This function is a wrapper for the
* clEnqueueNativeKernel() OpenCL function, the documentation of which
* provides additional information.
*
* @public @memberof ccl_kernel
*
* @param[in] cq A command queue wrapper object.
* @param[in] user_func A pointer to a host-callable user function.
* @param[in] args A pointer to the args list that `user_func` should be
* called with.
* @param[in] cb_args The size in bytes of the args list that args
* points to.
* @param[in] num_mos The number of ::CCLMemObj* objects that are passed
* in `mo_list`.
* @param[in] mo_list A list of ::CCLMemObj* objects (or `NULL`
* references), if num_mos > 0.
* @param[in] args_mem_loc A pointer to appropriate locations that
* `args` points to where `cl_mem` values (unwrapped from the respective
* ::CCLMemObj* objects) are stored. Before the user function is
* executed, the `cl_mem` values are replaced by pointers to global
* memory.
* @param[in,out] evt_wait_lst List of events that need to complete
* before this command can be executed. The list will be cleared and
* can be reused by client code.
* @param[out] err Return location for a ::CCLErr object, or `NULL` if error
* reporting is to be ignored.
* @return Event wrapper object that identifies this command.
* */
CCL_EXPORT
CCLEvent* ccl_kernel_enqueue_native(CCLQueue* cq,
void (CL_CALLBACK * user_func)(void*), void* args, size_t cb_args,
cl_uint num_mos, CCLMemObj* const* mo_list,
const void** args_mem_loc, CCLEventWaitList* evt_wait_lst,
CCLErr** err) {
/* Make sure cq is not NULL. */
g_return_val_if_fail(cq != NULL, NULL);
/* Make sure user_func is not NULL. */
g_return_val_if_fail(user_func != NULL, NULL);
/* Make sure that num_mos == 0 AND mo_list != NULL, OR, that
* num_mos > 0 AND mo_list != NULL */
g_return_val_if_fail(((num_mos == 0) && (mo_list == NULL))
|| ((num_mos > 0) && (mo_list != NULL)), NULL);
/* Make sure err is NULL or it is not set. */
g_return_val_if_fail(err == NULL || *err == NULL, NULL);
/* OpenCL status flag. */
cl_int ocl_status;
/* OpenCL event. */
cl_event event = NULL;
/* Event wrapper. */
CCLEvent* evt = NULL;
/* List of cl_mem objects. */
cl_mem* mem_list = NULL;
/* Unwrap memory objects. */
if (num_mos > 0) {
mem_list = g_slice_alloc(sizeof(cl_mem) * num_mos);
for (cl_uint i = 0; i < num_mos; ++i) {
mem_list[i] = mo_list[i] != NULL
? ccl_memobj_unwrap(mo_list[i])
: NULL;
}
}
/* Enqueue kernel. */
ocl_status = clEnqueueNativeKernel(ccl_queue_unwrap(cq), user_func,
args, cb_args, num_mos, (const cl_mem*) mem_list, args_mem_loc,
ccl_event_wait_list_get_num_events(evt_wait_lst),
ccl_event_wait_list_get_clevents(evt_wait_lst), &event);
ccl_if_err_create_goto(*err, CCL_OCL_ERROR,
CL_SUCCESS != ocl_status, ocl_status, error_handler,
"%s: unable to enqueue native kernel (OpenCL error %d: %s).",
CCL_STRD, ocl_status, ccl_err(ocl_status));
/* Wrap event and associate it with the respective command queue.
* The event object will be released automatically when the command
* queue is released. */
evt = ccl_queue_produce_event(cq, event);
/* Clear event wait list. */
ccl_event_wait_list_clear(evt_wait_lst);
/* If we got here, everything is OK. */
g_assert(err == NULL || *err == NULL);
goto finish;
error_handler:
/* If we got here there was an error, verify that it is so. */
g_assert(err == NULL || *err != NULL);
finish:
/* Release temporary cl_mem list. */
if (num_mos > 0)
g_slice_free1(sizeof(cl_mem) * num_mos, mem_list);
/* Return event wrapper. */
return evt;
}
/**
* Enqueues a kernel for execution on a device.
*
* Internally, this function calls the clSetKernelArg() OpenCL function
* for each argument defined with the ::ccl_kernel_set_arg() function,
* and the executes the kernel using the clEnqueueNDRangeKernel() OpenCL
* function.
*
* @warning This function is not thread-safe. For multi-threaded
* access to the same kernel function, create multiple instances of
* a kernel wrapper for the given kernel function with
* ::ccl_kernel_new(), one for each thread.
*
* @public @memberof ccl_kernel
*
* @param[in] krnl A kernel wrapper object.
* @param[in] cq A command queue wrapper object.
* @param[in] work_dim The number of dimensions used to specify the
* global work-items and work-items in the work-group.
* @param[in] global_work_offset Can be used to specify an array of
* `work_dim` unsigned values that describe the offset used to calculate
* the global ID of a work-item.
* @param[in] global_work_size An array of `work_dim` unsigned values
* that describe the number of global work-items in `work_dim`
* dimensions that will execute the kernel function.
* @param[in] local_work_size An array of `work_dim` unsigned values
* that describe the number of work-items that make up a work-group that
* will execute the specified kernel.
* @param[in,out] evt_wait_lst List of events that need to complete
* before this command can be executed. The list will be cleared and
* can be reused by client code.
* @param[out] err Return location for a ::CCLErr object, or `NULL` if error
* reporting is to be ignored.
* @return Event wrapper object that identifies this command.
* */
CCL_EXPORT
CCLEvent* ccl_kernel_enqueue_ndrange(CCLKernel* krnl, CCLQueue* cq,
cl_uint work_dim, const size_t* global_work_offset,
const size_t* global_work_size, const size_t* local_work_size,
CCLEventWaitList* evt_wait_lst, CCLErr** err) {
/* Make sure krnl is not NULL. */
g_return_val_if_fail(krnl != NULL, NULL);
/* Make sure cq is not NULL. */
g_return_val_if_fail(cq != NULL, NULL);
/* Make sure err is NULL or it is not set. */
g_return_val_if_fail(err == NULL || *err == NULL, NULL);
/* OpenCL status flag. */
cl_int ocl_status;
/* OpenCL event. */
cl_event event;
/* Event wrapper. */
CCLEvent* evt;
/* Iterator for table of kernel arguments. */
GHashTableIter iter;
gpointer arg_index_ptr, arg_ptr;
/* Set pending kernel arguments. */
if (krnl->args != NULL) {
g_hash_table_iter_init(&iter, krnl->args);
while (g_hash_table_iter_next(&iter, &arg_index_ptr, &arg_ptr)) {
cl_uint arg_index = GPOINTER_TO_UINT(arg_index_ptr);
CCLArg* arg = (CCLArg*) arg_ptr;
ocl_status = clSetKernelArg(ccl_kernel_unwrap(krnl), arg_index,
ccl_arg_size(arg), ccl_arg_value(arg));
ccl_if_err_create_goto(*err, CCL_OCL_ERROR,
CL_SUCCESS != ocl_status, ocl_status, error_handler,
"%s: unable to set kernel arg %d (OpenCL error %d: %s).",
CCL_STRD, arg_index, ocl_status, ccl_err(ocl_status));
g_hash_table_iter_remove(&iter);
}
}
/* Run kernel. */
ocl_status = clEnqueueNDRangeKernel(ccl_queue_unwrap(cq),
ccl_kernel_unwrap(krnl), work_dim, global_work_offset,
global_work_size, local_work_size,
ccl_event_wait_list_get_num_events(evt_wait_lst),
ccl_event_wait_list_get_clevents(evt_wait_lst), &event);
ccl_if_err_create_goto(*err, CCL_OCL_ERROR,
CL_SUCCESS != ocl_status, ocl_status, error_handler,
"%s: unable to enqueue kernel (OpenCL error %d: %s).",
CCL_STRD, ocl_status, ccl_err(ocl_status));
/* Wrap event and associate it with the respective command queue.
* The event object will be released automatically when the command
* queue is released. */
evt = ccl_queue_produce_event(cq, event);
/* Clear event wait list. */
ccl_event_wait_list_clear(evt_wait_lst);
/* If we got here, everything is OK. */
g_assert(err == NULL || *err == NULL);
goto finish;
error_handler:
/* If we got here there was an error, verify that it is so. */
g_assert(err == NULL || *err != NULL);
/* An error occurred, return NULL to signal it. */
evt = NULL;
finish:
/* Return evt. */
return evt;
}
/**
* Create a new sampler wrapper object using a list of properties.
*
* If a supported property is not specified, a default value is used.
* Some valid properties are `CL_SAMPLER_NORMALIZED_COORDS` (default
* value is `CL_TRUE`), `CL_SAMPLER_ADDRESSING_MODE` (default value is
* `CL_ADDRESS_CLAMP`) and `CL_SAMPLER_FILTER_MODE` (default value is
* `CL_FILTER_NEAREST`).
*
* This function mimicks the style of the OpenCL 2.0 sampler
* constructor, clCreateSamplerWithProperties(), but can be used with
* any version of OpenCL. Thus, The underlying OpenCL sampler object is
* created using:
*
* * clCreateSampler() - for platforms with OpenCL version <= 1.2
* * clCreateSamplerWithProperties() - for platforms with OpenCL version
* >= 2.0.
*
* @public @memberof ccl_sampler
*
* @param[in] ctx A context wrapper object.
* @param[in] sampler_properties A list of sampler property names and
* their corresponding values. Each sampler property name is immediately
* followed by the corresponding desired value. The list is terminated
* with 0. If a supported property is not specified, its default value
* will be used. If `NULL`, default values for supported sampler
* properties will be used.
* @param[out] err Return location for a ::CCLErr object, or `NULL` if error
* reporting is to be ignored.
* @return A new sampler wrapper object or `NULL` if an error occurs.
* */
CCL_EXPORT
CCLSampler* ccl_sampler_new_full(CCLContext* ctx,
const cl_sampler_properties *sampler_properties, CCLErr** err) {
/* Make sure err is NULL or it is not set. */
g_return_val_if_fail((err) == NULL || *(err) == NULL, NULL);
/* Make sure ctx is not NULL. */
g_return_val_if_fail(ctx != NULL, NULL);
/* New sampler wrapper object to create. */
CCLSampler* smplr = NULL;
/* OpenCL sampler object to create and wrap. */
cl_sampler sampler;
/* OpenCL function status. */
cl_int ocl_status;
#ifdef CL_VERSION_2_0
/* OpenCL platform version. */
double ocl_ver;
/* Internal error handling object. */
CCLErr* err_internal = NULL;
/* Get context platform version. */
ocl_ver = ccl_context_get_opencl_version(ctx, &err_internal);
ccl_if_err_propagate_goto(err, err_internal, error_handler);
/* Create the OpenCL sampler object. */
if (ocl_ver >= 200) {
/* Platform is OpenCL >= 2.0, use "new" API. */
sampler = clCreateSamplerWithProperties(
ccl_context_unwrap(ctx), sampler_properties, &ocl_status);
} else {
/* Platform is OpenCL <= 1.2, use "old" API. */
struct ccl_sampler_basic_properties sbp =
ccl_sampler_get_basic_properties(sampler_properties);
CCL_BEGIN_IGNORE_DEPRECATIONS
sampler = clCreateSampler(ccl_context_unwrap(ctx),
sbp.normalized_coords, sbp.addressing_mode, sbp.filter_mode,
&ocl_status);
CCL_END_IGNORE_DEPRECATIONS
}
#else
/* Create OpenCL sampler object. */
struct ccl_sampler_basic_properties sbp =
ccl_sampler_get_basic_properties(sampler_properties);
sampler = clCreateSampler(ccl_context_unwrap(ctx),
sbp.normalized_coords, sbp.addressing_mode, sbp.filter_mode,
&ocl_status);
#endif
/* Check for errors. */
ccl_if_err_create_goto(*err, CCL_OCL_ERROR,
CL_SUCCESS != ocl_status, ocl_status, error_handler,
"%s: unable to create sampler (OpenCL error %d: %s).",
CCL_STRD, ocl_status, ccl_err(ocl_status));
/* Create sampler wrapper. */
smplr = ccl_sampler_new_wrap(sampler);
/* If we got here, everything is OK. */
g_assert(err == NULL || *err == NULL);
goto finish;
error_handler:
/* If we got here there was an error, verify that it is so. */
g_assert(err == NULL || *err != NULL);
finish:
/* Return sampler wrapper. */
return smplr;
}
