Warum funktioniert mein CUDA kernel-crash (unspecified launch failure) mit einem anderen dataset-Größe?

Habe ich einen kernel zur Berechnung verschiedener Elemente einer matrix, basierend auf Ihrer position (diagonal-oder off-diagonal). Den kernel wie erwartet funktioniert, wenn die Berechnung der Matrizen der Größen:

  • 14 x 14 (ich verstehe, das ist klein und nicht sachgerechte Nutzung der GPU-Ressourcen, aber dies war rein für Testzwecke, um sicherzustellen, die Ergebnisse waren korrekt)
  • 118 x 118, und
  • 300 x 300

Allerdings, wenn ich versuche zu berechnen, eine matrix der Größe 2383 x 2383, wird der kernel Abstürzen. Insbesondere die Fehlermeldung "Unspecified launch failure" ausgegeben, auf die cudaMemcpy () - Zeile, um Ergebnisse vom Gerät zum host. Aus der Forschung, verstehe ich, dass dieser Fehler entsteht in der Regel im Falle eines out-of-bounds memory access (z.B. in einem array), was ich aber nicht verstehe ist, dass es funktioniert für die drei vorangegangenen Fällen, nicht aber für den 2383 x 2383 Fall. Der kernel code ist unten dargestellt:

__global__ void createYBus(float *R, float *X, float *B, int numberOfBuses, int numberOfBranches, int *fromBus, int *toBus, cuComplex *y)
{
    int rowIdx = blockIdx.y*blockDim.y + threadIdx.y;
    int colIdx = blockIdx.x*blockDim.x + threadIdx.x;
    int index = rowIdx*numberOfBuses + colIdx;
    if (rowIdx<numberOfBuses && colIdx<numberOfBuses)
    {
        for (int i=0; i<numberOfBranches; ++i)
        {
            if (rowIdx==fromBus[i] && colIdx==fromBus[i]) { //diagonal element
                y[index] = cuCaddf(y[index], make_cuComplex((R[i]/((R[i]*R[i])+(X[i]*X[i]))), (-(X[i]/((R[i]*R[i])+(X[i]*X[i])))+ (B[i]/2))));
            }
            if (rowIdx==toBus[i] && colIdx==toBus[i]) { //diagonal element
                y[index] = cuCaddf(y[index], make_cuComplex((R[i]/((R[i]*R[i])+(X[i]*X[i]))), (-(X[i]/((R[i]*R[i])+(X[i]*X[i])))+ (B[i]/2))));
            }
            if (rowIdx==fromBus[i] && colIdx==toBus[i]) { //off-diagonal element
                y[index] = make_cuComplex(-(R[i]/((R[i]*R[i])+(X[i]*X[i]))), X[i]/((R[i]*R[i])+(X[i]*X[i])));
            }
            if (rowIdx==toBus[i] && colIdx==fromBus[i]) { //off-diagonal element
                y[index] = make_cuComplex(-(R[i]/((R[i]*R[i])+(X[i]*X[i]))), X[i]/((R[i]*R[i])+(X[i]*X[i])));
            }
        }
    }
}

Globalen Speicher Zuordnungen erfolgen über Aufrufe von cudaMalloc(). Die Zuweisungen des Kodex lauten wie folgt:

cudaStat1 = cudaMalloc((void**)&dev_fromBus, numLines*sizeof(int));
cudaStat2 = cudaMalloc((void**)&dev_toBus, numLines*sizeof(int));
cudaStat3 = cudaMalloc((void**)&dev_R, numLines*sizeof(float));
cudaStat4 = cudaMalloc((void**)&dev_X, numLines*sizeof(float));
cudaStat5 = cudaMalloc((void**)&dev_B, numLines*sizeof(float));
cudaStat6 = cudaMalloc((void**)&dev_y, numberOfBuses*numberOfBuses*sizeof(cuComplex));
cudaStat7 = cudaMalloc((void**)&dev_Pd, numberOfBuses*sizeof(float));
cudaStat8 = cudaMalloc((void**)&dev_Qd, numberOfBuses*sizeof(float));
cudaStat9 = cudaMalloc((void**)&dev_Vmag, numberOfBuses*sizeof(float));
cudaStat10 = cudaMalloc((void**)&dev_theta, numberOfBuses*sizeof(float));
cudaStat11 = cudaMalloc((void**)&dev_Peq, numberOfBuses*sizeof(float));
cudaStat12 = cudaMalloc((void**)&dev_Qeq, numberOfBuses*sizeof(float));
cudaStat13 = cudaMalloc((void**)&dev_Peq1, numberOfBuses*sizeof(float));
cudaStat14 = cudaMalloc((void**)&dev_Qeq1, numberOfBuses*sizeof(float));
...
...
cudaStat15 = cudaMalloc((void**)&dev_powerMismatch, jacSize*sizeof(float));
cudaStat16 = cudaMalloc((void**)&dev_jacobian, jacSize*jacSize*sizeof(float));
cudaStat17 = cudaMalloc((void**)&dev_stateVector, jacSize*sizeof(float));
cudaStat18 = cudaMalloc((void**)&dev_PQindex, jacSize*sizeof(int));

