include/Geant4/LUPI_declareMacro.hpp

0001 /*
0002 # <<BEGIN-copyright>>
0003 # Copyright 2019, Lawrence Livermore National Security, LLC.
0004 # This file is part of the gidiplus package (https://github.com/LLNL/gidiplus).
0005 # gidiplus is licensed under the MIT license (see https://opensource.org/licenses/MIT).
0006 # SPDX-License-Identifier: MIT
0007 # <<END-copyright>>
0008 */
0009
0010 #ifndef LUPI_declare_macro_hpp_included
0011 #define LUPI_declare_macro_hpp_included
0012
0013 #include <LUPI_defines.hpp>
0014
0015 // Default, if LUPI_HIP_INLINE is not defined, is to use an attribute function
0016 // to inform HIP not to inline the function.
0017 // This is quite useful for the publicly installed header files for code robustness
0018 // However, when compiling the source, one should be able to disable
0019 // this within the library itself for faster code.  To  do so
0020 // the define -DLUPI_HIP_INLINE can be added to the compiler flags, and
0021 // the define will evaluate to nothing, so the compiler is welcome to do
0022 // its optimizations.
0023
0024 #ifdef LUPI_HIP_INLINE
0025     #define LUPI_HIP_INLINE_ATTRIBUTE
0026 #else
0027     #define LUPI_HIP_INLINE_ATTRIBUTE  __attribute__ ((noinline))
0028 #endif
0029
0030 #define gpuErrorCheck(ans) { gpuAssert((ans), __FILE__, __LINE__); }
0031
0032 #if defined(__HIP_DEVICE_COMPILE__) || defined(__CUDA_ARCH__)
0033     #define LUPI_ON_GPU 1
0034 #endif
0035
0036 #ifdef __CUDACC__
0037     #include <cstdio>
0038     #define LUPI_HOST __host__
0039     #define LUPI_DEVICE __device__
0040     #define LUPI_HOST_DEVICE __host__ __device__
0041     #define LUPI_THROW(arg) printf("%s", arg)
0042     #define LUPI_WARP_SIZE 32
0043     #define LUPI_THREADID threadIdx.x
0044 inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
0045 {
0046    if (code != cudaSuccess)
0047    {
0048       fprintf(stderr,"GPUASSERT: %s File: %s line: %d\n", cudaGetErrorString(code), file, line);
0049       if (abort) exit(code);
0050    }
0051 }
0052
0053 #elif HAVE_OPENMP_TARGET
0054     #define LUPI_HOST
0055     #define LUPI_DEVICE
0056     #define LUPI_HOST_DEVICE
0057     #define LUPI_THROW(arg) printf("%s", arg)
0058     #define LUPI_WARP_SIZE 1
0059     #define LUPI_THREADID
0060 inline void gpuAssert(int code, const char *file, int line, bool abort=true) {}
0061 #elif defined(__HIP__)
0062     #include <hip/hip_version.h>
0063     #include <hip/hip_runtime.h>
0064     #include <hip/hip_runtime_api.h>
0065     #include <hip/hip_common.h>
0066
0067     #define LUPI_HOST __host__
0068     #define LUPI_DEVICE __device__
0069     #define LUPI_HOST_DEVICE LUPI_HIP_INLINE_ATTRIBUTE __host__ __device__
0070     #define LUPI_THROW(arg)
0071     #define LUPI_WARP_SIZE 1
0072     #define LUPI_THREADID hipThreadIdx_x
0073 inline void gpuAssert(hipError_t code, const char *file, int line, bool do_abort=true)
0074 {
0075     if (code == hipSuccess) { return; }
0076     printf("GPUassert code %d: %s %s %d\n", code, hipGetErrorString(code), file, line);
0077     if (do_abort) { abort(); }
0078 }
0079
0080 #else
0081     #define LUPI_HOST
0082     #define LUPI_DEVICE
0083     #define LUPI_HOST_DEVICE
0084     #define LUPI_THROW(arg) throw arg
0085     #define LUPI_WARP_SIZE 1
0086     #define LUPI_THREADID
0087 inline void gpuAssert(LUPI_maybeUnused int code, LUPI_maybeUnused const char *file, LUPI_maybeUnused int line, LUPI_maybeUnused bool abort=true) {}
0088 #endif
0089
0090 #endif      // End of LUPI_declare_macro_hpp_included