CUDAFunctionLoad в Mathematica - проблема индексации - PullRequest
Новая
       25

CUDAFunctionLoad в Mathematica - проблема индексации

3 голосов
/ 28 мая 2011

Я пытаюсь отладить проблему с индексом, которая возникла на моей машине CUDA 

Cuda Machine Info:

{1 -> {Name-> Tesla C2050, тактовая частота-> 1147000, вычислительные возможности-> 2., перекрытие GPU-> 1, максимальные размеры блока -> {1024,1024,64}, максимальные размеры сетки- > {65535,65535,65535}, Максимальное количество потоков в блоке -> 1024, Максимальная общая память в блоке -> 49152, Общая постоянная память -> 65536, Размер деформации-> 32, Максимальный шаг -> 2147483647, Максимальное количество регистров в блоке - > 32768, выравнивание текстуры-> 512, счетчик мультипроцессора-> 14, число ядер-> 448, тайм-аут выполнения-> 0, интегрировано-> ложно, может отобразить память хоста-> True, режим вычисления-> по умолчанию, ширина Texture1D-> 65536, Ширина Texture2D-> 65536, Высота Texture2D-> 65535, Ширина Texture3D-> 2048, Высота Texture3D-> 2048, Глубина Texture3D-> 2048, Ширина массива Texture2D-> 16384, Высота массива Texture2D-> 16384, Срезы массива Texture2D- > 2048, выравнивание поверхности-> 512, параллельные ядра-> True, ECC включен-> True, общий объем памяти-> 2817982462},

Все, что делает этот код, устанавливает значения трехмерного массива, равные индексу, который использует 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;

}
}

По какой-то причине, когда размер моего трехмерного массива становится слишком большим, индексация прекращается.

Я пробовал разные размеры блока (blockDim.x от blockDim.y от blockDim.z):

8x8x8 дает правильное индексирование только до размера массива 12x12x12

9x9x9 дает только правильное индексирование до размера массива 14x14x14

10x10x10 дает правильное индексирование только до размера массива 15x15x15

Для размеров, превышающих эти, все размеры блоков разных размеров в конечном итоге снова начинают увеличиваться, но они никогда не достигают значения dim ^ 3-1 (что является максимальным индексом, которого должен достичь поток cuda)

Вот несколько графиков, иллюстрирующих это поведение:

Например: на оси x отображается размер трехмерного массива (который равен x x x), а на оси y - максимальное число индекса, которое обрабатывается во время выполнения cuda. Этот конкретный график предназначен для размеров блока 10x10x10.

enter image description here

Вот код (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

enter image description here

Я проверяю размеры трехмерного массива и смотрю, как далеко идет индексация, и сравниваю его с dim ^ 3-1. Например. для dim = 32 индекс cuda max равен 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. Кто-нибудь знает?

Любые предложения будут чрезвычайно полезны!

1 Ответ

1 голос
/ 01 июня 2011

Итак, в 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.

Надеюсь, в будущем это станет понятным и полезным для тех, кто сталкивается с этой проблемой.

...