test(skiplist): Add insert and check test

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

2 files changed, 157 insertions, 86 deletions

diff --git a/db/skiplist.cuh b/db/skiplist.cuh
index 4fcd220..8fe864a 100644
--- a/db/skiplist.cuh
+++ b/db/skiplist.cuh
@@ -59,21 +59,21 @@ class SkipList {
   __device__ void Insert(const Key& key);
 
   // Returns true iff an entry that compares equal to key is in the list.
-  bool Contains(const Key& key) const;
+  __device__ bool Contains(const Key& key) const;
 
   // Iteration over the contents of a skip list
   class Iterator {
    public:
     // Initialize an iterator over the specified list.
     // The returned iterator is not valid.
-    explicit Iterator(const SkipList* list);
+    __device__ __host__ explicit Iterator(const SkipList* list);
 
     // Returns true iff the iterator is positioned at a valid node.
-    bool Valid() const;
+    __device__ __host__ bool Valid() const;
 
     // Returns the key at the current position.
     // REQUIRES: Valid()
-    const Key& key() const;
+    __device__ const Key& key() const;
 
     // Advances to the next position.
     // REQUIRES: Valid()
@@ -88,11 +88,11 @@ class SkipList {
 
     // Position at the first entry in list.
     // Final state of iterator is Valid() iff list is not empty.
-    void SeekToFirst();
+    __device__ void SeekToFirst();
 
     // Position at the last entry in list.
     // Final state of iterator is Valid() iff list is not empty.
-    void SeekToLast();
+    __device__ void SeekToLast();
 
    private:
     const SkipList* list_;
@@ -209,18 +209,18 @@ __device__ typename SkipList<Key, Comparator>::Node* SkipList<Key, Comparator>::
 }
 
 template <typename Key, class Comparator>
-inline SkipList<Key, Comparator>::Iterator::Iterator(const SkipList* list) {
+__device__ __host__ inline SkipList<Key, Comparator>::Iterator::Iterator(const SkipList* list) {
   list_ = list;
   node_ = nullptr;
 }
 
 template <typename Key, class Comparator>
-inline bool SkipList<Key, Comparator>::Iterator::Valid() const {
+__device__ __host__ inline bool SkipList<Key, Comparator>::Iterator::Valid() const {
   return node_ != nullptr;
 }
 
 template <typename Key, class Comparator>
-inline const Key& SkipList<Key, Comparator>::Iterator::key() const {
+__device__ inline const Key& SkipList<Key, Comparator>::Iterator::key() const {
   assert(Valid());
   return node_->key;
 }
@@ -248,12 +248,12 @@ __device__ inline void SkipList<Key, Comparator>::Iterator::Seek(const Key& targ
 }
 
 template <typename Key, class Comparator>
-inline void SkipList<Key, Comparator>::Iterator::SeekToFirst() {
+__device__ inline void SkipList<Key, Comparator>::Iterator::SeekToFirst() {
   node_ = list_->head_->Next(0);
 }
 
 template <typename Key, class Comparator>
-inline void SkipList<Key, Comparator>::Iterator::SeekToLast() {
+__device__ inline void SkipList<Key, Comparator>::Iterator::SeekToLast() {
   node_ = list_->FindLast();
   if (node_ == list_->head_) {
     node_ = nullptr;
@@ -390,7 +390,7 @@ __device__ void SkipList<Key, Comparator>::Insert(const Key& key) {
 }
 
 template <typename Key, class Comparator>
-bool SkipList<Key, Comparator>::Contains(const Key& key) const {
+__device__ bool SkipList<Key, Comparator>::Contains(const Key& key) const {
   Node* x = FindGreaterOrEqual(key, nullptr);
   if (x != nullptr && Equal(key, x->key)) {
     return true;
diff --git a/db/skiplist_test.cu b/db/skiplist_test.cu
index d18f11f..2b42b48 100644
--- a/db/skiplist_test.cu
+++ b/db/skiplist_test.cu
@@ -50,80 +50,6 @@ struct Comparator {
 }
 
 TEST(SkipTest, InsertAndLookup) {
-  const int N = 2000;
-  const int R = 5000;
-  Random rnd(1000);
-  std::set<Key> keys;
-  Arena arena;
-  Comparator cmp;
-  SkipList<Key, Comparator> list(cmp, &arena);
-  for (int i = 0; i < N; i++) {
-    Key key = rnd.Next() % R;
-    if (keys.insert(key).second) {
-      list.Insert(key);
-    }
-  }
-
-  for (int i = 0; i < R; i++) {
-    if (list.Contains(i)) {
-      ASSERT_EQ(keys.count(i), 1);
-    } else {
-      ASSERT_EQ(keys.count(i), 0);
-    }
-  }
-
-  // Simple iterator tests
-  {
-    SkipList<Key, Comparator>::Iterator iter(&list);
-    ASSERT_TRUE(!iter.Valid());
-
-    iter.Seek(0);
-    ASSERT_TRUE(iter.Valid());
-    ASSERT_EQ(*(keys.begin()), iter.key());
-
-    iter.SeekToFirst();
-    ASSERT_TRUE(iter.Valid());
-    ASSERT_EQ(*(keys.begin()), iter.key());
-
-    iter.SeekToLast();
-    ASSERT_TRUE(iter.Valid());
-    ASSERT_EQ(*(keys.rbegin()), iter.key());
-  }
-
-  // Forward iteration test
-  for (int i = 0; i < R; i++) {
-    SkipList<Key, Comparator>::Iterator iter(&list);
-    iter.Seek(i);
-
-    // Compare against model iterator
-    std::set<Key>::iterator model_iter = keys.lower_bound(i);
-    for (int j = 0; j < 3; j++) {
-      if (model_iter == keys.end()) {
-        ASSERT_TRUE(!iter.Valid());
-        break;
-      } else {
-        ASSERT_TRUE(iter.Valid());
-        ASSERT_EQ(*model_iter, iter.key());
-        ++model_iter;
-        iter.Next();
-      }
-    }
-  }
-
-  // Backward iteration test
-  {
-    SkipList<Key, Comparator>::Iterator iter(&list);
-    iter.SeekToLast();
-
-    // Compare against model iterator
-    for (std::set<Key>::reverse_iterator model_iter = keys.rbegin();
-         model_iter != keys.rend(); ++model_iter) {
-      ASSERT_TRUE(iter.Valid());
-      ASSERT_EQ(*model_iter, iter.key());
-      iter.Prev();
-    }
-    ASSERT_TRUE(!iter.Valid());
-  }
 }
 
 // We want to make sure that with a single writer and multiple
@@ -294,6 +220,148 @@ __global__ void insert_skiplist(SkipList<Key, Comparator> * skipList, Random *de
   }
 }
 
+struct Node {
+  Key num;
+  Node* next;
+
+};
+
+template<typename Key>
+class MemorySet {
+ public:
+  explicit __device__ MemorySet(): first(nullptr), current(nullptr) {};
+
+  __device__ ~MemorySet() {
+    Node * crt = first, * prev;
+    while (crt != nullptr) {
+      prev = crt;
+      crt = crt->next;
+      delete prev;
+    }
+    first = nullptr;
+  }
+
+  __device__ bool insert(const Key & k) {
+    if (first == nullptr) {
+      first = new Node;
+      first->num = k;
+      first->next = nullptr;
+      current = first;
+      return true;
+    }
+    if (this->find(k)) {
+      return false;
+    }
+    current->next = new Node;
+    current = current->next;
+    current->num = k;
+    current->next = nullptr;
+    return true;
+  }
+
+  __device__ bool find(const Key & value) {
+    Node * crt = first;
+    while (crt != nullptr) {
+      if (crt->num == value) {
+        return true;
+      }
+      crt = crt->next;
+    }
+    return false;
+  }
+
+  __device__ Node * get_first() {
+    return this->first;
+  }
+
+  __device__ size_t count(const Key & k) {
+    if (this->find(k)) {
+      return 1;
+    }
+    return 0;
+  }
+
+ private:
+  size_t total;
+  Node * first;
+  Node * current;
+};
+
+template<typename T, typename U>
+__device__ void ASSERT_EQ_dev(T a, U b) {
+  assert(a == b);
+}
+
+__global__ void insert_and_lookup(SkipList<Key, Comparator> * skipList) {
+
+  const int N = 2000;
+  const int R = 5000;
+  Random rnd(1000);
+
+  MemorySet<Key> keys;
+  Arena arena;
+  Comparator cmp;
+  SkipList<Key, Comparator> list(cmp, &arena);
+  for (int i = 0; i < N; i++) {
+    Key key = rnd.Next() % R;
+    if (keys.insert(key)) {
+      list.Insert(key);
+    }
+  }
+
+  for (int i = 0; i < R; i++) {
+    if (list.Contains(i)) {
+      ASSERT_EQ_dev(keys.count(i), 1);
+    } else {
+      ASSERT_EQ_dev(keys.count(i), 0);
+    }
+  }
+
+  Node * cur = keys.get_first();
+  while (cur != nullptr) {
+    assert(list.Contains(cur->num));
+    cur = cur->next;
+  }
+/*
+
+  // Forward iteration test
+  for (int i = 0; i < R; i++) {
+    SkipList<Key, Comparator>::Iterator iter(&list);
+    iter.Seek(i);
+
+    // Compare against model iterator
+    std::set<Key>::iterator model_iter = keys.lower_bound(i);
+    for (int j = 0; j < 3; j++) {
+      if (model_iter == keys.end()) {
+        ASSERT_TRUE(!iter.Valid());
+        break;
+      } else {
+        ASSERT_TRUE(iter.Valid());
+        ASSERT_EQ(*model_iter, iter.key());
+        ++model_iter;
+        iter.Next();
+      }
+    }
+  }
+
+  // Backward iteration test
+  {
+    SkipList<Key, Comparator>::Iterator iter(&list);
+    iter.SeekToLast();
+
+    // Compare against model iterator
+    for (std::set<Key>::reverse_iterator model_iter = keys.rbegin();
+         model_iter != keys.rend(); ++model_iter) {
+      ASSERT_TRUE(iter.Valid());
+      ASSERT_EQ(*model_iter, iter.key());
+      iter.Prev();
+    }
+    ASSERT_TRUE(!iter.Valid());
+  }
+*/
+
+}
+
 __global__ void init(Arena ** pArena, SkipList<Key, Comparator> ** pSkipList) {
   Comparator cmp;
   *pArena = new Arena();
@@ -324,6 +392,9 @@ __host__ void host_insert_test() {
 
   cudaDeviceSynchronize();
 
+  insert_and_lookup<<<gridSize, blockSize>>>(skipList);
+  cudaDeviceSynchronize();
+
 }