Как работают поток CUDA, DMA Engine и Async Engine и взаимодействуют друг с другом? ⇐ C++

[Anonymous]

Я пытаюсь перекрывать передачу данных и выполнение ядра, используя CUDA C ++.
I создал массив, разделите его на 8 кусков, а затем назначьте каждый из чанков в соответствующий поток CUDA, используя следующее для цикла: < /p>

Код: Выделить всё

memset(a, 0, bytes);
for (int i = 0; i < nStreams; ++i) {
int offset = i * streamSize;
cudaMemcpyAsync(&d_a[offset], &a[offset],
streamBytes, cudaMemcpyHostToDevice,
stream[i]);
kernel(d_a, offset);
cudaMemcpyAsync(&a[offset], &d_a[offset],
streamBytes, cudaMemcpyDeviceToHost,
stream[i]);
}
< /code>
Поскольку мой графический процессор имеет только один асинхронный двигатель, теоретически, заказ передачи данных должен быть:
H2D[0] -> D2H[0] -> H2D[1] -> D2H[1] -> ...

Однако, используя систему NSIGHT для профилирования моей программы, результат подобен следующему:

Мы можем видеть, что второй h2 не заблокированным. Переводы идеально следуют теории. < /p>
Мне интересно, почему такой результат произойдет. Может ли кто -нибудь придумать значимое объяснение тому, что происходит под капюшоном? Спасибо большое!#include

__global__ void kernel(float *a, int offset)
{
int i = offset + threadIdx.x + blockIdx.x*blockDim.x;
float x = (float)i;
float s = sinf(x);
float c = cosf(x);
a = a + sqrtf(s*s+c*c);
}

int main(int argc, char **argv)
{
const int blockSize = 256, nStreams = 8;
const int n = 160 * 1024 * blockSize * nStreams;
const int streamSize = n / nStreams;
const int streamBytes = streamSize * sizeof(float);
const int bytes = n * sizeof(float);

int devId = 0;
if (argc > 1) devId = atoi(argv[1]);

cudaDeviceProp prop;
cudaGetDeviceProperties(&prop, devId);
printf("Device : %s\n", prop.name);
cudaSetDevice(devId);

// allocate pinned host memory and device memory
float *a, *d_a;
cudaMallocHost((void**)&a, bytes); // host pinned
cudaMalloc((void**)&d_a, bytes); // device

cudaStream_t stream[nStreams];
for (int i = 0; i < nStreams; ++i)
cudaStreamCreate(&stream);

// asynchronous version: loop over {copy, kernel, copy}
memset(a, 0, bytes);
for (int i = 0; i < nStreams; ++i) {
int offset = i * streamSize;
cudaMemcpyAsync(&d_a[offset], &a[offset],
streamBytes, cudaMemcpyHostToDevice,
stream);
kernel(d_a, offset);
cudaMemcpyAsync(&a[offset], &d_a[offset],
streamBytes, cudaMemcpyDeviceToHost,
stream);
}
// cleanup
for (int i = 0; i < nStreams; ++i)
cudaStreamDestroy(stream);
cudaFree(d_a);
cudaFreeHost(a);

return 0;
}
< /code>
Я попытался изменить размер куски данных, количество потоков CUDA. Результат сохраняется.>

Подробнее здесь: https://stackoverflow.com/questions/796 ... h-each-oth