【问题标题】:Issue in OpenCL Kernel functionOpenCL 内核函数中的问题
【发布时间】:2013-10-20 05:13:36
【问题描述】:

我是 Open-cl 的新手,我正在尝试为以下矩阵运算编写内核代码:

A is a 2X2 matrix:
A = [1  2] ----> row1
    [3  4] ----->row2

I need to compute: 
1) s1 = transpose(row1) X row1
2) s1 = transpose(row2) X row2
3) Sum = s1+s2

我为行级编写了内核代码(即我可以做 transpose(row1) X row1 ) - 这仅用于第一行

如何使用并行性计算每一行的值并在核函数中找到最终总和?

private static String programSource1 =
            "__kernel"+
            " void matrixMul(__global float* A, __global float* C,  int rowLength)"+
            "{"+
                "int row = get_global_id(1);"+
                "int col = get_global_id(0);"+              
                    "C[row*rowLength+col] = A[col] * A[row];"+

            "}";

【问题讨论】:

  • 最终实现是否仅适用于 2x2 矩阵数组或更大的矩阵?这完全改变了如何实现代码......

标签: opencl gpu jocl


【解决方案1】:
#define MAX_ROW_LENGTH 2 // or more 
__kernel void matrixMul(__global float* A, __global float* C,  
                                                 int rowLength)
{
     __local float buffer[MAX_ROW_LENGTH  * MAX_ROW_LENGTH];
     __local float s1[MAX_ROW_LENGTH * MAX_ROW_LENGTH];

    int col = get_global_id(0);
    int row = get_global_id(1);
    int rows = get_global_size(1);

    // read the matrix from global to local memory
    buffer[row * rowLength + col] = A[row * rowLength + col]; 
    s1[row * rowLength + col] = 0.0f;

    barrier(CLK_LOCAL_MEM_FENCE);

    for (int i = 0; i < rows; ++i)
    {
        s1[row * rowLength + col] += 
                buffer[i * rowLength + col] * buffer[i * rowLength + row];
    }
    C[row * rowLength + col] = s1[row*rowLength+col];
}

这里有一些内核代码可以满足您对小型矩阵的需求。内核使用本地内存来减少全局内存访问。对于这样的小问题(2x2 矩阵),这需要完成任何事情,但是如果您正在计算更大的矩阵,这可以稍微加快速度。然而,这是一个简短的例子,并没有优化。它有一些限制:

  • 此代码仅支持本地工作组大小等于全局 工作组大小(无块)
  • 如果矩阵变大,共享内存将限制 GPU 的利用率,并且
  • 如果您的矩阵变得非常大,它们的共享内存将不足

如果您不想删除本地内存,请将 for 循环中的缓冲区调用替换为 A 并直接写入 C 而不是 s1。

【讨论】:

  • 非常感谢,很好的解释。这让我对编写内核代码有了一个清晰的认识。
猜你喜欢
  • 1970-01-01
  • 1970-01-01
  • 2019-12-31
  • 1970-01-01
  • 1970-01-01
  • 1970-01-01
  • 2012-02-05
相关资源
最近更新 更多