start gpu

Src/AMReX_Arena.cpp

@@ -1,16 +1,70 @@
 #include "AMReX_Arena.H"
+#include "AMReX_Gpu.H"
 
 namespace amrex
 {
 
 void* allocate_host (std::size_t sz)
 {
+#if defined(AMREX_USE_CUDA)
+    void* p;
+    cudaHostAlloc(&p, sz, cudaHostAllocMapped);
+    return p;
+#elif defined(AMREX_USE_HIP)
+    void* p;
+    hipHostAlloc(&p, sz, hipHostAllocMapped | hipHostMallocNonCoherent);
+    return p;
+#elif defined(AMREX_USE_SYCL)
+    return sycl::malloc_host(...);
+#else
     return std::malloc(sz);
+#endif
 }
 
 void free_host (void* pt)
 {
+#if defined(AMREX_USE_CUDA)
+    cudaFreeHost(pt);
+#elif defined(AMREX_USE_HIP)
+    hipHostFree(pt);
+#elif defined(AMREX_USE_SYCL)
+    sycl::free(...);
+#else
     std::free(pt);
+#endif
 }
 
+#ifdef AMREX_USE_GPU
+
+void* allocate_device (std::size_t sz)
+{
+    void* p;
+#if defined(AMREX_USE_CUDA)
+    cudaMalloc(&p, sz);
+#elif defined(AMREX_USE_HIP)
+    hipMalloc(&p, sz);
+#elif defined(AMREX_USE_SYCL)
+    p = sycl::malloc_device(...);
+#else
+    static_assert(false);
+#endif
+    return p;
+}
+
+void free_device (void* pt)
+{
+#if defined(AMREX_USE_CUDA)
+    cudaFree(pt);
+#elif defined(AMREX_USE_HIP)
+    hipFree(pt);
+#elif defined(AMREX_USE_SYCL)
+    sycl::free(...);
+#else
+    static_assert(false);
+#endif
+
+}
+
+#endif
+
 }

Src/AMReX_BLassert.H

@@ -21,7 +21,7 @@ void Assert (const char* EX, const char* file, int line, const char* msg)
 {
     AMREX_IF_ON_HOST((Assert_host(EX,file,line,msg);))
     AMREX_IF_ON_DEVICE((amrex::ignore_unused(EX,file,line,msg);))
-    AMREX_IF_ON_DEVICE((assert(0)));
+    AMREX_IF_ON_DEVICE((assert(0);))
 }
 
 }

Src/AMReX_Gpu.H

@@ -1,5 +1,8 @@
 #ifndef AMREX_GPU_H_
 #define AMREX_GPU_H_
+
+#include <cstdlib>
+
 #if defined(AMREX_USE_GPU) && !defined(AMREX_USE_SYCL)
 
 #if defined(AMREX_USE_HIP)
@@ -70,12 +73,22 @@
                                               AMREX_WRONG_NUM_ARGS)(__VA_ARGS__)
 
 #ifdef AMREX_USE_GPU
-namespace amrex::Gpu {
+namespace amrex {
+
+#if defined(AMREX_USE_CUDA)
+using gpuStream_t = cudaStream_t;
+#elif defined(AMREX_USE_HIP)
+using gpuStream_t = hipStream_t;
+#elif defined(AMREX_USE_SYCL)
+using gpuStream_t = sycl::queue*;
+#endif
 
-// add init(); // to get stream I am supposed to use
+namespace Gpu {
 
-void htod_memcpy (....); // todo: need to synchronize too, use default stream
+    void setStream (gpuStream_t a_stream);
 
+    void htod_memcpy (void* p_d, void const* p_h, std::size_t sz);
+}
 }
 #endif
 

Src/AMReX_Gpu.cpp

@@ -0,0 +1,34 @@
+#include "AMReX_Gpu.H"
+
+#ifdef AMREX_USE_GPU
+
+namespace {
+#ifdef AMREX_USE_SYCL
+    amrex::gpuStream_t gpu_stream = nullptr;
+#else
+    amrex::gpuStream_t gpu_stream = 0;
+#endif
+}
+
+namespace amrex::Gpu {
+
+void setStream (gpuStream_t a_stream)
+{
+    gpu_stream = a_stream;
+}
+
+void htod_memcpy (void* p_d, void const* p_h, std::size_t sz)
+{
+#if defined(AMREX_USE_CUDA)
+    cudaMemcpyAsync(p_d, p_h, sz, cudaMemcpyHostToDevice, gpu_stream);
+    cudaStreamSynchronize(gpu_stream);
+#elif defined(AMREX_USE_HIP)
+#elif defined(AMREX_USE_SYCL)
+#else
+    static_assert(false);
+#endif
+}
+
+}
+
+#endif

Src/Make.package

@@ -18,6 +18,7 @@ CEXE_headers += AMReX_Arena.H \
 
 CEXE_sources += AMReX_Arena.cpp \
                 AMReX_BLassert.cpp \
+                AMReX_Gpu.cpp \
                 AMReX_Parser.cpp \
                 AMReX_Parser_Exe.cpp \
                 amrex_parser.lex.cpp \

Tests/GPU/GNUmakefile

@@ -0,0 +1,16 @@
+AMREX_HOME := ../..
+
+DEBUG	= FALSE
+
+DIM	= 3
+
+COMP    = gcc
+
+USE_CUDA  = TRUE
+USE_HIP   = FALSE
+USE_SYCL  = FALSE
+
+include $(AMREX_HOME)/Tools/GNUMake/Make.defs
+include ./Make.package
+include $(AMREX_HOME)/Src/Make.package
+include $(AMREX_HOME)/Tools/GNUMake/Make.rules

Tests/GPU/Make.package

@@ -0,0 +1 @@
+CEXE_sources += main.cpp

Tests/GPU/main.cpp

@@ -0,0 +1,10 @@
+#include "AMReX_Parser.H"
+#include <iostream>
+
+using namespace amrex;
+
+
+int main (int argc, char* argv[])
+{
+    amrex::ignore_unused(argc, argv);
+}

Tools/GNUMake/Make.defs

@@ -481,6 +481,9 @@ else ifeq ($(USE_CUDA),TRUE)
         SYSTEM_INCLUDE_LOCATIONS += $(MAKE_CUDA_PATH)/include
     endif
 
+    LINKFLAGS = $(NVCC_FLAGS) $(NVCC_ARCH_LINK_FLAGS) $(CXXFLAGS_FROM_HOST)
+    AMREX_LINKER = nvcc
+
     ifeq ($(nvcc_forward_unknowns),0)
 
       comm := ,