Opencl入门Demo

最近负责的几个项目需要使用opencl进行编程，进行了学习，并将学习后编写的主要Demo代码记录下来，供大家初步入门使用。

opencl的介绍，原理等这里就不说了，百度一下有很多，直接切入主题。

这个demo实现两个数组的相加操作。

1.进行平台的初始化相关操作

int initPlatform(TPlatformObject* tplatformObj)
{
    cl_int err= CL_SUCCESS;
    cl_uint num_platforms;
    char platformInfo[100];
    char deviceInfo[100];
    size_t nameLen;

    printInfo("InitPlatform\n");
    /*获取opencl执行的平台*/
    err=clGetPlatformIDs(0,NULL,&num_platforms);
    if(err != CL_SUCCESS){
        printErr("initPlatform_clGetPlatformIDS0 failed!err:%d",err);
        return err;
    }
    err=clGetPlatformIDs(1,&tplatformObj->platform,NULL);
    if(err != CL_SUCCESS){
        printErr("initPlatform_clGetPlatformIDS1 failed!err:%d",err);
        return err;
    }

    err=clGetPlatformInfo(tplatformObj->platform,CL_PLATFORM_VENDOR,100,platformInfo,&nameLen);
    if(err != CL_SUCCESS){
        printErr("initPlatform_clGetPlatforminfo failed!err:%d",err);
        return err;
    }
    printInfo("CL_PLATFORM_VENDOR:%s\n",platformInfo);

    /*获取平台执行设备*/
    err=clGetDeviceIDs(tplatformObj->platform,CL_DEVICE_TYPE_GPU,1,&tplatformObj->device,NULL);
    if(err != CL_SUCCESS){
        printErr("initPlatform_clGetPlatforminfo failed!err:%d",err);
        return err;
    }
    err=clGetDeviceInfo(tplatformObj->device,CL_DEVICE_VERSION,100,deviceInfo,&nameLen);
    printInfo("initPlatform_CL_DRIVER_VERSION:%s\n",deviceInfo);

    /*创建context*/
    tplatformObj->context = clCreateContext(NULL,1,&tplatformObj->device,NULL,NULL,&err);
    if(err != CL_SUCCESS){
        printErr("initPlatform_clCreateContext:%d",err);
        return err;
    }

    /*创建命令队列*/
    tplatformObj->queue = clCreateCommandQueue(tplatformObj->context,tplatformObj->device, 0, &err);
    if(err != CL_SUCCESS){
        printErr("initPlatform_clCreateCommandQueue:%d",err);
        return err;
    }

    return 0;
}

2.加载kernel函数，执行，直接整个文件拷贝过来了

#include "Rotate.h"

/*初始化工程*/
int initProgram(TPlatformObject* tplatformObj)
{
    FILE* fp=NULL;                /*文件指针指向*.cl内核代码文件*/
    char* program_buffer=NULL;    /*用于拷贝*.cl中的内容*/
    size_t program_size;          /**.cl文件大小*/
    cl_int errCode=0;

    fp=fopen("Operation.cl","r");
    if(fp == NULL){
        printErr("initProgram:fp is NULL\n");
        return -1;
    }
    fseek(fp,0,SEEK_END);         /*文件指针指向文件结尾*/
    program_size=ftell(fp);       /*计算出文件开头与fp指针位置之间的数据大小*/
    rewind(fp);                   /*让fp指针重新回到文件头位置*/
    program_buffer=(char*)malloc(program_size+1);   /*分配内存*/
    program_buffer[program_size]='\0';              /*添加结尾标志*/
    fread(program_buffer,sizeof(char),program_size,fp);
    fclose(fp);

    /*创建程序对象*/
    tplatformObj->program=clCreateProgramWithSource(tplatformObj->context,
        1,(const char**) &program_buffer,&program_size,&errCode);
    if(errCode != CL_SUCCESS){
        printErr("initProgram-Couldn't create the program,errcode=%d\n",errCode);
        free(program_buffer);
        program_buffer=NULL;
        return -1;
    }
    free(program_buffer);
    program_buffer=NULL;

    /*构建程序执行体*/
    errCode=clBuildProgram(tplatformObj->program,0,NULL,NULL,NULL,NULL);
    if(errCode != CL_SUCCESS){
        printErr("initProgram_clBuildProgram failed!errcode is :%d\n",errCode);
        return -1;
    }

    return 0;
}

