portelli/Grid
1
0
Fork 0
mirror of https://github.com/paboyle/Grid.git synced 2025-04-09 21:50:45 +01:00
Code Releases Activity

NVCC compile fixes

Browse Source
This commit is contained in:
Peter Boyle 2020-05-08 13:14:12 -04:00
parent f8b8e00090
commit 52081acfa5
2 changed files with 14 additions and 9 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

11
Grid/Grid_Eigen_Dense.h
View File

@ -12,25 +12,26 @@
#endif #endif
/* NVCC save and restore compile environment*/ /* NVCC save and restore compile environment*/
#ifdef __NVCC__ #ifdef GRID_CUDA
#pragma push #pragma push
#pragma diag_suppress code_is_unreachable #pragma diag_suppress code_is_unreachable
#pragma push_macro("GRID_SIMT") #pragma push_macro("__CUDA_ARCH__")
#pragma push_macro("__NVCC__") #pragma push_macro("__NVCC__")
#pragma push_macro("__CUDACC__") #pragma push_macro("__CUDACC__")
#undef __CUDA_ARCH__
#undef __NVCC__ #undef __NVCC__
#undef __CUDACC__ #undef __CUDACC__
#undef GRID_SIMT
#define __NVCC__REDEFINE__ #define __NVCC__REDEFINE__
#endif #endif
/* SYCL save and restore compile environment*/ /* SYCL save and restore compile environment*/
#ifdef __SYCL_DEVICE_ONLY__ #ifdef GRID_SYCL
#pragma push #pragma push
#pragma push_macro("__SYCL_DEVICE_ONLY__") #pragma push_macro("__SYCL_DEVICE_ONLY__")
#undef __SYCL_DEVICE_ONLY__ #undef __SYCL_DEVICE_ONLY__
#undef EIGEN_USE_SYCL
#define EIGEN_DONT_VECTORIZE #define EIGEN_DONT_VECTORIZE
//#undef EIGEN_USE_SYCL
#define __SYCL__REDEFINE__
#endif #endif

12
Grid/threads/Accelerator.h
View File

@ -73,6 +73,10 @@ void acceleratorThreads(uint32_t);
////////////////////////////////////////////// //////////////////////////////////////////////
// CUDA acceleration // CUDA acceleration
////////////////////////////////////////////// //////////////////////////////////////////////
#ifdef __NVCC__
#define GRID_CUDA
#endif
#ifdef GRID_CUDA #ifdef GRID_CUDA
#ifdef __CUDA_ARCH__ #ifdef __CUDA_ARCH__
@ -114,7 +118,7 @@ inline void *acceleratorAllocShared(size_t bytes)
void *ptr=NULL; void *ptr=NULL;
auto err = cudaMallocManaged((void **)&ptr,bytes); auto err = cudaMallocManaged((void **)&ptr,bytes);
if( err != cudaSuccess ) { if( err != cudaSuccess ) {
ptr = (_Tp *) NULL; ptr = (void *) NULL;
printf(" cudaMallocManaged failed for %d %s \n",bytes,cudaGetErrorString(err)); printf(" cudaMallocManaged failed for %d %s \n",bytes,cudaGetErrorString(err));
} }
return ptr; return ptr;
@ -124,7 +128,7 @@ inline void *acceleratorAllocDevice(size_t bytes)
void *ptr=NULL; void *ptr=NULL;
auto err = cudaMalloc((void **)&ptr,bytes); auto err = cudaMalloc((void **)&ptr,bytes);
if( err != cudaSuccess ) { if( err != cudaSuccess ) {
ptr = (_Tp *) NULL; ptr = (void *) NULL;
printf(" cudaMalloc failed for %d %s \n",bytes,cudaGetErrorString(err)); printf(" cudaMalloc failed for %d %s \n",bytes,cudaGetErrorString(err));
} }
return ptr; return ptr;
@ -232,7 +236,7 @@ inline void *acceleratorAllocShared(size_t bytes)
void *ptr=NULL; void *ptr=NULL;
auto err = hipMallocManaged((void **)&ptr,bytes); auto err = hipMallocManaged((void **)&ptr,bytes);
if( err != hipSuccess ) { if( err != hipSuccess ) {
ptr = (_Tp *) NULL; ptr = (void *) NULL;
printf(" hipMallocManaged failed for %d %s \n",bytes,hipGetErrorString(err)); printf(" hipMallocManaged failed for %d %s \n",bytes,hipGetErrorString(err));
} }
return ptr; return ptr;
@ -242,7 +246,7 @@ inline void *acceleratorAllocDevice(size_t bytes)
void *ptr=NULL; void *ptr=NULL;
auto err = hipMalloc((void **)&ptr,bytes); auto err = hipMalloc((void **)&ptr,bytes);
if( err != hipSuccess ) { if( err != hipSuccess ) {
ptr = (_Tp *) NULL; ptr = (void *) NULL;
printf(" hipMalloc failed for %d %s \n",bytes,hipGetErrorString(err)); printf(" hipMalloc failed for %d %s \n",bytes,hipGetErrorString(err));
} }
return ptr; return ptr;