diff --git a/runtime/devices/cuda.hpp b/runtime/devices/cuda.hpp
new file mode 100644
index 0000000000000000000000000000000000000000..a9ae18d4568dbade7b4677d444fb9e4922dac4aa
--- /dev/null
+++ b/runtime/devices/cuda.hpp
@@ -0,0 +1,40 @@
+#include <cuda_runtime.h>
+#include <iostream>
+//---
+#include "pairs.hpp"
+
+#pragma once
+
+#define CUDA_ASSERT(a) { pairs::cuda_assert((a), __LINE__, __FILE__); }
+
+namespace pairs {
+
+inline void cuda_assert(cudaError_t err, const char *file, int line) {
+    if(cuda != cudaSuccess) {
+        std::cerr << file << ":" << line << ": " << cudaGetErrorString(err) << std::endl;
+        exit(-1);
+    }
+}
+
+__host__ __device__ void *device_alloc(size_t size) {
+    void *ptr;
+    CUDA_ASSERT(cudaMalloc(&ptr, size));
+    return ptr;
+}
+
+__host__ __device__ void *device_realloc(void *ptr, size_t size) {
+    void *new_ptr;
+    CUDA_ASSERT(cudaFree(ptr));
+    CUDA_ASSERT(cudaMalloc(&new_ptr, size));
+    return new_ptr;
+}
+
+__host__ void copy_to_device(void *h_ptr, const void *d_ptr, size_t count) {
+    CUDA_ASSERT(cudaMemcpy(d_ptr, h_ptr, count, cudaMemcpyHostToDevice));
+}
+
+__host__ void copy_to_host(void *d_ptr, const void *h_ptr, size_t count) {
+    CUDA_ASSERT(cudaMemcpy(h_ptr, d_ptr, count, cudaMemcpyDeviceToHost));
+}
+
+}
diff --git a/runtime/devices/dummy.hpp b/runtime/devices/dummy.hpp
new file mode 100644
index 0000000000000000000000000000000000000000..470b5432a77ba7f4e663151e0601b4e4a244d08c
--- /dev/null
+++ b/runtime/devices/dummy.hpp
@@ -0,0 +1,10 @@
+#pragma once
+
+namespace pairs {
+
+void *device_alloc(size_t size) { return NULL; }
+void *device_realloc(void *ptr, size_t size) { return NULL; }
+void copy_to_device(void *h_ptr, const void *d_ptr, size_t count) {}
+void copy_to_host(void *d_ptr, const void *h_ptr, size_t count) {}
+
+}
diff --git a/src/pairs/code_gen/cgen.py b/src/pairs/code_gen/cgen.py
index f6827b830af5011be5c4e88566a2bd756c3b0b35..f11a8ceb5fb662cf9a607061da1bdaae5fcfce16 100644
--- a/src/pairs/code_gen/cgen.py
+++ b/src/pairs/code_gen/cgen.py
@@ -48,6 +48,12 @@ class CGen:
         self.print("#include \"runtime/pairs.hpp\"")
         self.print("#include \"runtime/read_from_file.hpp\"")
         self.print("#include \"runtime/vtk.hpp\"")
+
+        if self.target.is_gpu():
+            self.print("#include \"runtime/devices/cuda.hpp\"")
+        else:
+            self.print("#include \"runtime/devices/dummy.hpp\"")
+
         self.print("")
         self.print("using namespace pairs;")
         self.print("")
@@ -229,11 +235,13 @@ class CGen:
 
         if isinstance(ast_node, CopyToDevice):
             array_name = ast_node.prop.name()
-            self.print(f"pairs::copy_to_device({array_name}, d_{array_name})")
+            size = self.generate_expression(ast_node.size)
+            self.print(f"pairs::copy_to_device({array_name}, d_{array_name}, {size});")
 
         if isinstance(ast_node, CopyToHost):
             array_name = ast_node.prop.name()
-            self.print(f"pairs::copy_to_host(d_{array_name}, {array_name})")
+            size = self.generate_expression(ast_node.size)
+            self.print(f"pairs::copy_to_host(d_{array_name}, {array_name}, {size});")
 
         if isinstance(ast_node, For):
             iterator = self.generate_expression(ast_node.iterator)
diff --git a/src/pairs/ir/device.py b/src/pairs/ir/device.py
index b583bcd9891504e81a286ccb5ba8b0c2ce9d2507..dfc4851541623b3657f75b43d5d065fd9bdaa15d 100644
--- a/src/pairs/ir/device.py
+++ b/src/pairs/ir/device.py
@@ -1,10 +1,17 @@
+from functools import reduce
+import operator
 from pairs.ir.ast_node import ASTNode
+from pairs.ir.bin_op import BinOp
+from pairs.ir.sizeof import Sizeof
 
 
 class CopyToDevice(ASTNode):
     def __init__(self, sim, prop):
         super().__init__(sim)
+        sizes = prop.sizes()
         self.prop = prop
+        self.prim_size = Sizeof(sim, prop.type())
+        self.size = BinOp.inline(self.prim_size * (reduce(operator.mul, sizes) if isinstance(sizes, list) else sizes))
 
     def children(self):
         return [self.prop]
@@ -13,7 +20,10 @@ class CopyToDevice(ASTNode):
 class CopyToHost(ASTNode):
     def __init__(self, sim, prop):
         super().__init__(sim)
+        sizes = prop.sizes()
         self.prop = prop
+        self.prim_size = Sizeof(sim, prop.type())
+        self.size = BinOp.inline(self.prim_size * (reduce(operator.mul, sizes) if isinstance(sizes, list) else sizes))
 
     def children(self):
         return [self.prop]