fix(skiplist): Fix some device function not declear correctly

Signed-off-by: KunoiSayami <[email protected]>

1 files changed, 12 insertions, 3 deletions

diff --git a/util/arena.cuh b/util/arena.cuh
index 618b426..afbc575 100644
--- a/util/arena.cuh
+++ b/util/arena.cuh
@@ -25,7 +25,7 @@ class Arena {
   ~Arena();
 
   // Return a pointer to a newly allocated memory block of "bytes" bytes.
-  char* Allocate(size_t bytes);
+  __device__ char* Allocate(size_t bytes);
 
   // Allocate memory with the normal alignment guarantees provided by malloc.
   __device__ char* AllocateAligned(size_t bytes);
@@ -40,13 +40,22 @@ class Arena {
   __device__ char* AllocateFallback(size_t bytes);
   __device__ char* AllocateNewBlock(size_t block_bytes);
 
+  struct ArenaNode {
+    char * block;
+    ArenaNode * next;
+  };
+
   // Allocation state
   char* alloc_ptr_;
   size_t alloc_bytes_remaining_;
 
   // Array of new[] allocated memory blocks
   //thrust::host_vector<char *> blocks_;
-  std::vector<char*> blocks_;
+  //std::vector<char*> blocks_;
+
+  ArenaNode * head_;
+  ArenaNode * blocks_;
+
 
   // Total memory usage of the arena.
   //
@@ -55,7 +64,7 @@ class Arena {
   cuda::atomic<size_t> memory_usage_;
 };
 
-inline char* Arena::Allocate(size_t bytes) {
+__device__ inline char* Arena::Allocate(size_t bytes) {
   // The semantics of what to return are a bit messy if we allow
   // 0-byte allocations, so we disallow them here (we don't need
   // them for our internal use).