Есть ли способ заставить CUDA nvprof включать вызовы функций, такие как malloc, в свой статистический профилировщик?
Я пытался улучшить производительность своего приложения,Естественно, я использовал nvprof в качестве инструмента в этих усилиях.
В последнее время, пытаясь уменьшить объем памяти, занимаемой графическим процессором, моего приложения, я написал код, который потребовал вдвое больше времени для запуска,Однако новый код, вызвавший замедление, обнаруживался в профилировщике только в небольшом количестве (выборка команд показала, что около 10% времени тратится на новый код, но наивная мысль будет указывать, что 50% времени должно было быть потрачено на новый код).Возможно, новый код вызвал больше кеша, возможно, поместив реализацию в файл заголовка, чтобы он мог быть встроен, чтобы запутать профилировщик и т. Д. Однако, без всякой уважительной причины, я подозревал, что вызовы нового кода malloc.
Действительно, после того, как я сократил количество вызовов malloc, моя производительность возросла почти до того уровня, который был до включения нового кода.
Это привело меня к аналогичному вопросу , почемуРазве вызовы malloc не отображаются в статистическом профилировщике? Являются ли вызовы malloc чем-то вроде системного вызова графического процессора, который не наблюдается?
Ниже приведен примерПрограмма и ее программа, которая демонстрирует эту конкретную проблему.
#include <iostream>
#include <numeric>
#include <thread>
#include <stdlib.h>
#include <stdio.h>
static void CheckCudaErrorAux (const char *, unsigned, const char *, cudaError_t);
#define CUDA_CHECK_RETURN(value) CheckCudaErrorAux(__FILE__,__LINE__, #value, value)
__global__ void countup()
{
long sum = 0;
for (long i = 0; i < (1 << 23); ++i) {
sum += i;
}
printf("sum is %li\n", sum);
}
__global__ void malloc_a_lot() {
long sum = 0;
for (int i = 0; i < (1 << 17) * 3; ++i) {
int * v = (int *) malloc(sizeof(int));
sum += (long) v;
free(v);
}
printf("sum is %li\n", sum);
}
__global__ void both() {
long sum = 0;
for (long i = 0; i < (1 << 23); ++i) {
sum += i;
}
printf("sum is %li\n", sum);
sum = 0;
for (int i = 0; i < (1 << 17) * 3; ++i) {
int * v = (int *) malloc(sizeof(int));
sum += (long) v;
free(v);
}
printf("sum is %li\n", sum);
}
int main(void)
{
CUDA_CHECK_RETURN(cudaDeviceSynchronize());
std::chrono::time_point<std::chrono::system_clock> t1 = std::chrono::system_clock::now();
countup<<<8,1>>>();
CUDA_CHECK_RETURN(cudaDeviceSynchronize());
std::chrono::time_point<std::chrono::system_clock> t2 = std::chrono::system_clock::now();
malloc_a_lot<<<8,1>>>();
CUDA_CHECK_RETURN(cudaDeviceSynchronize());
std::chrono::time_point<std::chrono::system_clock> t3 = std::chrono::system_clock::now();
both<<<8,1>>>();
CUDA_CHECK_RETURN(cudaDeviceSynchronize());
std::chrono::time_point<std::chrono::system_clock> t4 = std::chrono::system_clock::now();
std::chrono::duration<double> duration_1_to_2 = t2 - t1;
std::chrono::duration<double> duration_2_to_3 = t3 - t2;
std::chrono::duration<double> duration_3_to_4 = t4 - t3;
printf("timer for countup() took %.3lf\n", duration_1_to_2.count());
printf("timer for malloc_a_lot() took %.3lf\n", duration_2_to_3.count());
printf("timer for both() took %.3lf\n", duration_3_to_4.count());
return 0;
}
static void CheckCudaErrorAux (const char *file, unsigned line, const char *statement, cudaError_t err)
{
if (err == cudaSuccess)
return;
std::cerr << statement<<" returned " << cudaGetErrorString(err) << "("<<err<< ") at "<<file<<":"<<line << std::endl;
exit (1);
}
Элитная версия результатов:
sum is 35184367894528...
sum is -319453208467532096...
sum is 35184367894528...
sum is -319453208467332416...
timer for countup() took 4.034
timer for malloc_a_lot() took 4.306
timer for both() took 8.343
Результат профилирования показан на следующем рисунке.Числа, которые отображаются при наведении курсора на светло-синие столбцы, соответствуют размеру столбцов.В частности, строка 41 имеет 16,515,077 выборок, связанных с ней, но строка 47 имеет только 633,996 выборок.
both Kernel Profile - PC Sampling">
Кстати, программа выше скомпилирована с отладочной информациейи, по-видимому, без оптимизации - режим «Debug» по умолчанию для компиляции в Nsight Eclipse.Если я компилирую в режиме «Release», оптимизация вызывается, и продолжительность вызова countup() очень близка к 0 секундам.