Jaki jest kanoniczny sposób sprawdzania błędów za pomocą CUDA runtime API?
Przeglądając odpowiedzi i komentarze na pytania CUDA oraz w CUDA tag wiki , widzę, że często sugeruje się, że status powrotu każdego wywołania API powinien być sprawdzany pod kątem błędów. Dokumentacja API zawiera takie funkcje jak cudaGetLastError
, cudaPeekAtLastError
, i cudaGetErrorString
, ale jaki jest najlepszy sposób, aby je połączyć, aby niezawodnie łapać i zgłaszać błędy bez konieczności stosowania dużej ilości dodatkowego kodu?
4 answers
Prawdopodobnie najlepszym sposobem na sprawdzenie błędów w kodzie runtime API jest zdefiniowanie funkcji obsługi stylu assert i makra wrappera w następujący sposób:
#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);
}
}
Można następnie zawinąć każde wywołanie API za pomocą makra gpuErrchk
, które przetworzy status zwracanego wywołania API, na przykład:
gpuErrchk( cudaMalloc(&a_d, size*sizeof(int)) );
Jeśli wystąpi błąd w wywołaniu, wiadomość tekstowa opisująca błąd oraz plik i linię w kodzie, w którym wystąpił błąd, zostanie wysłana do stderr
i aplikacja zakończy działanie. Możliwe, że można zmodyfikować gpuAssert
, aby wywołać wyjątek, a nie wywołać exit()
w bardziej zaawansowanej aplikacji, jeśli jest to wymagane.
Drugie związane z tym pytanie brzmi: jak sprawdzić błędy podczas uruchamiania jądra, które nie mogą być bezpośrednio zawinięte w wywołanie makra, takie jak standardowe wywołania runtime API. Dla jąder, coś takiego:
kernel<<<1,1>>>(a);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );
Najpierw sprawdzi, czy nie ma nieprawidłowego argumentu uruchamiania, a następnie zmusi hosta, aby poczekał, aż jądro się zatrzyma i sprawdzi, czy nie wystąpił błąd wykonania. Na synchronizacja może zostać wyeliminowana, jeśli masz kolejne wywołanie API blokujące, takie jak:
kernel<<<1,1>>>(a_d);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaMemcpy(a_h, a_d, size * sizeof(int), cudaMemcpyDeviceToHost) );
W takim przypadku wywołanie cudaMemcpy
może zwracać błędy, które wystąpiły podczas wykonywania jądra lub te z samej kopii pamięci. Może to być mylące dla początkujących i zalecałbym użycie jawnej synchronizacji po uruchomieniu jądra podczas debugowania, aby ułatwić zrozumienie, gdzie mogą pojawić się problemy.
Warning: date(): Invalid date.timezone value 'Europe/Kyiv', we selected the timezone 'UTC' for now. in /var/www/agent_stack/data/www/doraprojects.net/template/agent.layouts/content.php on line 54
2014-09-22 05:17:58
Powyższa odpowiedź Talonmies to świetny sposób na anulowanie aplikacji w stylu assert
.
Czasami możemy chcieć zgłosić i odzyskać stan błędu w kontekście C++ jako część większej aplikacji.
Oto dość zwięzły sposób, aby to zrobić, rzucając wyjątek C++ pochodzący z std::runtime_error
za pomocą thrust::system_error
:
#include <thrust/system_error.h>
#include <thrust/system/cuda/error.h>
#include <sstream>
void throw_on_cuda_error(cudaError_t code, const char *file, int line)
{
if(code != cudaSuccess)
{
std::stringstream ss;
ss << file << "(" << line << ")";
std::string file_and_line;
ss >> file_and_line;
throw thrust::system_error(code, thrust::cuda_category(), file_and_line);
}
}
Spowoduje to włączenie nazwy pliku, numeru linii i opisu cudaError_t
w języku angielskim do wyrzuconego wyjątku .what()
członek:
#include <iostream>
int main()
{
try
{
// do something crazy
throw_on_cuda_error(cudaSetDevice(-1), __FILE__, __LINE__);
}
catch(thrust::system_error &e)
{
std::cerr << "CUDA error after cudaSetDevice: " << e.what() << std::endl;
// oops, recover
cudaSetDevice(0);
}
return 0;
}
Wyjście:
$ nvcc exception.cu -run
CUDA error after cudaSetDevice: exception.cu(23): invalid device ordinal
Klient some_function
może odróżnić błędy CUDA od innych rodzajów błędów w razie potrzeby:
try
{
// call some_function which may throw something
some_function();
}
catch(thrust::system_error &e)
{
std::cerr << "CUDA error during some_function: " << e.what() << std::endl;
}
catch(std::bad_alloc &e)
{
std::cerr << "Bad memory allocation during some_function: " << e.what() << std::endl;
}
catch(std::runtime_error &e)
{
std::cerr << "Runtime error during some_function: " << e.what() << std::endl;
}
catch(...)
{
std::cerr << "Some other kind of error during some_function" << std::endl;
// no idea what to do, so just rethrow the exception
throw;
}
Ponieważ thrust::system_error
jest std::runtime_error
, możemy alternatywnie obsługiwać go w taki sam sposób jak szeroką klasę błędów, jeśli nie wymagamy precyzji poprzedniego przykładu:
try
{
// call some_function which may throw something
some_function();
}
catch(std::runtime_error &e)
{
std::cerr << "Runtime error during some_function: " << e.what() << std::endl;
}
Warning: date(): Invalid date.timezone value 'Europe/Kyiv', we selected the timezone 'UTC' for now. in /var/www/agent_stack/data/www/doraprojects.net/template/agent.layouts/content.php on line 54
2015-05-18 20:55:44
Sposób C++-canonical: nie sprawdzaj błędów...użyj wiązań C++, które rzucają wyjątki.
Kiedyś byłem irytowany tym problemem; i miałem rozwiązanie funkcji makro-cum-wrapper, tak jak w szponach i odpowiedziach Jareda,ale szczerze? To sprawia, że korzystanie z CUDA Runtime API jeszcze bardziej brzydkie i C-jak.
Więc podszedłem do tego w inny i bardziej fundamentalny sposób. Dla próbki wyniku, oto część próbki CUDAvectorAdd
- z kompletny błąd sprawdzanie każdego wywołania runtime API:
// (... prepare host-side buffers here ...)
auto current_device = cuda::device::current::get();
auto d_A = cuda::memory::device::make_unique<float[]>(current_device, numElements);
auto d_B = cuda::memory::device::make_unique<float[]>(current_device, numElements);
auto d_C = cuda::memory::device::make_unique<float[]>(current_device, numElements);
cuda::memory::copy(d_A.get(), h_A.get(), size);
cuda::memory::copy(d_B.get(), h_B.get(), size);
// (... prepare a launch configuration here... )
cuda::launch( vectorAdd, launch_config,
d_A.get(), d_B.get(), d_C.get(), numElements
);
cuda::memory::copy(h_C.get(), d_C.get(), size);
// (... verify results here...)
Ponownie - wszystkie potencjalne błędy są sprawdzane i zgłaszane za pomocą wyrzuconego wyjątku. Ten kod używa mojego
Thin Modern-wrappery C++ dla biblioteki CUDA Runtime API (Github)
Zauważ, że wyjątki zawierają zarówno wyjaśnienie ciągu znaków, jak i kod statusu CUDA runtime API po nieudanym wywołaniu.
Kilka linków do tego, jak błędy CUDA są automatycznie sprawdzane z tymi opakowaniami:
Warning: date(): Invalid date.timezone value 'Europe/Kyiv', we selected the timezone 'UTC' for now. in /var/www/agent_stack/data/www/doraprojects.net/template/agent.layouts/content.php on line 54
2018-09-24 08:33:24
Rozwiązanie omówione Tutaj zadziałało dla mnie dobrze. To rozwiązanie wykorzystuje wbudowane funkcje cuda i jest bardzo proste do wdrożenia.
Odpowiedni kod jest skopiowany poniżej:
#include <stdio.h>
#include <stdlib.h>
__global__ void foo(int *ptr)
{
*ptr = 7;
}
int main(void)
{
foo<<<1,1>>>(0);
// make the host block until the device is finished with foo
cudaDeviceSynchronize();
// check for error
cudaError_t error = cudaGetLastError();
if(error != cudaSuccess)
{
// print the CUDA error message and exit
printf("CUDA error: %s\n", cudaGetErrorString(error));
exit(-1);
}
return 0;
}
Warning: date(): Invalid date.timezone value 'Europe/Kyiv', we selected the timezone 'UTC' for now. in /var/www/agent_stack/data/www/doraprojects.net/template/agent.layouts/content.php on line 54
2016-03-15 19:35:45