Move hip functions to cpp file

src/targets/miopen/CMakeLists.txt

@@ -7,6 +7,7 @@ if(NOT TARGET MIOpen)
 endif()
 
 add_library(rtg_miopen
+    hip.cpp
     miopen_target.cpp
 )
 rocm_clang_tidy_check(rtg_miopen)

src/targets/miopen/hip.cpp

+
+#include <rtg/miopen/hip.hpp>
+
+namespace rtg {
+namespace miopen {
+
+hip_ptr gpu_allocate(std::size_t sz)
+{
+    void* result;
+    // TODO: Check status
+    hipMalloc(&result, sz);
+    return hip_ptr{result};
+}
+
+hip_ptr write_to_gpu(const void* x, std::size_t sz)
+{
+    auto result = gpu_allocate(sz);
+    // TODO: Check status
+    hipMemcpy(result.get(), x, sz, hipMemcpyHostToDevice);
+    return result;
+}
+
+rtg::argument to_gpu(rtg::argument arg)
+{
+    auto p = share(write_to_gpu(arg.data(), arg.get_shape().bytes()));
+    return {arg.get_shape(), [p]() mutable { return reinterpret_cast<char*>(p.get()); }};
+}
+
+rtg::argument from_gpu(rtg::argument arg)
+{
+    rtg::argument result;
+    arg.visit([&](auto x) {
+        using type = typename decltype(x)::value_type;
+        auto v     = read_from_gpu<type>(arg.data(), x.get_shape().bytes() / sizeof(type));
+        result     = {x.get_shape(), [v]() mutable { return reinterpret_cast<char*>(v.data()); }};
+    });
+    return result;
+}
+
+} // namespace miopen
+
+} // namespace rtg

src/targets/miopen/include/rtg/miopen/hip.hpp

@@ -2,29 +2,24 @@
 #define RTG_GUARD_RTGLIB_HIP_HPP
 
 #include <rtg/manage_ptr.hpp>
-
+#include <rtg/argument.hpp>
+#include <rtg/operators.hpp>
 #include <miopen/miopen.h>
 
+#include <vector>
+
 namespace rtg {
 namespace miopen {
 
 using hip_ptr = RTG_MANAGE_PTR(void, hipFree);
 
-inline hip_ptr gpu_allocate(std::size_t sz)
-{
-    void* result;
-    // TODO: Check status
-    hipMalloc(&result, sz);
-    return hip_ptr{result};
-}
+hip_ptr gpu_allocate(std::size_t sz);
 
-inline hip_ptr write_to_gpu(const void* x, std::size_t sz)
-{
-    auto result = gpu_allocate(sz);
-    // TODO: Check status
-    hipMemcpy(result.get(), x, sz, hipMemcpyHostToDevice);
-    return result;
-}
+hip_ptr write_to_gpu(const void* x, std::size_t sz);
+
+rtg::argument to_gpu(rtg::argument arg);
+
+rtg::argument from_gpu(rtg::argument arg);
 
 template <class T>
 hip_ptr write_to_gpu(const T& x)
@@ -43,23 +38,6 @@ std::vector<T> read_from_gpu(const void* x, std::size_t sz)
     return result;
 }
 
-inline rtg::argument to_gpu(rtg::argument arg)
-{
-    auto p = share(write_to_gpu(arg.data(), arg.get_shape().bytes()));
-    return {arg.get_shape(), [p]() mutable { return reinterpret_cast<char*>(p.get()); }};
-}
-
-inline rtg::argument from_gpu(rtg::argument arg)
-{
-    rtg::argument result;
-    arg.visit([&](auto x) {
-        using type = typename decltype(x)::value_type;
-        auto v     = read_from_gpu<type>(arg.data(), x.get_shape().bytes() / sizeof(type));
-        result     = {x.get_shape(), [v]() mutable { return reinterpret_cast<char*>(v.data()); }};
-    });
-    return result;
-}
-
 struct hip_allocate
 {
     std::string name() const { return "hip::allocate"; }