CUDA Speicher Probleme

Question

Ich habe einen CUDA-Kernel, die ich ohne spezielle Flaggen auf eine Cubin Datei bin Kompilieren:

nvcc text.cu -cubin

Es kompiliert, obwohl mit dieser Nachricht angezeigt:

  

Advisory: Kann nicht sagen, was Zeiger zeigt auf, globalen Speicherplatz vorausgesetzt

und ein Verweis auf eine Zeile in einer temporären CPP-Datei. Ich kann dies durch Kommentare für einige scheinbar willkürlichen Codes zu arbeiten, die mir keinen Sinn macht.

Der Kernel ist wie folgt:

__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;
    }
}

Wenn ich ersetzen Sie die Zeile

localMatchCounts[threadId] = localMatches;

, nachdem die erste for-Schleife mit dieser Zeile

localMatchCounts[threadId] = 5;

es kompiliert ohne bemerkt. Dies kann auch durch Kommentare aus scheinbar zufälligen Teilen der Schleife über der Linie erreicht werden. Ich habe auch versucht, mit einem normalen Array zu keinem Effekt auf dem lokalen Speicher-Array zu ersetzen. Kann mir jemand sagen, was das Problem ist?

Das System ist Vista 64bit, für das, was ihren Wert.

Edit: Ich habe den Code festgelegt, so dass es tatsächlich funktioniert, obwohl es immer noch die Compiler Mitteilung erzeugt. Es scheint nicht, als ob die Warnung ein Problem ist, zumindest in Bezug auf Korrektheit (es kann die Leistung beeinträchtigt).

Answer 1

Arrays von Zeigern wie char ** sind problematisch in Körnern, da die Kerne keinen Zugriff auf den Host-Speicher.
Es ist besser, ein einziges zuzuteilen kontinuierliche Puffer und es in einer Art und Weise zu unterteilen, die den parallelen Zugriff ermöglicht.
In diesem Fall würde ich einen 1D-Array definieren, die alle Zeichenfolge enthält, eine nach der anderen und einem anderen 1D-Array angeordnet ist, Größe 2 * numberOfStrings die die jede Saite Offset enthalten in der ersten Anordnung und ihre Gesamtlänge:

Zum Beispiel - Vorbereitung für den Kernel:

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]);
}

Im Kern:

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

Answer 2

Das Problem scheint mit dem char ** Parameter zugeordnet werden. Drehen dieses in ein char * löste die Warnung, so vermute ich, dass CUDA Probleme mit dieser Form von Daten haben könnten. Vielleicht cuda bevorzugt, dass man den spezifischen cuda 2D-Arrays in diesem Fall verwendet wird.