Fix cub::Debug for builds without RDC

[gevtushenko]

cub::Debug is used in most device algorithms and uses cudaGetLastError to reset errors occurred occurred during device-scope algorithms invocations. cudaGetLastError usage was introduced in 22b0573. The issue consists in cudaGetLastError usage in __host__ __device__ functions in build with -rdc=false:

__global__ void kernel() {
  cudaGetLastError();
}

int main() {
  kernel<<<1, 1>>>();
}

:nvcc main.cu
ptxas fatal   : Unresolved extern function 'cudaGetLastError'

The issue was caused by 3c20a56, so we have to backport this into 1.17. The fix consists in invoking cudaGetLastError only in the case of host invocations or device ones with -rdc=true.