CUDA - memcpy2d - неправильный шаг

Question

Я только начал программировать на CUDA и пытался выполнить код, показанный ниже.Идея состоит в том, чтобы скопировать 2-мерный массив на устройство, вычислить сумму всех элементов и впоследствии извлечь сумму (я знаю, что этот алгоритм не распараллелен. На самом деле он выполняет больше работы, чем необходимо. Однако это просто предназначенокак практика для memcopy).

#include<stdio.h>
#include<cuda.h>
#include <iostream>
#include <cutil_inline.h>

#define height 50
#define width 50

using namespace std;

// Device code
__global__ void kernel(float* devPtr, int pitch,int* sum)
{
int tempsum = 0;    
for (int r = 0; r < height; ++r) {
        int* row = (int*)((char*)devPtr + r * pitch);
        for (int c = 0; c < width; ++c) {
             int element = row[c];
             tempsum = tempsum + element;
        }
    }
*sum = tempsum;
}

//Host Code
int main()
{

int testarray[2][8] = {{4,4,4,4,4,4,4,4},{4,4,4,4,4,4,4,4}};
int* sum =0;
int* sumhost = 0;
sumhost = (int*)malloc(sizeof(int));

cout << *sumhost << endl;

float* devPtr;
size_t pitch;
cudaMallocPitch((void**)&devPtr, &pitch, width * sizeof(int), height);
cudaMemcpy2D(devPtr,pitch,testarray,0,8* sizeof(int),4,cudaMemcpyHostToDevice);

cudaMalloc((void**)&sum, sizeof(int));
kernel<<<1, 4>>>(devPtr, pitch, sum);
cutilCheckMsg("kernel launch failure");
cudaMemcpy(sumhost, sum, sizeof(int), cudaMemcpyDeviceToHost);

cout << *sumhost << endl;

return 0;
}

Этот код прекрасно компилируется (в версии-кандидате SDK 4.0).Однако, как только я пытаюсь выполнить, я получаю

0
cpexample.cu(43) : cutilCheckMsg() CUTIL CUDA error : kernel launch failure : invalid pitch argument.

, что вызывает сожаление, так как я понятия не имею, как это исправить ;-(. Насколько я знаю, высота тона - это смещение в памятичтобы обеспечить более быстрое копирование данных. Однако такой шаг используется только в памяти устройства, а не в памяти хоста, не так ли? Поэтому шаг моей памяти хоста должен быть 0, не так ли?

Кроме того, я также хотел бы задать еще два вопроса:

• Если я объявляю переменную типа int * sumhost (см. Выше), куда этот указатель указывает? Сначала на память хоста ипосле cudaMalloc в память устройства?
• cutilCheckMsg был очень удобен в этом случае. Существуют ли похожие функции для отладки, о которых я должен знать?

Answer 1

В этой строке вашего кода:

cudaMemcpy2D(devPtr,pitch,testarray,0,8* sizeof(int),4,cudaMemcpyHostToDevice);

Вы говорите, что значение исходного шага для testarray равно 0, но как это возможно, если формула для высоты тона равна T* elem = (T*)((char*)base_address + row * pitch) + column? Если мы заменим значение 0 для высоты тона в этой формуле, мы не получим правильные значения при поиске адреса с некоторым двумерным (x, y) смещением упорядоченной пары. Следует учитывать, что правило для значения высоты тона pitch = width + padding. На хосте отступ часто равен 0, но ширина не равна 0, если в вашем массиве нет ничего. Со стороны аппаратного обеспечения может быть дополнительное заполнение, поэтому значение шага может не совпадать с объявленной шириной массива. Поэтому вы можете сделать вывод, что pitch >= width в зависимости от значения заполнения. Таким образом, даже на стороне хоста значение для шага источника должно быть не меньше размера каждой строки в байтах, то есть в случае testarray оно должно быть 8*sizeof(int). Наконец, высота вашего 2D-массива в хосте также составляет всего 2 строк, а не 4.

В качестве ответа на ваш вопрос о том, что происходит с выделенными указателями, если вы назначаете указатель с помощью malloc(), тогда указателю присваивается значение адреса, которое находится в памяти хоста. Таким образом, вы можете разыменовать его на стороне хоста, но не на стороне устройства. С другой стороны, указатель, выделенный с помощью cudaMalloc(), получает указатель на память, находящуюся на устройстве. Поэтому, если вы разыменуете его на хосте, он не указывает на выделенную память на хосте, и это приведет к непредсказуемым результатам. Вполне допустимо передать этот адрес указателя ядру на устройстве, поскольку, когда он разыменовывается на стороне устройства, он указывает на память, локально доступную для устройства. В целом среда выполнения CUDA разделяет эти две области памяти, предоставляя функции копирования памяти, которые будут копировать данные между указателем и устройством и обратно, и использовать значения адресов из этих указателей в качестве источника и / или места назначения для копирования в зависимости от желаемого направления. (хост-устройство или устройство-хост). Теперь, если вы возьмете тот же int* и сначала выделите его с помощью malloc(), а затем (после, надеюсь, вызова free() на указателе) с cudaMalloc(), ваш указатель сначала будет иметь адрес, который указывает на память хоста а затем память устройства. Вам нужно будет отслеживать его состояние, чтобы избежать непредсказуемых результатов от разыменования адреса, находящегося на устройстве или хосте, в зависимости от того, была ли разыменована ссылка в коде хоста или коде устройства.