Простое добавление ядра cuda: незаконная память после 2432 вызовов ядра

Question

Я создаю простое ядро ​​cuda, которое вычисляет сумму элементов. Каждый поток добавляет входное значение в выходной буфер. Каждый поток вычисляет одно значение. Используется 2432 потока (19 блоков * 128 потоков).

Выходной буфер остается прежним, указатель входного буфера смещается на количество потоков после каждого выполнения ядра. Таким образом, у нас есть al oop, вызывающее ядро ​​добавления, пока мы не вычислим все входные данные.

Пример: все мои входные значения установлены на 1. Размер выходного буфера равен 2432. Размер входного буфера равен 2432 * 2000. 2000 раз вызывается ядро ​​добавления, чтобы добавить 1 к каждому полю вывода. Конечный результат вывода - 2000 для каждого поля. Я вызываю агрегат функций, который содержит for l oop, вызывая ядро ​​столько раз, сколько необходимо для передачи полных входных данных. До сих пор это работает, если я не вызываю ядро ​​слишком часто.

Однако, если я вызываю ядро ​​2500 раз, я получаю ошибку незаконного доступа к памяти cuda. ​​ профилирование nsight cuda

Как видите, время выполнения последнего успешного ядра увеличивается на 3 порядка. Впоследствии мои указатели становятся недействительными, и следующие вызовы приводят к CudaErrorIllegalAdress.

Я очистил код, чтобы получить минимальный рабочий пример:

 #include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <vector>
#include <stdio.h>
#include <iostream>

using namespace std;

template <class T> __global__ void addKernel_2432(int *in, int * out)
{
    int i = blockIdx.x * blockDim.x  + threadIdx.x;
    out[i] = out[i] + in[i];
}


static int aggregate(int* array, size_t size, int* out) {

    

    size_t const vectorCount = size / 2432;
    cout << "ITERATIONS: " << vectorCount << endl;
    
    
    for (size_t i = 0; i < vectorCount-1; i++)
    {

         addKernel_2432<int><<<19,128>>>(array, out);
        
        array += vectorCount;
       
    }
    addKernel_2432<int> << <19, 128 >> > (array, out);
    return 1;
    }

    int main()
    {
  
    int* dev_in1 = 0;
    size_t vectorCount = 2432;
    int * dev_out = 0;
    size_t datacount = 2432*2500;
   
    std::vector<int> hostvec(datacount);
   
    //create input buffer, filled with 1
    std::fill(hostvec.begin(), hostvec.end(), 1);
    
    //allocate input buffer and output buffer
    cudaMalloc(&dev_in1, datacount*sizeof(int));
    cudaMalloc(&dev_out, vectorCount * sizeof(int));

    //set output buffer to 0
    cudaMemset(dev_out, 0, vectorCount * sizeof(int));

    //copy input buffer to GPU
    cudaMemcpy(dev_in1, hostvec.data(), datacount * sizeof(int), cudaMemcpyHostToDevice);
    
    //call kernel datacount / vectorcount times
    aggregate(dev_in1, datacount, dev_out);
    
    //return data to check for corectness
    cudaMemcpy(hostvec.data(), dev_out, vectorCount*sizeof(int), cudaMemcpyDeviceToHost);
   
    if (cudaSuccess != cudaMemcpy(hostvec.data(), dev_out, vectorCount * sizeof(int), cudaMemcpyDeviceToHost))
    {
        cudaError err = cudaGetLastError();
        cout << " CUDA ERROR: " << cudaGetErrorString(err) << endl;
    }
    else
    {
        cout << "NO CUDA ERROR" << endl;
        cout << "RETURNED SUM DATA" << endl;
        for (int i = 0; i < 2432; i++)
        {
            cout << hostvec[i] << " ";
        }

    }
   
    cudaDeviceReset();
    return 0;
}

Если вы скомпилируете и запустите его, вы получите ошибка. Изменение:

size_t datacount = 2432 * 2500;

to

size_t datacount = 2432 * 2400;

, и это дает правильные результаты.

Я ищу любые идеи, почему он ломается после 2432 вызовов ядра.

То, что я обнаружил до сих пор в поисковой системе: Неправильный набор целевой архитектуры. Я использую 1070ti. Моя цель установлена ​​на: compute_61, sm_61 В свойствах проекта Visual Studio. Это ничего не меняет.

Я что-то пропустил? Есть ли ограничение, сколько раз можно вызывать ядро, пока cuda не аннулирует указатель? Спасибо за помощь. Я использовал windows, Visual Studio 2019 и среду выполнения CUDA 11.

Это результат в обоих случаях. Успехи и неудачи:

[ успех 2400 элементов

Ошибка: [ ошибка 2500 элементов

Answer 1

static int aggregate(int* array, size_t size, int* out) {
    size_t const vectorCount = size / 2432;
    for (size_t i = 0; i < vectorCount-1; i++)
    {
        array += vectorCount;
    }
}

Это не vectorCount, а количество итераций, на которое вы случайно увеличили. Прекрасно работает, пока vectorCount <= 2432 (но дает неверные результаты), и приводит к переполнению буфера выше.

array += 2432 - это то, что вы намеревались написать.