Я тестировал новый CUDA 8 вместе с графическим процессором Pascal Titan X и ожидаю ускорения для моего кода, но по какой-то причине он заканчивается медленнее. Я на Ubuntu 16.04.
Вот минимальный код, который может воспроизвести результат:
CUDASample.cuh
class CUDASample{
public:
void AddOneToVector(std::vector<int> &in);
};
CUDASample.cu
__global__ static void CUDAKernelAddOneToVector(int *data)
{
const int x = blockIdx.x * blockDim.x + threadIdx.x;
const int y = blockIdx.y * blockDim.y + threadIdx.y;
const int mx = gridDim.x * blockDim.x;
data[y * mx + x] = data[y * mx + x] + 1.0f;
}
void CUDASample::AddOneToVector(std::vector<int> &in){
int *data;
cudaMallocManaged(reinterpret_cast<void **>(&data),
in.size() * sizeof(int),
cudaMemAttachGlobal);
for (std::size_t i = 0; i < in.size(); i++){
data[i] = in.at(i);
}
dim3 blks(in.size()/(16*32),1);
dim3 threads(32, 16);
CUDAKernelAddOneToVector<<<blks, threads>>>(data);
cudaDeviceSynchronize();
for (std::size_t i = 0; i < in.size(); i++){
in.at(i) = data[i];
}
cudaFree(data);
}
main.cpp
std::vector<int> v;
for (int i = 0; i < 8192000; i++){
v.push_back(i);
}
CUDASample cudasample;
cudasample.AddOneToVector(v);
Единственное отличие - флаг NVCC, который для Pascal Titan X:
-gencode arch=compute_61,code=sm_61-std=c++11;
а для старого Maxwell Titan X:
-gencode arch=compute_52,code=sm_52-std=c++11;
EDIT: Вот результаты для работы с графическим профилем NVIDIA.
Для старого Maxwell Titan время передачи памяти составляет около 205 мс, а запуск ядра составляет около 268 нас.
Для Pascal Titan время для передачи памяти составляет около 202 мс, а запуск ядра - безумно длинный 8343 нас, что заставляет меня думать, что что-то не так.
Я также выделил проблему, заменив cudaMallocManaged на старый добрый cudaMalloc и сделал некоторые профилирования и наблюдал некоторый интересный результат.
CUDASample.cu
__global__ static void CUDAKernelAddOneToVector(int *data)
{
const int x = blockIdx.x * blockDim.x + threadIdx.x;
const int y = blockIdx.y * blockDim.y + threadIdx.y;
const int mx = gridDim.x * blockDim.x;
data[y * mx + x] = data[y * mx + x] + 1.0f;
}
void CUDASample::AddOneToVector(std::vector<int> &in){
int *data;
cudaMalloc(reinterpret_cast<void **>(&data), in.size() * sizeof(int));
cudaMemcpy(reinterpret_cast<void*>(data),reinterpret_cast<void*>(in.data()),
in.size() * sizeof(int), cudaMemcpyHostToDevice);
dim3 blks(in.size()/(16*32),1);
dim3 threads(32, 16);
CUDAKernelAddOneToVector<<<blks, threads>>>(data);
cudaDeviceSynchronize();
cudaMemcpy(reinterpret_cast<void*>(in.data()),reinterpret_cast<void*>(data),
in.size() * sizeof(int), cudaMemcpyDeviceToHost);
cudaFree(data);
}
Для старого Maxwell Titan время для передачи памяти составляет около 5 мс в обоих направлениях, а запуск ядра составляет около 264 нас.
Для Pascal Titan время для передачи памяти составляет около 5 мс в обоих направлениях, а запуск ядра - около 194, что на самом деле приводит к увеличению производительности. Я надеюсь увидеть...
Почему Pascal GPU так медленно работает при запуске ядер CUDA, когда используется cudaMallocManaged? Это будет пародия, если мне нужно вернуть весь мой существующий код, который использует cudaMallocManaged в cudaMalloc. Этот эксперимент также показывает, что время передачи памяти с использованием cudaMallocManaged намного медленнее, чем использование cudaMalloc, что также кажется неправильным. Если использование этого результата приводит к медленному времени выполнения, даже код проще, это должно быть неприемлемым, поскольку вся цель использования CUDA вместо простого С++ - ускорить процесс. Что я делаю неправильно и почему я наблюдаю этот результат?
