Hendrik/xgboost
1
0
Fork 0
Code Issues Pull Requests Packages Projects Releases Wiki Activity

Mimic cuda assert output in span check. (#4762)

Browse Source
This commit is contained in:
Jiaming Yuan 2019-08-13 01:44:54 -04:00 committed by GitHub
parent c5b229632d
commit c0ffe65f5c
No known key found for this signature in database
GPG Key ID: 4AEE18F83AFDEB23
1 changed files with 12 additions and 9 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

21
src/common/span.h
View File

@ -70,15 +70,18 @@ namespace common {
// Usual logging facility is not available inside device code. // Usual logging facility is not available inside device code.
// TODO(trivialfis): Make dmlc check more generic. // TODO(trivialfis): Make dmlc check more generic.
// assert is not supported in mac as of CUDA 10.0 // assert is not supported in mac as of CUDA 10.0
#define KERNEL_CHECK(cond) \ #define KERNEL_CHECK(cond) \
do { \ do { \
if (!(cond)) { \ if (!(cond)) { \
printf("\nKernel error:\n" \ printf("\nKernel error:\n" \
"In: %s, \tline: %d\n" \ "In: %s: %d\n" \
"\t%s\n\tExpecting: %s\n", \ "\t%s\n\tExpecting: %s\n" \
__FILE__, __LINE__, __PRETTY_FUNCTION__, #cond); \ "\tBlock: [%d, %d, %d], Thread: [%d, %d, %d]\n\n", \
asm("trap;"); \ __FILE__, __LINE__, __PRETTY_FUNCTION__, #cond, \
} \ blockIdx.x, blockIdx.y, blockIdx.z, \
threadIdx.x, threadIdx.y, threadIdx.z); \
asm("trap;"); \
} \
} while (0); } while (0);
#ifdef __CUDA_ARCH__ #ifdef __CUDA_ARCH__