cublasSdot работает медленнее, чем cublasSgemm
В моем игрушечном примере я сначала умножаю матрицы размера 32x32, 100 000 раз, и после этого я вычисляю скалярные произведения двух векторов размера 1024, 100 000 раз снова. Для первого я использовал cublasSgemmдля второго — cublasSdot,
В результате время для первого расчета 530 msecдля второго — 10 000 msec, Однако для умножения матриц нам необходимо выполнить 32^3 операции (умножение-сложение), а для скалярного произведения просто 1024=32^2 операции.
Так почему я получаю такой результат? Вот код:
__device__ float res;
void randomInit(float *data, int size)
{
for (int i = 0; i < size; ++i)
data[i] = rand() / (float)RAND_MAX;
}
int main(){
cublasHandle_t handle;
float out;
cudaError_t cudaerr;
cudaEvent_t start1, stop1,start2,stop2;
cublasStatus_t stat;
int size = 32;
int num = 100000;
float *h_A = new float[size*size];
float *h_B = new float[size*size];
float *h_C = new float[size*size];
float *d_A, *d_B, *d_C;
const float alpha = 1.0f;
const float beta = 0.0f;
randomInit(h_A, size*size);
randomInit(h_B, size*size);
cudaMalloc((void **)&d_A, size *size *sizeof(float));
cudaMalloc((void **)&d_B, size *size * sizeof(float));
cudaMalloc((void **)&d_C, size *size * sizeof(float));
stat = cublasCreate(&handle);
cudaEventCreate(&start1);
cudaEventCreate(&stop1);
cudaEventCreate(&start2);
cudaEventCreate(&stop2);
cublasSgemm(handle, CUBLAS_OP_N, CUBLAS_OP_N, size, size, size, &alpha, d_A, size,
d_B, size, &beta, d_C, size);
cudaEventRecord(start1, NULL);
cudaMemcpy(d_A, h_A, size *size * sizeof(float), cudaMemcpyHostToDevice);
cudaMemcpy(d_B, h_B, size *size * sizeof(float), cudaMemcpyHostToDevice);
for (int i = 0; i < num; i++){
cublasSgemm(handle, CUBLAS_OP_N, CUBLAS_OP_N, size, size, size, &alpha, d_A,
size, d_B, size, &beta, d_C, size);
}
cudaMemcpy(h_C, d_C, size*size*sizeof(float), cudaMemcpyDeviceToHost);
cudaEventRecord(stop1, NULL);
cudaEventSynchronize(stop1);
float msecTotal1 = 0.0f;
cudaEventElapsedTime(&msecTotal1, start1, stop1);
std::cout <<"total time for MAtMul:" << msecTotal1 << "\n";
cudaEventRecord(start2, NULL);
cudaMemcpy(d_A, h_A, size *size * sizeof(float), cudaMemcpyHostToDevice);
cudaMemcpy(d_B, h_B, size *size * sizeof(float), cudaMemcpyHostToDevice);
for (int i = 0; i < num; i++){
cublasSdot(handle, 1024, d_A , 1, d_B , 1, &res);
}
cudaEventRecord(stop2, NULL);
cudaEventSynchronize(stop2);
float msecTotal2 = 0.0f;
cudaEventElapsedTime(&msecTotal2, start2, stop2);
std::cout << "total time for dotVec:" << msecTotal2 << "\n";
cublasDestroy(handle);
cudaFree(d_A);
cudaFree(d_B);
cudaFree(d_C);
delete[] h_A;
delete[] h_B;
delete[] h_C;
return 1;
}
Обновление: я пытался также выполнить точечный продукт с cublasSgemm рассматривая вектор как 1 by 1024 матрица. Результат 3550 msec, что лучше, но все же в 7 раз больше, чем при первом расчете.
Решение
Одна проблема заключается в том, что вы неправильно обрабатываете режим указателя для вызова cublasSdot,
Вы хотите прочитать эта секция руководства.
Кроме того это:
cublasSdot(handle, 1024, d_A , 1, d_B , 1, &res);
^^^^
незаконно при любых обстоятельствах. В CUDA не разрешено брать адрес переменной устройства в коде хоста. Вы, конечно, можете сделать это, но результаты являются мусором.
Когда я изменяю ваш код следующим образом:
cublasSetPointerMode(handle, CUBLAS_POINTER_MODE_DEVICE);
float *dres;
cudaMalloc(&dres, sizeof(float));
cudaEventRecord(start2, NULL);
cudaMemcpy(d_A, h_A, size *size * sizeof(float), cudaMemcpyHostToDevice);
cudaMemcpy(d_B, h_B, size *size * sizeof(float), cudaMemcpyHostToDevice);
for (int i = 0; i < num; i++){
if(cublasSdot(handle, 1024, d_A , 1, d_B , 1, dres) != CUBLAS_STATUS_SUCCESS) {std::cout << ".";}
}
Я получаю примерно 2: 1 соотношение времени выполнения для cublasSdot в cublasSgemm что может быть правдоподобным, особенно для этих размеров. Под капотом точечная операция подразумевает параллельное сокращение. 1024 потока могут вычислить частичные результаты, но тогда требуется параллельное сокращение в 1024 потока. Gemm не нуждается в параллельном сокращении, и поэтому может быть быстрее. Для получения 1024 результатов в каждом потоке можно назначить 1024 потока. Для алгоритма, связанного с памятью, разница между 32 ^ 2 и 32 ^ 3 операциями может быть не такой значительной, но параллельное сокращение подразумевает значительные дополнительные операции. Когда я тогда изменю size в вашей программе от 32 до 128 я вижу обратное соотношение, и умножение матрицы действительно становится в 3 раза длиннее скалярного произведения.
Другие решения