Added DevicePointer unit test.

src/UnitTests/Pointers/CMakeLists.txt

@@ -14,4 +14,8 @@ if( BUILD_CUDA )
                         OPTIONS ${CXX_TESTS_FLAGS} )
    TARGET_LINK_LIBRARIES( SharedPointerCudaTest ${GTEST_BOTH_LIBRARIES} )
    ADD_TEST( SharedPointerCudaTest ${EXECUTABLE_OUTPUT_PATH}/SharedPointerCudaTest${CMAKE_EXECUTABLE_SUFFIX} )
+   CUDA_ADD_EXECUTABLE( DevicePointerCudaTest DevicePointerCudaTest.cu 
+                        OPTIONS ${CXX_TESTS_FLAGS} )
+   TARGET_LINK_LIBRARIES( DevicePointerCudaTest ${GTEST_BOTH_LIBRARIES} )
+   ADD_TEST( DevicePointerCudaTest ${EXECUTABLE_OUTPUT_PATH}/DevicePointerCudaTest${CMAKE_EXECUTABLE_SUFFIX} )
 endif( BUILD_CUDA )

src/UnitTests/Pointers/DevicePointerCudaTest.cu

+/***************************************************************************
+                          DevicePointerCudaTest.cpp  -  description
+                             -------------------
+    begin                : Nov 26, 2019
+    copyright            : (C) 2019 by Tomas Oberhuber
+    email                : tomas.oberhuber@fjfi.cvut.cz
+ ***************************************************************************/
+
+/* See Copyright Notice in tnl/Copyright */
+
+#include <cstdlib>
+#include <TNL/Devices/Host.h>
+#include <TNL/Pointers/DevicePointer.h>
+#include <TNL/Containers/StaticArray.h>
+#include <TNL/Containers/Array.h>
+
+#ifdef HAVE_GTEST
+#include <gtest/gtest.h>
+#endif
+
+#include <TNL/Devices/Cuda.h>
+
+using namespace TNL;
+
+#ifdef HAVE_GTEST
+TEST( DevicePointerCudaTest, ConstructorTest )
+{
+#ifdef HAVE_CUDA
+   using TestType =  TNL::Containers::StaticArray< 2, int  >;
+   TestType obj1;
+   Pointers::DevicePointer< TestType, Devices::Cuda > ptr1( obj1 );
+
+   ptr1->x() = 0;
+   ptr1->y() = 0;
+   ASSERT_EQ( ptr1->x(), 0 );
+   ASSERT_EQ( ptr1->y(), 0 );
+
+   TestType obj2( 1,2 );
+   Pointers::DevicePointer< TestType, Devices::Cuda > ptr2( obj2 );
+   ASSERT_EQ( ptr2->x(), 1 );
+   ASSERT_EQ( ptr2->y(), 2 );
+
+   ptr1 = ptr2;
+   ASSERT_EQ( ptr1->x(), 1 );
+   ASSERT_EQ( ptr1->y(), 2 );
+#endif
+};
+
+TEST( DevicePointerCudaTest, getDataTest )
+{
+#ifdef HAVE_CUDA
+   using TestType = TNL::Containers::StaticArray< 2, int  >;
+   TestType obj1( 1, 2 );
+   Pointers::DevicePointer< TestType, Devices::Cuda > ptr1( obj1 );
+
+   Pointers::synchronizeSmartPointersOnDevice< Devices::Cuda >();
+
+   TestType aux;
+
+   cudaMemcpy( ( void*) &aux, &ptr1.getData< Devices::Cuda >(), sizeof( TestType ), cudaMemcpyDeviceToHost );
+
+   ASSERT_EQ( aux[ 0 ], 1 );
+   ASSERT_EQ( aux[ 1 ], 2 );
+#endif  // HAVE_CUDA
+};
+
+#ifdef HAVE_CUDA
+__global__ void copyArrayKernel( const TNL::Containers::Array< int, Devices::Cuda >* inArray,
+                                 int* outArray )
+{
+   if( threadIdx.x < 2 )
+   {
+      outArray[ threadIdx.x ] = ( *inArray )[ threadIdx.x ];
+   }
+}
+
+__global__ void copyArrayKernel2( const Pointers::DevicePointer< TNL::Containers::Array< int, Devices::Cuda > > inArray,
+                                  int* outArray )
+{
+   if( threadIdx.x < 2 )
+   {
+      outArray[ threadIdx.x ] = ( *inArray )[ threadIdx.x ];
+   }
+}
+#endif
+
+TEST( DevicePointerCudaTest, getDataArrayTest )
+{
+#ifdef HAVE_CUDA
+   using TestType = TNL::Containers::Array< int, Devices::Cuda  >;
+   TestType obj;
+   Pointers::DevicePointer< TestType > ptr( obj );
+
+   ptr->setSize( 2 );
+   ptr->setElement( 0, 1 );
+   ptr->setElement( 1, 2 );
+
+   Pointers::synchronizeSmartPointersOnDevice< Devices::Cuda >();
+
+   int *testArray_device, *testArray_host;
+   cudaMalloc( ( void** ) &testArray_device, 2 * sizeof( int ) );
+   copyArrayKernel<<< 1, 2 >>>( &ptr.getData< Devices::Cuda >(), testArray_device );
+   testArray_host = new int [ 2 ];
+   cudaMemcpy( testArray_host, testArray_device, 2 * sizeof( int ), cudaMemcpyDeviceToHost );
+
+   ASSERT_EQ( testArray_host[ 0 ], 1 );
+   ASSERT_EQ( testArray_host[ 1 ], 2 );
+
+   copyArrayKernel2<<< 1, 2 >>>( ptr, testArray_device );
+   cudaMemcpy( testArray_host, testArray_device, 2 * sizeof( int ), cudaMemcpyDeviceToHost );
+
+   ASSERT_EQ( testArray_host[ 0 ], 1 );
+   ASSERT_EQ( testArray_host[ 1 ], 2 );
+
+   delete[] testArray_host;
+   cudaFree( testArray_device );
+
+#endif
+};
+
+TEST( DevicePointerCudaTest, nullptrAssignement )
+{
+#ifdef HAVE_CUDA
+   using TestType = Pointers::DevicePointer< double, Devices::Cuda >;
+   double o1 = 5;
+   TestType p1( o1 ), p2( nullptr );
+
+   // This should not crash
+   p1 = p2;
+
+   ASSERT_FALSE( p1 );
+   ASSERT_FALSE( p2 );
+#endif
+}
+
+TEST( DevicePointerCudaTest, swap )
+{
+#ifdef HAVE_CUDA
+   using TestType = Pointers::DevicePointer< double, Devices::Cuda >;
+   double o1( 1 ), o2( 2 );
+   TestType p1( o1 ), p2( o2 );
+
+   p1.swap( p2 );
+
+   ASSERT_EQ( *p1, 2 );
+   ASSERT_EQ( *p2, 1 );
+#endif
+}
+
+#endif
+
+
+#include "../main.h"