/*获取创建的kernel函数句柄*/
int CreateKernel(TPlatformObject* tplatformObj,TRotObj* tRotObj)
{
    cl_int errCode=0;
    tRotObj->VecAdd=clCreateKernel(tplatformObj->program,"vecadd",&errCode);
    if(errCode != CL_SUCCESS){
        printErr("CreateKernel failed! errCode:%d\n",errCode);
        return -1;
    }
    return 0;
}

/*创建buffer*/
int CreateBuffer(TPlatformObject* tplatformObj,TRotObj* tRotObj)
{
    cl_int errCode=0;
    tRotObj->cl_A=clCreateBuffer(tplatformObj->context,CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
        1024*sizeof(float),(void*)tRotObj->A,&errCode);
    if(errCode != CL_SUCCESS){
        printErr("clCreateBuffer cl_A failed!!");
    }

    tRotObj->cl_B=clCreateBuffer(tplatformObj->context,CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
        1024*sizeof(float),(void*)tRotObj->B,&errCode);
    if(errCode != CL_SUCCESS){
        printErr("clCreateBuffer cl_B failed!!");
    }

    tRotObj->cl_C=clCreateBuffer(tplatformObj->context,CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
        1024*sizeof(float),(void*)tRotObj->C,&errCode);
    if(errCode != CL_SUCCESS){
        printErr("clCreateBuffer cl_B failed!!");
    }

    return 0;
}

/*为kernel函数设置参数*/
int SetKernelArgs(TRotObj* tRotObj)
{
    cl_int errCode=CL_SUCCESS;
    errCode=clSetKernelArg(tRotObj->VecAdd,0,sizeof(cl_mem),&tRotObj->cl_A);
    errCode=clSetKernelArg(tRotObj->VecAdd,1,sizeof(cl_mem),&tRotObj->cl_B);
    errCode=clSetKernelArg(tRotObj->VecAdd,2,sizeof(cl_mem),&tRotObj->cl_C);
    return errCode;
}

int RotateOpen(TPlatformObject* tplatformObj,TRotObj* tRotObj)
{
    int ret=0;
    int i=0;
    ret=initProgram(tplatformObj);
    if(ret != CL_SUCCESS){
        printErr("RotateOpen initProgram failed!!");
    }
    printInfo("initProgram done!!\n");

    /*创建kernel可执行函数*/
    if(CreateKernel(tplatformObj,tRotObj) != CL_SUCCESS){
        printErr("RotateOpen CreateKernel failed!!");
        return -1;
    }

    /*测试用数组*/
    float* A=(float*)malloc(1024*sizeof(float));
    float* B=(float*)malloc(1024*sizeof(float));
    float* C=(float*)malloc(1024*sizeof(float));
    memset(A,1,sizeof(float)*1024);
    memset(B,2,sizeof(float)*1024);
    for(i=0;i<1024;i++){
        A[i]=i;
        B[i]=i;
    }
    memset(C,0,sizeof(float)*1024);
    tRotObj->A=A;
    tRotObj->B=B;
    tRotObj->C=C;
    /*创建buffer*/
    if(CreateBuffer(tplatformObj,tRotObj) != 0){
        printErr("RotateOpen CreateBuffer failed!!");
        return -1;
    }

    /*设置kernel的参数*/
    if(SetKernelArgs(tRotObj)){
        printErr("RotateOpen SetKernelArgs failed!!");
        return -1;
    }

    /*执行kernel函数*/
    cl_uint work_dim=1; /*工作项维数*/
    size_t global_work_size=1024;
    ret=clEnqueueNDRangeKernel(tplatformObj->queue,tRotObj->VecAdd,work_dim,
        NULL,&global_work_size,NULL,0,NULL,NULL);
    if(ret != CL_SUCCESS){
        printErr("RotateOpen clEnqueueNDRangeKernel failed!!");
        return -1;
    }
    clFinish(tplatformObj->queue);/*执行结束*/

    /*读取计算出的结果*/
    ret=clEnqueueReadBuffer(tplatformObj->queue,tRotObj->cl_C,CL_TRUE,0,1024*4,tRotObj->C,0,NULL,NULL);

    /*打印程序结果*/
    for(i=0;i<global_work_size;i++){
        printf("%f ",tRotObj->C[i]);
        if((i+1)%16==0){
            printf("\n");
        }
    }

    /*释放内存*/
    clReleaseCommandQueue(tplatformObj->queue);
    clReleaseContext(tplatformObj->context);
    clReleaseDevice(tplatformObj->device);
    clReleaseKernel(tRotObj->VecAdd);
    clReleaseMemObject(tRotObj->cl_A);
    clReleaseMemObject(tRotObj->cl_B);
    free(tRotObj->A);
    free(tRotObj->B);
    free(tRotObj->C);
    return ret;
}