内核就是程序中声明的一个函数。对于程序中的任一函数,都可以通过加上限定符__kernel将其标识为内核。内核对象中封装了程序中的某个__kernel函数以及执行此函数时所需的参数。
另外一个查询内核信息的函数:
示例程序:
1、创建内核
cl_kernel clCreateKernel (cl_program program,
const char *kernel_name,
cl_int *errcode_ret)
此函数用来创建内核对象。
- program,是一个程序对象,带有成功构建的执行体。
- kernel_name,是程序中一个声明时带有限定符__kernel的函数名。
- errcode,返回相应的错误码
- CL_INVALID_PROGRAM,program无效
- CL_INVALID_PROGRAM_EXECUTABLE,没有为program成功构建执行体。
- CL_INVALID_KERNEL_NAME,program中找不到kernel_name。
- CL_INVALID_KERNEL_DEFINITION,kernel_name的函数定义在已经具有对应执行体的那些设备上有所区别。
- CL_INVALID_VALUE,kernel_name是NULL。
- CL_OUT_OF_RESOURCES,为设备上的OpenCL操作分配内存失败。
- CL_OUT_OF_HOST_MEMORY,为主机上的OpenCL操作分配内存失败。
clCreateKernel一次只为一个内核函数创建内核对象。要是有多个内核函数时,需要一个一个的进行创建吗?
OpenCL提供了另外一个函数来解决这个问题。
cl_int clCreateKernelInProgram (cl_program program,
cl_uint num_kernels,
cl_kernel *kernels,
cl_uint *num_kernls_ret)
这个函数是为program中所有的内核函数创建对应的内核对象。如果某个__kernel函数的定义在已经具有执行体的那些设备上不完全一样,则不会为其创建内核对象。
- program,一个程序对象,具有成功构建的执行体。
- num_kernels,程序对象中内核函数的数目。
- kernels,用来存储多返回的内核对象。如果为NULL,则忽略;否则,num_kernels的值必须大于或等于program中内核的数目。
- num_kernels_ret,program中的内核的数目,如果为NULL,则忽略。
这个函数一般调用两次,第一次确定program中的内核数;第二次具体创建内核对象。下面的代码块展示了这个函数的使用方法:
cl_uint numKernels;
errNum = clCreateKernelInProgram(program, NULL, NULL, &numKernels);
cl_kernel *kernels = new cl_kernel[numKernels];
errNum = clCreateKernelInProgram(program, numKernels, kernels, NULL);
2、设置内核参数
想要执行内核,就必须设置内核参数。
cl_int clSetKernelArg (cl_kernel kernel,
cl_uint arg_index,
size_t arg_size,
const void *arg_value)
这个函数用来设置内核的某个参数。
- kernel,是内核对象。
- arg_index是参数索引。内核参数的索引值从最左边的0开始,一直到n-1,其中n是参数的总数。
- arg_size,参数的大小。这个大小由内核函数中参数的声明方式如何声明来确定。
- arg_value,传入内核函数的参数的一个指针。
3、查询和管理内核对象
cl_int clGetKernelInfo (cl_kernel kernel,
cl_kernel_info param_name,
size_t param_value_size,
void *param_value,
size_t *param_value_size_ret)
这个函数用来查询内核对象的相关信息。
- kernel,索要查询的内核对象。
- param_name,指定所要查询的信息。
- param_value,执行的内存用来存储查询结果。如果为NULL,则忽略。
- pram_value_size,param_value所指向内存块的大小。其值必须大于param_name对应的类型的大小。
- param_value_siez_ret,返回查询结果的实际大小。
这个函数也需要调用两次。
clGetKernelInfo所支持的param_name
cl_kernel_info | 返回类型 |
CL_KERNEL_FUNCTION_NAME | char [] |
返回内核函数的名字。 | |
CL_KERNEL_NUM_ARGS | cl_uint |
返回内核参数的个数。 | |
CL_KERNLE_REFERENCE_COUNT | cl_uint |
返回kernel的引用计数。 | |
CL_KERNEL_CONTEXT | cl_context |
返回kernel所在的上下文。 | |
CL_KERNEL_PROGRAM | cl_program |
返回kernel所关联的程序对象。 | |
CL_KERNEL_ATTRIBUTES | char [] |
返回程序源码中声明内核函数时所有通过__attributes__指定的特性。 |
另外一个查询内核信息的函数:
clGetKernelWorkGroupInfo (cl_kernel kernel,
cl_device_id device,
cl_kernel_work_group_info param_name,
size_t param_value_size,
void *param_value,
size_t *param_value_size_ret)
这个函数是查询内核对象针对某个设备的相关信息。
clGetKernelWorkGroupInfo所支持的param_name
cl_kernel_work_group_info | 返回类型 |
CL_KERNEL_GLOBAL_WORK_SIZE | size_t[3] |
利用此机制,应用可以查询用来在device上执行内核的全局索引空间的大小(即clEnqueueNDRangeKernel的参数global_work_size)。 这要求device是自定义设备或所执行的内核是内建的,否则clGetKernelWorkGroupInfo会返回CL_INVALID_VALUE. | |
CL_KERNEL_WORK_GROUP_SIZE | size_t |
应用可以查询用来在device上执行内核的工作组的大小。 | |
CL_KERNEL_COMPILE_WORK_GROUP_SIZE | size_t [3] |
返回限定符__arrtibute__((reqd_work_group_size(X, Y, Z))) 所指定的工作组的大小。 | |
CL_KERNEL_LOCAL_MEM_SIZE | cl_ulong |
返回内核所使用的局部内存的大小。 对于任一带有限定符__local的指针参数,如果没有为其指定内存大小,则假定为0。 | |
CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE | size_t |
返回所期望的工作组大小的粒度。仅为性能的建议。 | |
CL_KERNEL_PRIVATE_MEM_SIZE | cl_ulong |
内核中每个工作项至少要使用多少私有内存。 |
4、查询内核参数信息
cl_int clGetKernelArgInfo (cl_kernel kernel,
cl_uint arg_index,
cl_kernel_arg_info param_name,
size_t param_value_size,
void *param_value,
size_t *param_value_size_ret)
此函数会返回内核参数的相关信息。只有满足下列条件时,内核参数信息才可用:
- kernel所关联的程序对象是用clCreateProgramWithSource创建的。
- 用cl{Build | Compile}program构建程序执行体时,在参数options中指定了-cl-kernel-arg-info。
cl_kernel_arg_info的类型:
cl_kernel_arg_info | 返回类别 |
CL_KERNEL_ARG_ADDRESS_QUALIFIER | cl_kernel_arg_address_qualifier |
返回参数的地址限定符。所返回的值可以是下列之一:
| |
CL_KERNEL_ARG_ACCESS_QUALIFIER | cl_kernel_arg_access_qualifier |
返回参数的访问限定符。可以是下列值之一:
| |
CL_KERNEL_ARG_ARG_TYPE_NAME | char [] |
返回arg_index所指定的参数的类型名。 | |
CL_KERNEL_ARG_TYPE_QUALIFIER | cl_kernel_arg_type_qualifier |
返回参数的类型限定符。可能是:
| |
CL_KERNEL_ARG_NAME | char [] |
返回参数的名字。 |
示例程序:
__kernel void hello_kernel(__global const float *a,
__global const float *b,
__global float *result)
{
int gid = get_global_id(0);
result[gid] = a[gid] + b[gid];
}
kernel = clCreateKernel(program, "hello_kernel", &errNum);
if (kernel == NULL || errNum != CL_SUCCESS)
{
cerr << "Failed to create kernel." << endl;
Cleanup(context, commandQueue, program, kernel, memObjects);
return -1;
}
errNum = clSetKernelArg(kernel, 0, sizeof(cl_mem), &memObjects[0]);
errNum |= clSetKernelArg(kernel, 1, sizeof(cl_mem), &memObjects[1]);
errNum |= clSetKernelArg(kernel, 2, sizeof(cl_mem), &memObjects[2]);
if (errNum != CL_SUCCESS)
{
cerr << "Error setting kernel arguments." << endl;
Cleanup(context, commandQueue, program, kernel, memObjects);
return -1;
}