Obtaining Runtime Error Codes

All Runtime APIs return an error code. However, for an asynchronous API, the API returns a value before the device task is complete. Therefore, the API cannot report asynchronous task errors that may occur on the device. The API can only return errors that occur on the host before the device task is executed, for example, parameter verification failure. If an asynchronous task error occurs, the corresponding error code will be returned when an unrelated Runtime API is called later.

Therefore, the only way to check for asynchronous errors immediately after an asynchronous function call is to call the aclrtSynchronizeDevice API (or use any other synchronization mechanism described in Explicit Synchronization) immediately after the function call and check the error code returned by the aclrtSynchronizeDevice API.

The Runtime module maintains an error variable for each host thread. This variable is initialized to ACL_RT_SUCCESS and is overwritten by the error code each time an error (whether it is a parameter verification error or an asynchronous error) occurs. The aclrtPeekAtLastError API returns the value of this variable. The aclrtGetLastError API also returns this variable, but it also resets it to ACL_RT_SUCCESS.

 1
 2
 3
 4
 5
 6
 7
 8
 9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31

// Specify the device. All Runtime APIs return their error codes, which can be used to check whether the execution is successful.
aclError error = aclrtSetDevice(0);

// Create a stream.
aclrtStream stream;
error = aclrtCreateStream(&stream);

// Set the ACL_STOP_ON_FAILURE mode.
aclrtSetStreamFailureMode(stream, ACL_STOP_ON_FAILURE);

// Deliver the task in the stream. The return code only indicates whether the task is successfully delivered. Generally, it indicates the parameter verification failure on the host and cannot indicate the actual execution error on the device.
error = aclrtMemcpyAsync(devPtr, devSize, hostPtr, hostSize, ACL_MEMCPY_HOST_TO_DEVICE, stream);
error = myKernel<<<8, nullptr, stream>>>();
error = aclrtMemcpyAsync(hostPtr, hostSize, devPtr, devSize, ACL_MEMCPY_DEVICE_TO_HOST, stream);

// Block application running until the compute device completes computation, and obtains the error code of the current asynchronous task.
error = aclrtSynchronizeDevice();

// Obtain the latest error of the current thread.
error = aclrtPeekAtLastError(ACL_RT_THREAD_LEVEL);

// Obtain the latest error of the current thread and reset the status to ACL_RT_SUCCESS.
error = aclrtGetLastError(ACL_RT_THREAD_LEVEL);

// Obtain ErrorMsg and output it to logs.
char *errMsg = aclGetRecentErrMsg();
......

// Destroy resources.
error = aclrtDestroyStream(stream);
error = aclrtResetDevice(0);

Note: In continuing upon failure mode (for details, see Task Stopping Upon Failure), if an exception occurs during the execution of a task in a stream, other tasks that have not been executed in the stream can still be executed, and new tasks can still be delivered to the stream and other streams in the same context. In this case, error information returned by aclrtPeekAtLastError and aclrtGetLastError may not be about the first error.