wo cudaStatN Typ cudaError_t, um Fehler abzufangen. Die letzten vier Verteilungen wurden später im code auf und sind für einen anderen kernel. Allerdings sind diese Zuordnungen wurden getan, bevor der kernel in der Frage genannt wurde.

Den Start-Parameter sind wie folgt:

dim3 dimBlock(16, 16); //number of threads 
dim3 dimGrid((numberOfBuses+15)/16, (numberOfBuses+15)/16);  //number of blocks

//launch kernel once data has been copied to GPU
createYBus<<<dimGrid, dimBlock>>>(dev_R, dev_X, dev_B, numberOfBuses, numLines, dev_fromBus, dev_toBus, dev_y);

//copy results back to CPU
cudaStat6 = cudaMemcpy(y_bus, dev_y, numberOfBuses*numberOfBuses*sizeof(cuComplex), cudaMemcpyDeviceToHost);
if (cudaStat6 != cudaSuccess) {
    cout<<"Device memcpy failed"<<endl;
    cout<<cudaGetErrorString(cudaStat6)<<endl;
    return 1;
}

Entfernte ich die timing-code nur um zu zeigen, block-und grid-Maßen und-Fehlerüberprüfung-Technik verwendet.

Habe ich auch einen host (C++ code) version dieser Funktion, und ich bin die Weitergabe der Daten für beide Funktionen und Vergleich-Ergebnisse, Erstens, um sicherzustellen, der kernel erzeugt die richtigen Ergebnisse, und zweitens im Hinblick auf die Ausführung der Zeit, Leistungen zu vergleichen. Ich doppelt überprüft die Daten für den 2383 x 2383 Fall (es ist gelesen, die aus einer text-Datei und kopiert global memory) und ich bin nicht der Suche nach irgendwelchen Anomalien in der array-Zugriffe/Indizierung.

Ich bin mit Visual Studio 2010, also versuchte ich mit Nsight den Fehler zu finden (ich bin nicht allzu versiert mit Nsight). Die übersicht Bericht-übersicht heißt es: "Es wurde 1 runtime-API-Aufruf-Fehler gemeldet. (Bitte sehen Sie die CUDA-Runtime-API-Aufrufe Berichts für weitere Informationen). In der Liste der Laufzeit-API-Aufrufe, cudaMemcpy gibt Fehler 4 - nicht sicher, ob die Thread-ID (5012) von Bedeutung ist in der Tabelle - diese Zahl schwankt mit jedem Lauf. CUDA-memcheck Werkzeug (in der Befehlszeile) das folgende liefert:

Thank you for using this program
========= Program hit cudaErrorLaunchFailure (error 4) due to "unspecified launch failure" on CUDA API call to cudaMemcpy.
=========     Saved host backtrace up to driver entry point at error
=========
========= ERROR SUMMARY: 1 error

Ich weiß, dass mein kernel ist nicht das effizienteste, da es viele global memory-Zugriffe. Warum wird der kernel-Absturz für dieses größere matrix? Gibt es eine out-of-bounds array access, die mir fehlt? Jegliche Hilfe würde sehr geschätzt werden.

  • Es könnte sein, dass der memory-allocation-Größe hat einige Nichtübereinstimmung mit den Parametern. Könnte Sie auch Aktien Zuteilung Anrufe ?
  • Sind Sie sicher, dass Ihr den Wert "int index", in den kernel, richtig ? Sollte es nicht davon abhängen, die Blöcke und Dimensionen statt ?
  • Ich bearbeitet die post zu zählen Zuweisung fordert. Ich verstehe, dass der Globale Speicher ist groß, aber ist es möglich, dass mir Zuweisung von zu viel Speicher? Ich bin nicht immer alle Fehler für cudaMalloc() Aufrufe obwohl.
  • Ich bin die Einleitung einer 2D-raster mit 2D-Blöcke. Der Globale thread-ID wird bei der Berechnung der x-und y-Richtung, dann index kombiniert die global thread ID Formeln, index die volle matrix in abgeflachter form.
InformationsquelleAutor danieljovan | 2016-04-28
Schreibe einen Kommentar