OpenFPM_pdata  4.1.0
Project that contain the implementation of distributed structures
util_debug.cuh
Go to the documentation of this file.
1 /******************************************************************************
2  * Copyright (c) 2011, Duane Merrill. All rights reserved.
3  * Copyright (c) 2011-2018, NVIDIA CORPORATION. All rights reserved.
4  *
5  * Redistribution and use in source and binary forms, with or without
6  * modification, are permitted provided that the following conditions are met:
7  * * Redistributions of source code must retain the above copyright
8  * notice, this list of conditions and the following disclaimer.
9  * * Redistributions in binary form must reproduce the above copyright
10  * notice, this list of conditions and the following disclaimer in the
11  * documentation and/or other materials provided with the distribution.
12  * * Neither the name of the NVIDIA CORPORATION nor the
13  * names of its contributors may be used to endorse or promote products
14  * derived from this software without specific prior written permission.
15  *
16  * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" AND
17  * ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE IMPLIED
18  * WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE ARE
19  * DISCLAIMED. IN NO EVENT SHALL NVIDIA CORPORATION BE LIABLE FOR ANY
20  * DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES
21  * (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES;
22  * LOSS OF USE, DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND
23  * ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
24  * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
25  * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
26  *
27  ******************************************************************************/
28 
37 #pragma once
38 
39 #include <stdio.h>
40 #include "util_namespace.cuh"
41 #include "util_arch.cuh"
42 
44 CUB_NS_PREFIX
45 
47 namespace cub {
48 
49 
56 #if (defined(DEBUG) || defined(_DEBUG)) && !defined(CUB_STDERR)
58  #define CUB_STDERR
59 #endif
60 
61 
62 
68 __host__ __device__ __forceinline__ cudaError_t Debug(
69  cudaError_t error,
70  const char* filename,
71  int line)
72 {
73  (void)filename;
74  (void)line;
75 #ifdef CUB_STDERR
76  if (error)
77  {
78  #if (CUB_PTX_ARCH == 0)
79  fprintf(stderr, "CUDA error %d [%s, %d]: %s\n", error, filename, line, cudaGetErrorString(error));
80  fflush(stderr);
81  #elif (CUB_PTX_ARCH >= 200)
82  printf("CUDA error %d [block (%d,%d,%d) thread (%d,%d,%d), %s, %d]\n", error, blockIdx.z, blockIdx.y, blockIdx.x, threadIdx.z, threadIdx.y, threadIdx.x, filename, line);
83  #endif
84  }
85 #endif
86  return error;
87 }
88 
89 
93 #ifndef CubDebug
94  #define CubDebug(e) cub::Debug((cudaError_t) (e), __FILE__, __LINE__)
95 #endif
96 
97 
101 #ifndef CubDebugExit
102  #define CubDebugExit(e) if (cub::Debug((cudaError_t) (e), __FILE__, __LINE__)) { exit(1); }
103 #endif
104 
105 
109 #if !defined(_CubLog)
110  #if !(defined(__clang__) && defined(__CUDA__))
111  #if (CUB_PTX_ARCH == 0)
112  #define _CubLog(format, ...) printf(format,__VA_ARGS__);
113  #elif (CUB_PTX_ARCH >= 200)
114  #define _CubLog(format, ...) printf("[block (%d,%d,%d), thread (%d,%d,%d)]: " format, blockIdx.z, blockIdx.y, blockIdx.x, threadIdx.z, threadIdx.y, threadIdx.x, __VA_ARGS__);
115  #endif
116  #else
117  // XXX shameless hack for clang around variadic printf...
118  // Compilies w/o supplying -std=c++11 but shows warning,
119  // so we sielence them :)
120  #pragma clang diagnostic ignored "-Wc++11-extensions"
121  #pragma clang diagnostic ignored "-Wunnamed-type-template-args"
122  template <class... Args>
123  inline __host__ __device__ void va_printf(char const* format, Args const&... args)
124  {
125  #ifdef __CUDA_ARCH__
126  printf(format, blockIdx.z, blockIdx.y, blockIdx.x, threadIdx.z, threadIdx.y, threadIdx.x, args...);
127  #else
128  printf(format, args...);
129  #endif
130  }
131  #ifndef __CUDA_ARCH__
132  #define _CubLog(format, ...) va_printf(format,__VA_ARGS__);
133  #else
134  #define _CubLog(format, ...) va_printf("[block (%d,%d,%d), thread (%d,%d,%d)]: " format, __VA_ARGS__);
135  #endif
136  #endif
137 #endif
138 
139 
140 
141  // end group UtilMgmt
143 
144 } // CUB namespace
145 CUB_NS_POSTFIX // Optional outer namespace(s)
Optional outer namespace(s)
__host__ __device__ __forceinline__ cudaError_t Debug(cudaError_t error, const char *filename, int line)
CUB error reporting macro (prints error messages to stderr)
Definition: util_debug.cuh:68