3

我正在尝试调试我在 CUDA 机器上遇到的索引问题

Cuda Machine Info:

{1->{Name->Tesla C2050,时钟频率->1147000,Compute Capabilities->2.,GPU Overlap->1,最大块尺寸->{1024,1024,64},最大网格尺寸->{65535 ,65535,65535},Maximum Threads Per Block->1024,Maximum Shared Memory Per Block->49152,Total Constant Memory->65536,Warp Size->32,Maximum Pitch->2147483647,Maximum Registers Per Block->32768, Texture Alignment->512,Multiprocessor Count->14,Core Count->448,Execution Timeout->0,Integrated->False,Can Map Host Memory->True,Compute Mode->Default,Texture1D Width->65536,Texture2D宽度->65536,Texture2D 高度->65535,Texture3D 宽度->2048,Texture3D 高度->2048,Texture3D 深度->2048,Texture2D 阵列宽度->16384,Texture2D 阵列高度->16384,Texture2D 阵列切片->2048, Surface Alignment->512,Concurrent Kernels->True,ECC Enabled->True,Total Memory->2817982462},

这段代码所做的只是将 3D 数组的值设置为等于 CUDA 使用的索引：

__global __ void cudaMatExp(
float *matrix1, float *matrixStore, int lengthx, int lengthy, int lengthz){

long UniqueBlockIndex = blockIdx.y * gridDim.x + blockIdx.x;

long index = UniqueBlockIndex * blockDim.z * blockDim.y * blockDim.x +
    threadIdx.z * blockDim.y * blockDim.x + threadIdx.y * blockDim.x +
    threadIdx.x;

if (index < lengthx*lengthy*lengthz) {

matrixStore[index] =  index;

}
}

出于某种原因，一旦我的 3D 数组的维度变得太大，索引就会停止。

我尝试了不同的块尺寸（blockDim.x by blockDim.y by blockDim.z）：

8x8x8 只提供正确的索引到数组维度 12x12x12

9x9x9 只提供正确的索引到数组维度 14x14x14

10x10x10 只提供正确的索引到数组维度 15x15x15

对于大于这些的尺寸，所有不同的块大小最终会再次开始增加，但它们永远不会达到 dim^3-1 的值（这是 cuda 线程应该达到的最大索引）

以下是一些说明此行为的图表：

例如：这是在 x 轴上绘制 3D 数组的维度（即 x x x），在 y 轴上绘制在 cuda 执行期间处理的最大索引号。此特定图适用于 10x10x10 的块尺寸。

在此处输入图像描述

这是生成该图的（Mathematica）代码，但是当我运行这个时，我使用了 1024x1x1 的块尺寸：

CUDAExp = CUDAFunctionLoad[codeexp, "cudaMatExp",
  {{"Float", _,"Input"}, {"Float", _,"Output"},
    _Integer, _Integer, _Integer},
  {1024, 1, 1}]; (*These last three numbers are the block dimensions*)

max = 100; (* the maximum dimension of the 3D array *)
hold = Table[1, {i, 1, max}];
compare = Table[i^3, {i, 1, max}];
Do[
   dim = ii;
   AA  = CUDAMemoryLoad[ConstantArray[1.0, {dim, dim, dim}], Real, 
                                     "TargetPrecision" -> "Single"];
   BB  = CUDAMemoryLoad[ConstantArray[1.0, {dim, dim, dim}], Real, 
                                     "TargetPrecision" -> "Single"];

   hold[[ii]] = Max[Flatten[
                  CUDAMemoryGet[CUDAExp[AA, BB, dim, dim, dim][[1]]]]];

 , {ii, 1, max}]

ListLinePlot[{compare, Flatten[hold]}, PlotRange -> All]

这是相同的图，但现在绘制 x^3 以比较它应该在的位置。请注意，它在数组的维度大于 32 后发散

在此处输入图像描述

我测试了 3D 数组的维度，并查看了索引的距离并将其与 dim^3-1 进行比较。例如，对于 dim=32，cuda 最大索引为 32767（即 32^3 -1），但对于 dim=33，cuda 输出为 33791，而应为 35936（33^3 -1）。请注意，33791-32767 = 1024 = blockDim.x

问题：

有没有办法正确索引一个尺寸大于 Mathematica 中块尺寸的数组？

现在，我知道有些人在他们的索引方程中使用 __mul24(threadIdx.y,blockDim.x) 来防止位乘法错误，但在我的情况下似乎没有帮助。

另外，我看到有人提到您应该使用 -arch=sm_11 编译代码，因为默认情况下它是为计算能力 1.0 编译的。不过，我不知道 Mathematica 是否是这种情况。我假设 CUDAFunctionLoad[] 知道使用 2.0 功能进行编译。有谁知道吗？

任何建议都会非常有帮助！

4

1 回答 1

1

因此，Mathematica 有一种处理网格尺寸的隐藏方式，要将网格尺寸固定为可行的东西，您必须在调用的函数的末尾添加另一个数字。

该参数表示要启动的线程数（或网格维度乘以块维度）。

例如，在我上面的代码中：

CUDAExp = 
  CUDAFunctionLoad[codeexp, 
   "cudaMatExp", {
           {"Float", _, "Input"}, {"Float", _,"Output"}, 
                        _Integer, _Integer, _Integer}, 
     {8, 8, 8}, "ShellOutputFunction" -> Print];

(8,8,8) 表示块的维度。

当你调用CUDAExp[]mathematica时，你可以添加一个参数来表示要启动的线程数：

在这个例子中，我终于让它与以下内容一起工作：

// AA and BB are 3D arrays of 0 with dimensions dim^3
dim = 64;
CUDAExp[AA, BB, dim, dim, dim, 4089];

请注意，当您使用 CUDAFunctionLoad[] 进行编译时，它只需要 5 个输入，第一个是您传递给它的数组（维度dim x dim x dim），第二个是存储它的内存的位置。第三，第四和第五是维度。

当您将其传递给第 6 个时，mathematica 将其转换为gridDim.x * blockDim.x，因此，由于我知道我需要 gridDim.x = 512 才能处理数组中的每个元素，因此我将此数字设置为等于 512 * 8 = 4089。

我希望这对将来遇到此问题的人来说是清楚和有用的。

于 2011-05-31T20:18:27.143 回答