CUDA памяти неприятности

Question

У меня есть CUDA ядро, которое я компиляции в cubin файл без каких-либо специальных флагов:

nvcc text.cu -cubin 

собирает, хотя с этим сообщением:

Advisory: Не могу сказать, на какой указатель указывает, если предполагается глобальное пространство памяти

и ссылка на строку в некотором временном файле cpp. Я могу заставить это работать, комментируя какой-то, казалось бы, произвольный код, который не имеет для меня никакого смысла.

Ядро выглядит следующим образом:

__global__ void string_search(char** texts, int* lengths, char* symbol, int* matches, int symbolLength) 
{ 
    int localMatches = 0; 
    int blockId = blockIdx.x + blockIdx.y * gridDim.x; 
    int threadId = threadIdx.x + threadIdx.y * blockDim.x; 
    int blockThreads = blockDim.x * blockDim.y; 

    __shared__ int localMatchCounts[32]; 

    bool breaking = false; 
    for(int i = 0; i < (lengths[blockId] - (symbolLength - 1)); i += blockThreads) 
    { 
     if(texts[blockId][i] == symbol[0]) 
     { 
      for(int j = 1; j < symbolLength; j++) 
      { 
       if(texts[blockId][i + j] != symbol[j]) 
       { 
        breaking = true; 
        break; 
       } 
      } 
      if (breaking) continue; 
      localMatches++; 
     } 
    } 

    localMatchCounts[threadId] = localMatches; 

    __syncthreads(); 

    if(threadId == 0) 
    { 
     int sum = 0; 
     for(int i = 0; i < 32; i++) 
     { 
      sum += localMatchCounts[i]; 
     } 
     matches[blockId] = sum; 
    } 
} 

Если я заменить строку

localMatchCounts[threadId] = localMatches; 

после первого цикла с этой линией

localMatchCounts[threadId] = 5; 

компилируется без каких-либо уведомлений. Это также может быть достигнуто путем комментирования кажущихся случайными частей цикла над линией. Я также попытался заменить локальный массив памяти нормальным массивом на отсутствие эффекта. Может ли кто-нибудь сказать мне, в чем проблема?

Система Vista 64bit, для чего ее стоит.

Редактировать: я исправил код, чтобы он действительно работал, хотя он все еще вызывает уведомление компилятора. Не похоже, что предупреждение является проблемой, по крайней мере, в отношении правильности (это может повлиять на производительность).

Answer 1

Массивы указателей типа char ** проблематичны в ядрах, поскольку ядра не имеют доступа к памяти хоста.
Лучше выделить один непрерывный буфер и разделить его таким образом, чтобы обеспечить параллельный доступ.
В этом случае я бы определить массив 1D, который содержит все строки, расположенные друг за другие и другой массив 1D, размером 2 * numberOfStrings, который содержит смещение каждой строки в пределах первого массива и его длиной:

Для пример - подготовка к ядру:

 
char* buffer = st[0] + st[1] + st[2] + ....; 
int* metadata = new int[numberOfStrings * 2]; 
int lastpos = 0; 
for (int cnt = 0; cnt < 2* numberOfStrings; cnt+=2) 
{ 
    metadata[cnt] = lastpos; 
    lastpos += length(st[cnt]); 
    metadata[cnt] = length(st[cnt]); 
} 

В ядре:

 
currentIndex = threadId + blockId * numberOfBlocks; 
char* currentString = buffer + metadata[2 * currentIndex]; 
int currentStringLength = metadata[2 * currentIndex + 1]; 

Answer 2

Проблема, похоже, связана с параметром char **. Включение этого в char * решило предупреждение, поэтому я подозреваю, что cuda может иметь проблемы с этой формой данных. Возможно, cuda предпочитает, что в этом случае используются конкретные 2D массивы cuda.