0

我有两个内核（它们都只运行一次，因此示例中的 globalWorkSize 为 1）：

第一个内核 ( kernel_Calc()) 计算一些值并将它们存储在__global内存中。在这个例子中，它计算（设置转换 3D 空间中的一个点的转换矩阵）一个转换矩阵并转换原点。

inline 
float4 mul( const float4 M[ 4 ], const float4 v)
{
   float4 r;
   r.x = dot( v, M[ 0 ] );
   r.y = dot( v, M[ 1 ] );
   r.z = dot( v, M[ 2 ] );
   r.w = dot( v, M[ 3 ] );
   return r;
}

__kernel
void kernel_Calc( __global float4* g_TransformationMatrices, __global float3* g_Point3D )
{
    __private float4 transformationMatrix[ 4 ];

    transformationMatrix [ 0 ] = (float4) ( 1.0f, 0.0f, 0.0f, 0.0f );
    transformationMatrix [ 1 ] = (float4) ( 0.0f, 1.0f, 0.0f, 10.0f );
    transformationMatrix [ 2 ] = (float4) ( 0.0f, 0.0f, 1.0f, 0.0f );
    transformationMatrix [ 3 ] = (float4) ( 0.0f, 0.0f, 0.0f, 1.0f );

    g_TransformationMatrices[ 0 ] = transformationMatrix[ 0 ];
    g_TransformationMatrices[ 1 ] = transformationMatrix[ 1 ];
    g_TransformationMatrices[ 2 ] = transformationMatrix[ 2 ];
    g_TransformationMatrices[ 3 ] = transformationMatrix[ 3 ];


    float4 point4D = (float4) ( 0.0f, 0.0f, 0.0f, 1.0f );
    float4 point4DTransformed = mul( transformationMatrix, point4D);

    g_Point3D[ 0 ] = (float3) ( point4DTransformed.x / point4DTransformed.w ,
                                point4DTransformed.y / point4DTransformed.w ,
                                point4DTransformed.z / point4DTransformed.w );
}

在主机端，我将计算出的__global缓冲区复制到具有函数的__constant缓冲区（CL_MEM_READ_ONLY缓冲区）中clEnqueueCopyBuffer()。（我这样做是因为我希望从__constant内存中读取比从内存中读取更快__global。缓冲区复制可以在设备端使用此功能完成，而无需复制__global回主机然后将其复制到__constant.)

第二个内核 ( kernel_Test()) 尝试将计算值加载到可以在主机端读取的__global变量 ( ) 中。__global float4* test这sizeStruct是一个用户定义的结构，它只包含一个整数（这是矩阵和转换点的数量）。第二个和第三个参数是内存中被函数__constant填满的缓冲区。clEnqueueCopyBuffer()

struct sizeStruct
{
    int m_Size;
};

__kernel 
void kernel_Test( __constant struct sizeStruct* c_SS,
                  __constant float4* c_TransformationMatrices,
                  __constant float3* c_Points3D,
                  __global float4 *test )
{                   
    test[ 0 ] = c_TransformationMatrices[ 0 ];
    test[ 1 ] = c_TransformationMatrices[ 1 ];
    test[ 2 ] = c_TransformationMatrices[ 2 ];
    test[ 3 ] = c_TransformationMatrices[ 3 ];
}

问题是当我运行程序时，测试变量包含以下内容：

1.000000, 0.000000, 0.000000, 0.000000
0.000000, 0.000000, 0.000000, 0.000000
0.000000, 0.000000, 0.000000, 0.000000
0.000000, 0.000000, 0.000000, 0.000000

但它应该包含：

1.000000, 0.000000, 0.000000, 0.000000
0.000000, 1.000000, 0.000000, 10.000000
0.000000, 0.000000, 1.000000, 0.000000
0.000000, 0.000000, 0.000000, 1.000000

我检查了__constant变量（通过将它们复制到主机内存）它们包含正确的数据。该代码是我的程序的简化版本。这就是它可能包含不必要的操作和参数的原因。该示例已按照我的描述进行了测试和工作。

更有趣的是，当我将__constant float3* c_Points3D内核参数更改为__global float3* c_Points3D内核参数（但仍使用已填充clEnqueueCopyBuffer()函数的 read_only 缓冲区）时，它工作正常。__constant struct sizeStruct* c_SS当我删除参数时它也有效。因此，kernel_Test 参数的地址空间似乎有问题，但问题出现在__constant->__global复制。

我在 nvidia geforce gtx 690 上运行程序，但我可以将设备（和平台）更改为 intel i7-3930k（使用 intel sdk）。使用 intel-i7 cpu 一切正常，内核代码没有任何变化。

Q1：为什么会出现这种奇怪的行为？有人知道我在做什么错吗？

Q2：使用地址空间限定符创建缓冲区cl_mem_read_only和使用 with是否合法？__global

4

1 回答 1

0

Q1：为什么会出现这种奇怪的行为？有人知道我在做什么错吗？

没有什么明显的想法，你能发布一个在 Windows 上编译的完整的最小工作示例吗？我想在 AMD Radeon GPU 上进行测试。

Q2：使用 cl_mem_read_only 创建缓冲区并使用 __global 地址空间限定符是否合法？

是的，这是合法的，尽管您需要添加const以指定缓冲区是只读的，请参阅 OpenCL 1.2 规范的第 6.5.1 节。

于 2013-07-28T01:56:50.620 回答