Единая память CUDA и Windows 10 - PullRequest
Новая
       22

Единая память CUDA и Windows 10

2 голосов
/ 17 февраля 2020

При использовании CudaMallocManaged () для выделения массива структур с массивами внутри я получаю сообщение об ошибке «недостаточно памяти», хотя у меня достаточно свободной памяти. Вот некоторый код, который повторяет мою проблему: 

#include <iostream>
#include <cuda.h>

#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
{
   if (code != cudaSuccess) 
   {
      fprintf(stderr,"GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
      if (abort) exit(code);
   }
}

#define N 100000
#define ARR_SZ 100

struct Struct
{
    float* arr;
};

int main()
{
    Struct* struct_arr;

    gpuErrchk( cudaMallocManaged((void**)&struct_arr, sizeof(Struct)*N) );
    for(int i = 0; i < N; ++i)
        gpuErrchk( cudaMallocManaged((void**)&(struct_arr[i].arr), sizeof(float)*ARR_SZ) ); //out of memory...

    for(int i = 0; i < N; ++i)
        cudaFree(struct_arr[i].arr);
    cudaFree(struct_arr);

    /*float* f;
    gpuErrchk( cudaMallocManaged((void**)&f, sizeof(float)*N*ARR_SZ) ); //this works ok
    cudaFree(f);*/

    return 0;
}

Кажется, что нет проблемы, когда я вызываю cudaMallocManaged () один раз, чтобы выделить один кусок памяти, как я показываю в последней части комментируемого кода. У меня GeForce GTX 1070 Ti, и я использую Windows 10. Друг попытался скомпилировать тот же код в P C с Linux, и он работал правильно, хотя у него была та же проблема в другом P C с Windows 10. WDDM TDR отключен. Любая помощь будет оценена. Спасибо.

1 Ответ

3 голосов
/ 17 февраля 2020

Существует гранулярность выделения.

Это означает, что если вы запрашиваете 1 байт или 400 байт, то, что фактически израсходовано, будет что-то вроде 4096 65536 байт. Таким образом, группа очень небольших выделений будет фактически использовать память с гораздо большей скоростью, чем вы могли бы прогнозировать на основе запрошенного размера выделения. Решение состоит не в том, чтобы делать очень маленькие выделения, а в том, чтобы распределять их большими кусками.

Альтернативная стратегия здесь также будет заключаться в том, чтобы сгладить ваше выделение и выделить из него части для каждого из ваших массивов: 

#include <iostream>
#include <cstdio>

#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
{
   if (code != cudaSuccess)
   {
      fprintf(stderr,"GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
      if (abort) exit(code);
   }
}

#define N 100000
#define ARR_SZ 100

struct Struct
{
    float* arr;
};

int main()
{
    Struct* struct_arr;
    float* f;

    gpuErrchk( cudaMallocManaged((void**)&struct_arr, sizeof(Struct)*N) );
    gpuErrchk( cudaMallocManaged((void**)&f, sizeof(float)*N*ARR_SZ) );
    for(int i = 0; i < N; ++i)
        struct_arr[i].arr = f+i*ARR_SZ;
    cudaFree(struct_arr);
    cudaFree(f);

    return 0;
}

ARR_SZ делится на 4 означает, что различные созданные указатели также могут быть преобразованы в более крупные векторные типы, например float2 или float4, если ваше использование имело намерение сделать это.

Возможная причина, по которой оригинальный код работает на linux, заключается в том, что управляемая память на linux при правильной настройке может переподписывать физическую память графического процессора. В результате фактический предел выделения намного выше, чем то, что может предложить встроенная память графического процессора. Возможно также, что в случае linux имеется немного больше свободной памяти, или, возможно, гранулярность распределения в linux отличается (меньше).

Исходя из вопроса в комментариях, я решил оценить гранулярность выделения, используя этот код: 

#include <iostream>
#include <cstdio>

#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
inline void gpuAssert(cudaError_t code, const char* file, int line, bool abort = true)
{
    if (code != cudaSuccess)
    {
        fprintf(stderr, "GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
        if (abort) exit(code);
    }
}

#define N 100000
#define ARR_SZ 100

struct Struct
{
    float* arr;
};

int main()
{
    Struct* struct_arr;
    //float* f;

    gpuErrchk(cudaMallocManaged((void**)& struct_arr, sizeof(Struct) * N));
#if 0
    gpuErrchk(cudaMallocManaged((void**)& f, sizeof(float) * N * ARR_SZ));
    for (int i = 0; i < N; ++i)
        struct_arr[i].arr = f + i * ARR_SZ;
#else
    size_t fre, tot;
    gpuErrchk(cudaMemGetInfo(&fre, &tot));
    std::cout << "Free: " << fre << " total: " << tot << std::endl;
    for (int i = 0; i < N; ++i)
        gpuErrchk(cudaMallocManaged((void**) & (struct_arr[i].arr), sizeof(float) * ARR_SZ)); 
    gpuErrchk(cudaMemGetInfo(&fre, &tot));
    std::cout << "Free: " << fre << " total: " << tot << std::endl;

    for (int i = 0; i < N; ++i)
        cudaFree(struct_arr[i].arr);
#endif
    cudaFree(struct_arr);
    //cudaFree(f);

    return 0;
}

Когда я компилирую проект отладки с этим кодом и запускаю его на настольном компьютере windows 10 с графическим процессором RTX 2070 (8 ГБ памяти, так же, как GTX 1070 Ti) ) Я получаю следующий вывод: 

Microsoft Windows [Version 10.0.17763.973]
(c) 2018 Microsoft Corporation. All rights reserved.

C:\Users\Robert Crovella>cd C:\Users\Robert Crovella\source\repos\test12\x64\Debug

C:\Users\Robert Crovella\source\repos\test12\x64\Debug>test12
Free: 7069866393 total: 8589934592
Free: 516266393 total: 8589934592

C:\Users\Robert Crovella\source\repos\test12\x64\Debug>test12
Free: 7069866393 total: 8589934592
Free: 516266393 total: 8589934592

C:\Users\Robert Crovella\source\repos\test12\x64\Debug>

  1. Обратите внимание, что на моем компьютере остается только 0,5 ГБ зарегистрированной свободной памяти после 100 000 выделений. Поэтому, если по какой-либо причине ваш 8 ГБ графический процессор запускается с меньшим количеством свободной памяти (вполне возможно), вы можете столкнуться с ошибкой нехватки памяти, даже если я этого не сделал.

  2. Расчет детализации распределения выглядит следующим образом: 

    7069866393 - 516266393 / 100000 = 65536 bytes per allocation(!)

    Таким образом, моя предыдущая оценка в 4096 байт на распределение была далека, по крайней мере, на 1 порядок, на моей машине / тестовой установке.

  3. Степень детализации распределения может варьироваться в зависимости от:

    • windows или linux
    • WDDM или T CC
    • x86 или Power9
    • управляемый против обычного cudaMalloc
    • возможно, другие факторы (например, версия CUDA)

    , поэтому мой совет будущим читателям не будет предполагать, что это всегда 65536 байтов на выделение, минимум.

...