CFD/hypre
1
0
Fork 0
Code Issues Pull Requests Packages Projects Releases Wiki Activity

fix domp build

Browse Source
This commit is contained in:
Ruipeng Li 2022-03-28 22:55:18 -07:00
parent ce30e2faa4
commit e5393bc0eb
3 changed files with 17 additions and 11 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

15
src/utilities/_hypre_utilities.hpp
View File

@ -229,10 +229,9 @@ using dim3 = sycl::range<1>;
{ \
if ( gridsize[0] == 0 || blocksize[0] == 0 ) \
{ \
hypre_printf("Error %s %d: Invalid SYCL 1D launch parameters grid/block (%d) (%d)\n", \
/* hypre_printf("Warning %s %d: Zero SYCL 1D launch parameters grid/block (%d) (%d)\n", \
__FILE__, __LINE__, \
gridsize[0], blocksize[0]); \
assert(0); exit(1); \
gridsize[0], blocksize[0]); */ \
} \
else \
{ \
@ -565,24 +564,24 @@ struct hypre_GpuMatData
#endif //#if defined(HYPRE_USING_GPU)
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* WM: generic device functions (cuda/hip/sycl)
* WM: Q - what about device openmp?
* generic device functions (cuda/hip/sycl)
* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - */
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP) || defined(HYPRE_USING_SYCL)
template <typename T>
static __device__ __forceinline__
T read_only_load( const T *ptr )
{
/* WM: Q - where/when is __CUDA_ARCH__ defined? */
#if defined(__CUDA_ARCH__) && __CUDA_ARCH__ >= 350
return __ldg( ptr );
#else
return *ptr;
#endif
}
#endif
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* WM: cuda/hip functions
* cuda/hip functions
* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - */
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP)
@ -1157,6 +1156,7 @@ struct TupleComp3
* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - */
/* device_utils.c */
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP) || defined(HYPRE_USING_SYCL)
dim3 hypre_GetDefaultDeviceBlockDimension();
dim3 hypre_GetDefaultDeviceGridDimension( HYPRE_Int n, const char *granularity, dim3 bDim );
@ -1170,6 +1170,7 @@ hypreDevice_StableSortTupleByTupleKey(HYPRE_Int N, T1 *keys1, T2 *keys2, T3 *val
template <typename T1, typename T2, typename T3> HYPRE_Int hypreDevice_ReduceByTupleKey(HYPRE_Int N,
T1 *keys1_in, T2 *keys2_in, T3 *vals_in, T1 *keys1_out, T2 *keys2_out, T3 *vals_out);
#endif
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP)

3
src/utilities/device_utils.c
View File

@ -14,6 +14,8 @@
* generic device functions (cuda/hip/sycl)
* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - */
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP) || defined(HYPRE_USING_SYCL)
dim3
hypre_GetDefaultDeviceBlockDimension()
{
@ -602,6 +604,7 @@ hypreDevice_ReduceByTupleKey(HYPRE_Int N, T1 *keys1_in, T2 *keys2_in, T3 *vals
template HYPRE_Int hypreDevice_ReduceByTupleKey(HYPRE_Int N, HYPRE_Int *keys1_in,
HYPRE_Int *keys2_in, HYPRE_Complex *vals_in, HYPRE_Int *keys1_out, HYPRE_Int *keys2_out,
HYPRE_Complex *vals_out);
#endif
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* cuda/hip functions

10
src/utilities/device_utils.h
View File

@ -507,24 +507,24 @@ struct hypre_GpuMatData
#endif //#if defined(HYPRE_USING_GPU)
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* WM: generic device functions (cuda/hip/sycl)
* WM: Q - what about device openmp?
* generic device functions (cuda/hip/sycl)
* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - */
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP) || defined(HYPRE_USING_SYCL)
template <typename T>
static __device__ __forceinline__
T read_only_load( const T *ptr )
{
/* WM: Q - where/when is __CUDA_ARCH__ defined? */
#if defined(__CUDA_ARCH__) && __CUDA_ARCH__ >= 350
return __ldg( ptr );
#else
return *ptr;
#endif
}
#endif
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* WM: cuda/hip functions
* cuda/hip functions
* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - */
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP)
@ -1099,6 +1099,7 @@ struct TupleComp3
* - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - - */
/* device_utils.c */
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP) || defined(HYPRE_USING_SYCL)
dim3 hypre_GetDefaultDeviceBlockDimension();
dim3 hypre_GetDefaultDeviceGridDimension( HYPRE_Int n, const char *granularity, dim3 bDim );
@ -1112,6 +1113,7 @@ hypreDevice_StableSortTupleByTupleKey(HYPRE_Int N, T1 *keys1, T2 *keys2, T3 *val
template <typename T1, typename T2, typename T3> HYPRE_Int hypreDevice_ReduceByTupleKey(HYPRE_Int N,
T1 *keys1_in, T2 *keys2_in, T3 *vals_in, T1 *keys1_out, T2 *keys2_out, T3 *vals_out);
#endif
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP)