1 files changed, 52 insertions, 27 deletions

diff --git a/db/skiplist_test.cu b/db/skiplist_test.cu
index 512c6f7..b5511f2 100644
--- a/db/skiplist_test.cu
+++ b/db/skiplist_test.cu
@@ -393,20 +393,6 @@ class TestClass {
   cuda::atomic<size_t> atomic;
 };
 
-__global__ void init(Arena ** pArena, SkipList<Key, Comparator> ** pSkipList) {
-  Comparator cmp;
-  printf("init\n");
-  *pArena = new Arena();
-  //new TestClass();
-  printf("init arena\n");
-  // new = SkipList *
-  *pSkipList = new SkipList<Key, Comparator>(cmp, *pArena);
-  printf("init2\n");
-  //pSkipList->Insert(16807);
-  printf("skiplist: %p\n", *pSkipList);
-  printf("init3\n");
-}
-
 constexpr size_t SKIPLIST_TEST_SIZE = 10000;
 constexpr size_t TEST_STEP = SKIPLIST_TEST_SIZE / 10;
 
@@ -421,21 +407,60 @@ __global__ void testParallel(SkipList<Key, Comparator> * skipList, Key * keys) {
   printf("done: %u\n", start);
 }
 
+__global__ void testSingle(SkipList<Key, Comparator>* skipList, Key * keys) {
+  for (unsigned  i = 0; i < SKIPLIST_TEST_SIZE; i ++) {
+    skipList->Insert(keys[i]);
+  }
+}
+
 
-__global__ void test_keys_is_sort(SkipList<Key, Comparator> * skiplist, const Key * sorted_keys) {
+__global__ void testKeysIsEqualLists(SkipList<Key, Comparator> * skiplist, const Key * sorted_keys) {
   SkipList<Key, Comparator>::Iterator iter(skiplist);
 
   iter.SeekToFirst();
 
   for (unsigned i = 0; i < SKIPLIST_TEST_SIZE ; i++ ) {
-    printf("%lu %lu\n", iter.key(), sorted_keys[i]);
     assert(iter.key() == sorted_keys[i]);
     iter.Next();
   }
   assert(!iter.Valid());
 }
 
-void host_insert_test() {
+__global__ void initSkipList(Arena ** pArena, SkipList<Key, Comparator> ** pSkipList) {
+  Comparator cmp;
+  *pArena = new Arena();
+  *pSkipList = new SkipList<Key, Comparator>(cmp, *pArena);
+
+  printf("%p %p \n", *pArena, *pSkipList);
+
+}
+
+__global__ void freeSkipList(Arena *** pArena, SkipList<Key, Comparator> *** pSkipList) {
+    cudaFree(**pArena);
+    cudaFree(**pSkipList);
+    cudaFree(*pArena);
+    cudaFree(*pSkipList);
+}
+
+TEST(SkipTest, TestInitSkiplist) {
+
+  Arena ** pArena;
+  SkipList<Key, Comparator> ** pSkipList;
+
+  cudaMallocManaged((void**)&pArena, sizeof(void*));
+  cudaMallocManaged((void**)&pSkipList, sizeof(void*));
+
+    initSkipList<<<1, 1>>>(pArena, pSkipList);
+    cudaDeviceSynchronize();
+    //printf("%p %p\n", *pArena, *pSkipList);
+
+    cudaFree(*pArena);
+    cudaFree(*pSkipList);
+    cudaFree(pArena);
+    cudaFree(pSkipList);
+}
+
+TEST(SkipTest, TestCudaInsert) {
   //Key * keys;
   //SkipList<Key, Comparator> list(cmp, &arena);
   /*
@@ -476,22 +501,22 @@ void host_insert_test() {
 
   memcpy(sorted_keys, keys, SKIPLIST_TEST_SIZE * sizeof(Key));
   cudaMemcpy(device_keys, keys, SKIPLIST_TEST_SIZE * sizeof(Key), cudaMemcpyHostToDevice);
-  std::sort(sorted_keys, &sorted_keys[SKIPLIST_TEST_SIZE - 1]);
-
 
   dim3 gridSize(1, 1);
   dim3 blockSize(1, 1);
 
-  init<<<1, 1>>>(pArena, skipList);
+  initSkipList<<<1, 1>>>(pArena, skipList);
   //sleep(5);
   cudaDeviceSynchronize();
   //insert_skiplist<<<gridSize, blockSize>>>(skipList, device_rnd);
 
-  testParallel<<<gridSize, blockSize>>>(*skipList, device_keys);
+  //testParallel<<<gridSize, blockSize>>>(*skipList, device_keys);
+  testSingle<<<1, 1>>>(*skipList, device_keys);
   cudaDeviceSynchronize();
 
+  std::sort(sorted_keys, sorted_keys + SKIPLIST_TEST_SIZE);
   cudaMemcpy(device_keys, sorted_keys, SKIPLIST_TEST_SIZE * sizeof(Key), cudaMemcpyHostToDevice);
-  test_keys_is_sort<<<1, 1>>>(*skipList, device_keys);
+  testKeysIsEqualLists<<<1, 1>>>(*skipList, device_keys);
   //insert_and_lookup<<<gridSize, blockSize>>>(skipList);
   cudaDeviceSynchronize();
   cudaFree(device_keys);
@@ -499,11 +524,6 @@ void host_insert_test() {
   cudaFree(pArena);
 }
 
-
-TEST(SkipTest, TestCudaInsert) {
-  host_insert_test();
-}
-
 __global__ void test_empty() {
   Arena arena;
   Comparator cmp;
@@ -519,6 +539,11 @@ __global__ void test_empty() {
   assert(!iter.Valid());
 }
 
+TEST(SkipTest, TestCudaEmpty) {
+    test_empty<<<1, 1>>>();
+    cudaDeviceSynchronize();
+}
+
 __global__ void test_fail() {
   std::printf("show some message");
   assert(0);