Gabriele Cosmo
91 lines
3.0 KiB
C++
91 lines
3.0 KiB
C++
/*
|
|
# <<BEGIN-copyright>>
|
|
# Copyright 2019, Lawrence Livermore National Security, LLC.
|
|
# This file is part of the gidiplus package (https://github.com/LLNL/gidiplus).
|
|
# gidiplus is licensed under the MIT license (see https://opensource.org/licenses/MIT).
|
|
# SPDX-License-Identifier: MIT
|
|
# <<END-copyright>>
|
|
*/
|
|
|
|
#ifndef LUPI_declare_macro_hpp_included
|
|
#define LUPI_declare_macro_hpp_included
|
|
|
|
#include <LUPI_defines.hpp>
|
|
|
|
// Default, if LUPI_HIP_INLINE is not defined, is to use an attribute function
|
|
// to inform HIP not to inline the function.
|
|
// This is quite useful for the publicly installed header files for code robustness
|
|
// However, when compiling the source, one should be able to disable
|
|
// this within the library itself for faster code. To do so
|
|
// the define -DLUPI_HIP_INLINE can be added to the compiler flags, and
|
|
// the define will evaluate to nothing, so the compiler is welcome to do
|
|
// its optimizations.
|
|
|
|
#ifdef LUPI_HIP_INLINE
|
|
#define LUPI_HIP_INLINE_ATTRIBUTE
|
|
#else
|
|
#define LUPI_HIP_INLINE_ATTRIBUTE __attribute__ ((noinline))
|
|
#endif
|
|
|
|
#define gpuErrorCheck(ans) { gpuAssert((ans), __FILE__, __LINE__); }
|
|
|
|
#if defined(__HIP_DEVICE_COMPILE__) || defined(__CUDA_ARCH__)
|
|
#define LUPI_ON_GPU 1
|
|
#endif
|
|
|
|
#ifdef __CUDACC__
|
|
#include <cstdio>
|
|
#define LUPI_HOST __host__
|
|
#define LUPI_DEVICE __device__
|
|
#define LUPI_HOST_DEVICE __host__ __device__
|
|
#define LUPI_THROW(arg) printf("%s", arg)
|
|
#define LUPI_WARP_SIZE 32
|
|
#define LUPI_THREADID threadIdx.x
|
|
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
|
|
{
|
|
if (code != cudaSuccess)
|
|
{
|
|
fprintf(stderr,"GPUASSERT: %s File: %s line: %d\n", cudaGetErrorString(code), file, line);
|
|
if (abort) exit(code);
|
|
}
|
|
}
|
|
|
|
#elif HAVE_OPENMP_TARGET
|
|
#define LUPI_HOST
|
|
#define LUPI_DEVICE
|
|
#define LUPI_HOST_DEVICE
|
|
#define LUPI_THROW(arg) printf("%s", arg)
|
|
#define LUPI_WARP_SIZE 1
|
|
#define LUPI_THREADID
|
|
inline void gpuAssert(int code, const char *file, int line, bool abort=true) {}
|
|
#elif defined(__HIP__)
|
|
#include <hip/hip_version.h>
|
|
#include <hip/hip_runtime.h>
|
|
#include <hip/hip_runtime_api.h>
|
|
#include <hip/hip_common.h>
|
|
|
|
#define LUPI_HOST __host__
|
|
#define LUPI_DEVICE __device__
|
|
#define LUPI_HOST_DEVICE LUPI_HIP_INLINE_ATTRIBUTE __host__ __device__
|
|
#define LUPI_THROW(arg)
|
|
#define LUPI_WARP_SIZE 1
|
|
#define LUPI_THREADID hipThreadIdx_x
|
|
inline void gpuAssert(hipError_t code, const char *file, int line, bool do_abort=true)
|
|
{
|
|
if (code == hipSuccess) { return; }
|
|
printf("GPUassert code %d: %s %s %d\n", code, hipGetErrorString(code), file, line);
|
|
if (do_abort) { abort(); }
|
|
}
|
|
|
|
#else
|
|
#define LUPI_HOST
|
|
#define LUPI_DEVICE
|
|
#define LUPI_HOST_DEVICE
|
|
#define LUPI_THROW(arg) throw arg
|
|
#define LUPI_WARP_SIZE 1
|
|
#define LUPI_THREADID
|
|
inline void gpuAssert(LUPI_maybeUnused int code, LUPI_maybeUnused const char *file, LUPI_maybeUnused int line, LUPI_maybeUnused bool abort=true) {}
|
|
#endif
|
|
|
|
#endif // End of LUPI_declare_macro_hpp_included
|