|
|
|
@ -158,7 +158,13 @@ HOSTDEVICE int64_t& indexer<0>(Dim<0>& dim, int idx) {
|
|
|
|
|
#else
|
|
|
|
|
PADDLE_ASSERT(false);
|
|
|
|
|
#endif
|
|
|
|
|
#if (defined __CUDA_ARCH__) && (CUDA_VERSION < 8000)
|
|
|
|
|
// On CUDA versions previous to 8.0, only __shared__ variables
|
|
|
|
|
// could be declared as static in the device code.
|
|
|
|
|
int64_t head = 0;
|
|
|
|
|
#else
|
|
|
|
|
static int64_t head = 0;
|
|
|
|
|
#endif
|
|
|
|
|
return head;
|
|
|
|
|
}
|
|
|
|
|
|
|
|
|
@ -184,7 +190,13 @@ HOSTDEVICE int64_t indexer<0>(const Dim<0>& dim, int idx) {
|
|
|
|
|
#else
|
|
|
|
|
PADDLE_ASSERT(false);
|
|
|
|
|
#endif
|
|
|
|
|
#if (defined __CUDA_ARCH__) && (CUDA_VERSION < 8000)
|
|
|
|
|
// On CUDA versions previous to 8.0, only __shared__ variables
|
|
|
|
|
// could be declared as static in the device code.
|
|
|
|
|
int64_t head = 0;
|
|
|
|
|
#else
|
|
|
|
|
static int64_t head = 0;
|
|
|
|
|
#endif
|
|
|
|
|
return head;
|
|
|
|
|
}
|
|
|
|
|
|
|
|
|
|