使用AM5728/AM57XX openCL实现矩阵转置

下面对4x4矩阵进行转置,调用openCL,4X4矩阵采用二维数组进行存储,在程序设计上,让转置过程分10次转置完成,就是一次转一行,因此OpenCL的工作维数是二维。
  openCL程序分为两个部份,一部份是内核代码,负责具体算法。另一部份是主程序负责初始化OpenCL和准备数据。主程序加载内核代码,并按照既定方法进行运算。

核函数代码如下
__kernel void matrix_transposition(__global int* a, __global int* b)
{
    int col = get_global_id(0);
    int row = get_global_id(1);
  
    b[col*4+row] = a[row*4+col]; 
}

主机端代码如下
int main()
{
    cl_device_id device;
    cl_platform_id platform_id = NULL;
    cl_context context;
    cl_command_queue cmdQueue;
    cl_mem bufferA,bufferB;
    cl_program program;
    cl_kernel kernel = NULL;
    
    size_t ocl_string_size;
    char *ocl_string;


    ocl_string = (char *)malloc(1024*1024);
  
    //我们使用的是二维向量
    //设定向量大小(维数)
    size_t globalWorkSize[2];
    globalWorkSize[0] = dim_x ;
    globalWorkSize[1] = dim_y;
  
  
    /*
        定义输入变量和输出变量,并设定初值
    */
    int buf_A[dim_x][dim_y];
    int buf_B[dim_x][dim_y];
  
    size_t datasize = sizeof(int) * dim_x * dim_y;
  
    int n=0;
    int m=0;
  
    for(n=0;n<dim_x;n++)
    {
        for(m=0;m<dim_y;m++)
        {
            buf_A[m][n] = m + n*dim_x;
        }
    }
  
    ocl_string_size = get_ocl_string("matrix_transposition.cl", ocl_string);
    //step 1:初始化OpenCL
    clGetPlatformIDs(1, &platform_id, NULL);
    clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_ACCELERATOR, 1, &device, NULL);
  
    //step 2:创建上下文
    context = clCreateContext(NULL,1,&device,NULL,NULL,NULL);
  
    //step 3:创建命令队列
    cmdQueue = clCreateCommandQueue(context,device,0,NULL);
  
    //step 4:创建数据缓冲区
    bufferA = clCreateBuffer(context,
                             CL_MEM_READ_ONLY,
                             datasize,NULL,NULL);
  
    bufferB = clCreateBuffer(context,
                             CL_MEM_WRITE_ONLY,
                             datasize,NULL,NULL);
  
    //step 5:将数据上传到缓冲区
    clEnqueueWriteBuffer(cmdQueue,
                         bufferA,CL_FALSE,
                         0,datasize,
                         buf_A,0,
                         NULL,NULL);
  
    //step 6:加载编译代码,创建内核调用函数
    program = clCreateProgramWithSource(context,1,
                                        (const char**)&ocl_string,
                                        &ocl_string_size,NULL);
  
    clBuildProgram(program,1,&device,NULL,NULL,NULL);
  
    kernel = clCreateKernel(program,"matrix_transposition",NULL);
  
    //step 7:设置参数,执行内核
    clSetKernelArg(kernel,0,sizeof(cl_mem),&bufferA);
    clSetKernelArg(kernel,1,sizeof(cl_mem),&bufferB);
  
    //<span style="color: #ff0000;"><strong>注意这里第三个参数已经改成2,表示二维数据。</strong></span>
    clEnqueueNDRangeKernel(cmdQueue,kernel,
                           2,NULL,
                           globalWorkSize,
                           NULL,0,NULL,NULL);
  
    //step 8:取回计算结果
    clEnqueueReadBuffer(cmdQueue,bufferB,CL_TRUE,0,
                        datasize,buf_B,0,NULL,NULL);
  
    //输出计算结果
    for(n=0;n<dim_x;n++)
    {
        for(m=0;m<dim_y;m++)
        {
            printf("%d ", buf_A[m][n]);
        }
        printf("\n");
    }
  
  
    for(n=0;n<dim_x;n++)
    {
        for(m=0;m<dim_y;m++)
        {
            printf("%d ", buf_B[m][n]);
        }
        printf("\n");
    }
  
    //释放所有调用和内存
  
    clReleaseKernel(kernel);
    clReleaseProgram(program);
    clReleaseCommandQueue(cmdQueue);
    clReleaseMemObject(bufferA);
    clReleaseMemObject(bufferB);
    clReleaseContext(context);
  
  
    return 0;
}
将上述程序编译,结果如下
0 1 2 3
4 5 6 7
8 9 10 11
12 13 14 25
矩阵转置后
0 4 8 12
1 5 9 13
2 6 10 14
3 7 11 15
该矩阵的行和列转换成功,即第 n 行变成第 n 列

粽子糖果 发表于11-29 09:57 浏览65535次
分享到:

已有0条评论

暂时还没有回复哟,快来抢沙发吧

添加一条新评论

只有登录用户才能评论,请先登录注册哦!

话题作者

粽子糖果
粽子糖果(总统)
金币:41624个|学分:51977个
立即注册
畅学电子网,带你进入电子开发学习世界
专业电子工程技术学习交流社区,加入畅学一起充电加油吧!

x

畅学电子网订阅号