From bc00b1252bfa32b87b4a6cce3ff270e1d278cf48 Mon Sep 17 00:00:00 2001 From: Scott Zuyderduyn Date: Thu, 28 Jan 2016 12:37:18 -0500 Subject: [PATCH] last minute fix to get naked pointer iterators to work, and remove dep to estd:: library for test/ progs --- include/ecuda/algo/copy.hpp | 3 +- include/ecuda/algo/find_if.hpp | 1 + include/ecuda/global.hpp | 29 ++++--- include/ecuda/iterator.hpp | 30 +++---- include/ecuda/type_traits.hpp | 1 + test/test_algorithms.cu | 143 ++++++++++++++++----------------- test/test_matrix.cu | 10 +-- 7 files changed, 110 insertions(+), 107 deletions(-) diff --git a/include/ecuda/algo/copy.hpp b/include/ecuda/algo/copy.hpp index 1ae3b46..3c3bc51 100644 --- a/include/ecuda/algo/copy.hpp +++ b/include/ecuda/algo/copy.hpp @@ -667,7 +667,7 @@ __HOST__ __DEVICE__ inline OutputIterator copy( { // memory is now guaranteed to be regularly aligned so we can use cudaMemcpy2D typedef typename ecuda::add_pointer::type pointer; - pointer dest = naked_cast( result.operator->() ); + pointer dest = get_iterator_pointer( result ); typedef typename ecuda::add_pointer::type const_pointer; const_pointer src = naked_cast( first.operator->() ); @@ -805,6 +805,7 @@ __HOST__ __DEVICE__ inline OutputIterator copy( InputIterator first, InputIterat return impl::copy( first, last, result, ecuda::pair() ); } + } // namespace ecuda #endif diff --git a/include/ecuda/algo/find_if.hpp b/include/ecuda/algo/find_if.hpp index 035da95..afcc189 100644 --- a/include/ecuda/algo/find_if.hpp +++ b/include/ecuda/algo/find_if.hpp @@ -52,6 +52,7 @@ namespace ecuda { /// \cond DEVELOPER_DOCUMENTATION namespace impl { +ECUDA_SUPPRESS_HD_WARNINGS template __HOST__ __DEVICE__ InputIterator find_if( InputIterator first, InputIterator last, UnaryPredicate p, ecuda::true_type ) // device memory diff --git a/include/ecuda/global.hpp b/include/ecuda/global.hpp index ba72a54..a28e838 100644 --- a/include/ecuda/global.hpp +++ b/include/ecuda/global.hpp @@ -70,16 +70,17 @@ either expressed or implied, of the FreeBSD Project. #define ECUDA_CPP11_AVAILABLE #endif +#ifdef __CUDACC__ +// Macro function currently throws an ecuda::cuda_error exception containing a +// description of the problem error code. +#define CUDA_CALL(x) do { if((x)!=cudaSuccess) { std::ostringstream oss; oss << __FILE__; oss << ":"; oss << __LINE__; oss << " "; oss << cudaGetErrorString(cudaGetLastError()); throw ::ecuda::cuda_error(x,oss.str()); /*std::runtime_error(oss.str());*/ }} while(0); +#else /// /// Macro function that captures a CUDA error code and then does something /// with it. All calls to functions in the CUDA API that return an error code /// should use this. /// #define CUDA_CALL(x) x // cannot do CUDA calls when emulating with host only -#ifdef __CUDACC__ -// Macro function currently throws an ecuda::cuda_error exception containing a -// description of the problem error code. -#define CUDA_CALL(x) do { if((x)!=cudaSuccess) { std::ostringstream oss; oss << __FILE__; oss << ":"; oss << __LINE__; oss << " "; oss << cudaGetErrorString(cudaGetLastError()); throw ::ecuda::cuda_error(x,oss.str()); /*std::runtime_error(oss.str());*/ }} while(0); #endif #define S(x) #x @@ -90,6 +91,9 @@ either expressed or implied, of the FreeBSD Project. /// #define EXCEPTION_MSG(x) "" __FILE__ ":" S__LINE__ " " x +#ifdef __CUDACC__ +#define CUDA_CHECK_ERRORS() do { cudaError_t error = cudaGetLastError(); if( error != cudaSuccess ) throw ::ecuda::cuda_error(error,std::string(cudaGetErrorString(error))); } while(0); +#else /// /// Macro that performs a check for any outstanding CUDA errors. This macro /// should be declared after any CUDA API calls that do not return an error code @@ -97,17 +101,8 @@ either expressed or implied, of the FreeBSD Project. /// has not been made is safe. /// #define CUDA_CHECK_ERRORS() do {} while(0); // cannot check CUDA errors when emulating with host only -#ifdef __CUDACC__ -#define CUDA_CHECK_ERRORS() do { cudaError_t error = cudaGetLastError(); if( error != cudaSuccess ) throw ::ecuda::cuda_error(error,std::string(cudaGetErrorString(error))); } while(0); #endif -/// -/// Macro that calls a CUDA kernel function, waits for completion, and throws -/// an ecuda::cuda_error exception if any errors are reported by cudaGetLastError(). -/// -#define CUDA_CALL_KERNEL_AND_WAIT(...) do {\ - __VA_ARGS__;\ - } while( 0 ); // cannot do CUDA calls when emulating with host only #ifdef __CUDACC__ #define CUDA_CALL_KERNEL_AND_WAIT(...) do {\ __VA_ARGS__;\ @@ -115,6 +110,14 @@ either expressed or implied, of the FreeBSD Project. cudaDeviceSynchronize();\ { cudaError_t error = cudaGetLastError(); if( error != cudaSuccess ) throw ::ecuda::cuda_error(error,std::string(cudaGetErrorString(error))); }\ } while(0); +#else +/// +/// Macro that calls a CUDA kernel function, waits for completion, and throws +/// an ecuda::cuda_error exception if any errors are reported by cudaGetLastError(). +/// +#define CUDA_CALL_KERNEL_AND_WAIT(...) do {\ + __VA_ARGS__;\ + } while( 0 ); // cannot do CUDA calls when emulating with host only #endif /** Replace nullptr with NULL if nvcc still doesn't support C++11. */ diff --git a/include/ecuda/iterator.hpp b/include/ecuda/iterator.hpp index 228d267..22c1771 100644 --- a/include/ecuda/iterator.hpp +++ b/include/ecuda/iterator.hpp @@ -174,13 +174,13 @@ class device_contiguous_iterator : public device_iterator( const reverse_device_iterator& other ) const { return parentIterator > other.parentIterator; } __HOST__ __DEVICE__ inline bool operator<=( const reverse_device_iterator& other ) const { return operator<(other) || operator==(other); } __HOST__ __DEVICE__ inline bool operator>=( const reverse_device_iterator& other ) const { return operator>(other) || operator==(other); } - __HOST__ __DEVICE__ inline reverse_device_iterator& operator+=( int x ) { parentIterator -= x; return *this; } - __HOST__ __DEVICE__ inline reverse_device_iterator& operator-=( int x ) { parentIterator += x; return *this; } + __HOST__ __DEVICE__ inline reverse_device_iterator& operator+=( difference_type x ) { parentIterator -= x; return *this; } + __HOST__ __DEVICE__ inline reverse_device_iterator& operator-=( difference_type x ) { parentIterator += x; return *this; } - __DEVICE__ reference operator[]( int x ) const { return parentIterator.operator[]( -x-1 ); } + __DEVICE__ reference operator[]( difference_type x ) const { return parentIterator.operator[]( -x-1 ); } __HOST__ __DEVICE__ reverse_device_iterator& operator=( const reverse_device_iterator& other ) { diff --git a/include/ecuda/type_traits.hpp b/include/ecuda/type_traits.hpp index 7211af2..be7e24f 100644 --- a/include/ecuda/type_traits.hpp +++ b/include/ecuda/type_traits.hpp @@ -199,6 +199,7 @@ template class striding_padded_ptr; // forward declaratio /// so that a cast to a naked pointer T* is achieved by reinterpret_cast(ptr.get().get()). /// template __HOST__ __DEVICE__ T naked_cast( U* ptr ) { return reinterpret_cast(ptr); } +template __HOST__ __DEVICE__ T naked_cast( T* ptr ) { return ptr; } //template __HOST__ __DEVICE__ T naked_cast( const naked_ptr& ptr ) { return naked_cast(ptr.get()); } template __HOST__ __DEVICE__ T naked_cast( const unique_ptr& ptr ) { return naked_cast(ptr.get()); } template __HOST__ __DEVICE__ T naked_cast( const shared_ptr& ptr ) { return naked_cast(ptr.get()); } diff --git a/test/test_algorithms.cu b/test/test_algorithms.cu index 20132c9..4873157 100644 --- a/test/test_algorithms.cu +++ b/test/test_algorithms.cu @@ -4,12 +4,16 @@ #include #include "../include/ecuda/ecuda.hpp" -#include -#include // https://github.com/philsquared/Catch #include +//const std::size_t N = 1000; +const std::size_t R = 5; +const std::size_t C = 5; +const std::size_t D = 5; + + #ifdef __CUDACC__ template __global__ void testIterators( const typename ecuda::matrix::kernel src, typename ecuda::matrix::kernel dest ) @@ -118,12 +122,12 @@ std::ostream& operator<<( std::ostream& out, const ecuda::matrix& mat ) } template -std::ostream& operator<<( std::ostream& out, const estd::matrix& mat ) +std::ostream& operator<<( std::ostream& out, const T** mat ) { - for( std::size_t i = 0; i < mat.number_rows(); ++i ) { - for( std::size_t j = 0; j < mat.number_columns(); ++j ) { + for( std::size_t i = 0; i < R; ++i ) { + for( std::size_t j = 0; j < C; ++j ) { if( j ) out << " "; - out << mat(i,j); + out << mat[i][j]; } out << std::endl; } @@ -131,14 +135,14 @@ std::ostream& operator<<( std::ostream& out, const estd::matrix& mat ) } template -std::ostream& operator<<( std::ostream& out, const ecuda::cube& cbe ) +std::ostream& operator<<( std::ostream& out, const T*** cbe ) { - for( std::size_t i = 0; i < cbe.number_rows(); ++i ) { + for( std::size_t i = 0; i < R; ++i ) { out << "ROW[" << i << "]:" << std::endl; - for( std::size_t j = 0; j < cbe.number_columns(); ++j ) { - for( std::size_t k = 0; k < cbe.number_depths(); ++k ) { + for( std::size_t j = 0; j < C; ++j ) { + for( std::size_t k = 0; k < D; ++k ) { if( k ) out << " "; - out << cbe(i,j,k); + out << cbe[i][j][k]; } out << std::endl; } @@ -147,7 +151,7 @@ std::ostream& operator<<( std::ostream& out, const ecuda::cube& cbe ) } template -std::ostream& operator<<( std::ostream& out, const estd::cube& cbe ) +std::ostream& operator<<( std::ostream& out, const ecuda::cube& cbe ) { for( std::size_t i = 0; i < cbe.number_rows(); ++i ) { out << "ROW[" << i << "]:" << std::endl; @@ -162,25 +166,20 @@ std::ostream& operator<<( std::ostream& out, const estd::cube& cbe ) return out; } -//const std::size_t N = 1000; -const std::size_t R = 5; -const std::size_t C = 5; -const std::size_t D = 5; - void test_reverse() { data_type hostArray[R]; std::vector hostVector( R ); - estd::matrix hostMatrix( R, C ); - estd::cube hostCube( R, C, D ); + matrix_index hostMatrix[R][C]; + cube_index hostCube[R][C][D]; for( std::size_t i = 0; i < R; ++i ) { hostArray[i] = data_type(i); hostVector[i] = data_type(i); for( std::size_t j = 0; j < C; ++j ) { - hostMatrix(i,j) = matrix_index(i,j); + hostMatrix[i][j] = matrix_index(i,j); for( std::size_t k = 0; k < D; ++k ) { - hostCube(i,j,k) = cube_index(i,j,k); + hostCube[i][j][k] = cube_index(i,j,k); } } } @@ -191,15 +190,15 @@ void test_reverse() ecuda::cube deviceCube( R, C, D ); ecuda::copy( hostArray, hostArray+R, deviceArray.begin() ); ecuda::copy( hostVector.begin(), hostVector.end(), deviceVector.begin() ); - ecuda::copy( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::copy( hostCube.begin(), hostCube.end(), deviceCube.begin() ); + ecuda::copy( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); + ecuda::copy( hostCube[0][0], hostCube[R-1][C-1]+D, deviceCube.begin() ); // reverse { ecuda::reverse( hostArray, hostArray+R ); ecuda::reverse( hostVector.begin(), hostVector.end() ); - ecuda::reverse( hostMatrix.begin(), hostMatrix.end() ); - ecuda::reverse( hostCube.begin(), hostCube.end() ); + ecuda::reverse( hostMatrix[0], hostMatrix[R-1]+C ); + ecuda::reverse( hostCube[0][0], hostCube[R-1][C-1]+D ); ecuda::reverse( deviceArray.begin(), deviceArray.end() ); ecuda::reverse( deviceVector.begin(), deviceVector.end() ); ecuda::reverse( deviceMatrix.begin(), deviceMatrix.end() ); @@ -215,15 +214,15 @@ void test_find() { data_type hostArray[R]; std::vector hostVector( R ); - estd::matrix hostMatrix( R, C ); - estd::cube hostCube( R, C, D ); + matrix_index hostMatrix[R][C]; + cube_index hostCube[R][C][D]; for( std::size_t i = 0; i < R; ++i ) { hostArray[i] = data_type(i); hostVector[i] = data_type(i); for( std::size_t j = 0; j < C; ++j ) { - hostMatrix(i,j) = matrix_index(i,j); + hostMatrix[i][j] = matrix_index(i,j); for( std::size_t k = 0; k < D; ++k ) { - hostCube(i,j,k) = cube_index(i,j,k); + hostCube[i][j][k] = cube_index(i,j,k); } } } @@ -234,14 +233,14 @@ void test_find() ecuda::cube deviceCube( R, C, D ); ecuda::copy( hostArray, hostArray+R, deviceArray.begin() ); ecuda::copy( hostVector.begin(), hostVector.end(), deviceVector.begin() ); - ecuda::copy( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::copy( hostCube.begin(), hostCube.end(), deviceCube.begin() ); + ecuda::copy( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); + ecuda::copy( hostCube[0][0], hostCube[R-1][C-1]+D, deviceCube.begin() ); std::cout << "H(1) = " << ( ecuda::find( hostArray, hostArray+R, 1 ) - hostArray ) << std::endl; std::cout << "D(1) = " << ( ecuda::find( deviceArray.begin(), deviceArray.end(), 1 ) - deviceArray.begin() ) << std::endl; - std::cout << "H(1,1) = " << ( ecuda::find( hostMatrix.begin(), hostMatrix.end(), matrix_index(1,1) ) - hostMatrix.begin() ) << std::endl; + std::cout << "H(1,1) = " << ( ecuda::find( hostMatrix[0], hostMatrix[R-1]+C, matrix_index(1,1) ) - hostMatrix[0] ) << std::endl; std::cout << "D(1,1) = " << ( ecuda::find( deviceMatrix.begin(), deviceMatrix.end(), matrix_index(1,1) ) - deviceMatrix.begin() ) << std::endl; - std::cout << "H(1,1,1) = " << ( ecuda::find( hostCube.begin(), hostCube.end(), cube_index(1,1,1) ) - hostCube.begin() ) << std::endl; + std::cout << "H(1,1,1) = " << ( ecuda::find( hostCube[0][0], hostCube[R-1][C-1]+D, cube_index(1,1,1) ) - hostCube[0][0] ) << std::endl; std::cout << "D(1,1,1) = " << ( ecuda::find( deviceCube.begin(), deviceCube.end(), cube_index(1,1,1) ) - deviceCube.begin() ) << std::endl; } @@ -251,15 +250,15 @@ int main( int, char** ) data_type hostArray[R]; std::vector hostVector( R ); - estd::matrix hostMatrix( R, C ); - estd::cube hostCube( R, C, D ); + matrix_index hostMatrix[R][C]; + cube_index hostCube[R][C][D]; for( std::size_t i = 0; i < R; ++i ) { hostArray[i] = data_type(i); hostVector[i] = data_type(i); for( std::size_t j = 0; j < C; ++j ) { - hostMatrix(i,j) = matrix_index(i,j); + hostMatrix[i][j] = matrix_index(i,j); for( std::size_t k = 0; k < D; ++k ) { - hostCube(i,j,k) = cube_index(i,j,k); + hostCube[i][j][k] = cube_index(i,j,k); } } } @@ -297,28 +296,28 @@ int main( int, char** ) std::cout << " " << std::boolalpha << ecuda::equal( hostVector.begin(), hostVector.end(), deviceVector.begin() ); std::cout << std::endl; - ecuda::copy( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); + ecuda::copy( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); std::cout << " matrix =>"; - std::cout << " " << std::boolalpha << ecuda::equal( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::fill( hostMatrix.begin(), hostMatrix.end(), matrix_index() ); - std::cout << " " << std::boolalpha << !ecuda::equal( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::copy( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix.begin() ); - std::cout << " " << std::boolalpha << ecuda::equal( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::fill( hostMatrix.begin(), hostMatrix.end(), matrix_index() ); + std::cout << " " << std::boolalpha << ecuda::equal( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); + ecuda::fill( hostMatrix[0], hostMatrix[R-1]+C, matrix_index() ); + std::cout << " " << std::boolalpha << !ecuda::equal( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); + ecuda::copy( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix[0] ); + std::cout << " " << std::boolalpha << ecuda::equal( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); + ecuda::fill( hostMatrix[0], hostMatrix[R-1]+C, matrix_index() ); ecuda::fill( deviceMatrix.begin(), deviceMatrix.end(), matrix_index() ); - std::cout << " " << std::boolalpha << ecuda::equal( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); + std::cout << " " << std::boolalpha << ecuda::equal( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); std::cout << std::endl; - ecuda::copy( hostCube.begin(), hostCube.end(), deviceCube.begin() ); + ecuda::copy( hostCube[0][0], hostCube[R-1][C-1]+D, deviceCube.begin() ); std::cout << " cube =>"; - std::cout << " " << std::boolalpha << ecuda::equal( hostCube.begin(), hostCube.end(), deviceCube.begin() ); - ecuda::fill( hostCube.begin(), hostCube.end(), cube_index() ); - std::cout << " " << std::boolalpha << !ecuda::equal( hostCube.begin(), hostCube.end(), deviceCube.begin() ); - ecuda::copy( deviceCube.begin(), deviceCube.end(), hostCube.begin() ); - std::cout << " " << std::boolalpha << ecuda::equal( hostCube.begin(), hostCube.end(), deviceCube.begin() ); - ecuda::fill( hostCube.begin(), hostCube.end(), cube_index() ); + std::cout << " " << std::boolalpha << ecuda::equal( hostCube[0][0], hostCube[R-1][C-1]+D, deviceCube.begin() ); + ecuda::fill( hostCube[0][0], hostCube[R-1][C-1]+D, cube_index() ); + std::cout << " " << std::boolalpha << !ecuda::equal( hostCube[0][0], hostCube[R-1][C-1]+D, deviceCube.begin() ); + ecuda::copy( deviceCube.begin(), deviceCube.end(), hostCube[0][0] ); + std::cout << " " << std::boolalpha << ecuda::equal( hostCube[0][0], hostCube[R-1][C-1]+D, deviceCube.begin() ); + ecuda::fill( hostCube[0][0], hostCube[R-1][C-1]+D, cube_index() ); ecuda::fill( deviceCube.begin(), deviceCube.end(), cube_index() ); - std::cout << " " << std::boolalpha << ecuda::equal( hostCube.begin(), hostCube.end(), deviceCube.begin() ); + std::cout << " " << std::boolalpha << ecuda::equal( hostCube[0][0], hostCube[R-1][C-1]+D, deviceCube.begin() ); std::cout << std::endl; std::cout << std::endl; @@ -326,31 +325,31 @@ int main( int, char** ) // copy { - ecuda::copy( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::copy( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix.begin() ); - ecuda::copy( hostMatrix.begin(), hostMatrix.end(), hostMatrix.begin() ); + ecuda::copy( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); + ecuda::copy( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix[0] ); + ecuda::copy( hostMatrix[0], hostMatrix[R-1]+C, hostMatrix[0] ); ecuda::copy( deviceMatrix.begin(), deviceMatrix.end(), deviceMatrix.begin() ); } // equal { - ecuda::equal( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::equal( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix.begin() ); - ecuda::equal( hostMatrix.begin(), hostMatrix.end(), hostMatrix.begin() ); + ecuda::equal( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); + ecuda::equal( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix[0] ); + ecuda::equal( hostMatrix[0], hostMatrix[R-1]+C, hostMatrix[0] ); ecuda::equal( deviceMatrix.begin(), deviceMatrix.end(), deviceMatrix.begin() ); } // fill { - ecuda::fill( hostMatrix.begin(), hostMatrix.end(), matrix_index() ); + ecuda::fill( hostMatrix[0], hostMatrix[R-1]+C, matrix_index() ); ecuda::fill( deviceMatrix.begin(), deviceMatrix.end(), matrix_index() ); } // lexicographical_compare { - ecuda::lexicographical_compare( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin(), deviceMatrix.end() ); - ecuda::lexicographical_compare( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix.begin(), hostMatrix.end() ); - ecuda::lexicographical_compare( hostMatrix.begin(), hostMatrix.end(), hostMatrix.begin(), hostMatrix.end() ); + ecuda::lexicographical_compare( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin(), deviceMatrix.end() ); + ecuda::lexicographical_compare( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix[0], hostMatrix[R-1]+C ); + ecuda::lexicographical_compare( hostMatrix[0], hostMatrix[R-1]+C, hostMatrix[0], hostMatrix[R-1]+C ); ecuda::lexicographical_compare( deviceMatrix.begin(), deviceMatrix.end(), deviceMatrix.begin(), deviceMatrix.end() ); } @@ -359,7 +358,7 @@ int main( int, char** ) // find_if { - ecuda::find_if( hostMatrix.begin(), hostMatrix.end(), UnaryPredicate() ); + ecuda::find_if( hostMatrix[0], hostMatrix[R-1]+C, UnaryPredicate() ); ecuda::find_if( deviceMatrix.begin(), deviceMatrix.end(), UnaryPredicate() ); } @@ -367,39 +366,39 @@ int main( int, char** ) // any_of { - ecuda::any_of( hostMatrix.begin(), hostMatrix.end(), UnaryPredicate() ); + ecuda::any_of( hostMatrix[0], hostMatrix[R-1]+C, UnaryPredicate() ); ecuda::any_of( deviceMatrix.begin(), deviceMatrix.end(), UnaryPredicate() ); } // none_of { - ecuda::none_of( hostMatrix.begin(), hostMatrix.end(), UnaryPredicate() ); + ecuda::none_of( hostMatrix[0], hostMatrix[R-1]+C, UnaryPredicate() ); ecuda::none_of( deviceMatrix.begin(), deviceMatrix.end(), UnaryPredicate() ); } // for_each { - ecuda::for_each( hostMatrix.begin(), hostMatrix.end(), UnaryPredicate() ); + ecuda::for_each( hostMatrix[0], hostMatrix[R-1]+C, UnaryPredicate() ); ecuda::for_each( deviceMatrix.begin(), deviceMatrix.end(), UnaryPredicate() ); } // count { - ecuda::count( hostMatrix.begin(), hostMatrix.end(), 0 ); + ecuda::count( hostMatrix[0], hostMatrix[R-1]+C, 0 ); ecuda::count( deviceMatrix.begin(), deviceMatrix.end(), 0 ); } // count_if { - ecuda::count_if( hostMatrix.begin(), hostMatrix.end(), UnaryPredicate() ); + ecuda::count_if( hostMatrix[0], hostMatrix[R-1]+C, UnaryPredicate() ); ecuda::count_if( deviceMatrix.begin(), deviceMatrix.end(), UnaryPredicate() ); } // mismatch { - ecuda::mismatch( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::mismatch( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix.begin() ); - ecuda::mismatch( hostMatrix.begin(), hostMatrix.end(), hostMatrix.begin() ); + ecuda::mismatch( hostMatrix[0], hostMatrix[R-1]+C, deviceMatrix.begin() ); + ecuda::mismatch( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix[0] ); + ecuda::mismatch( hostMatrix[0], hostMatrix[R-1]+C, hostMatrix[0] ); ecuda::mismatch( deviceMatrix.begin(), deviceMatrix.end(), deviceMatrix.begin() ); } diff --git a/test/test_matrix.cu b/test/test_matrix.cu index c60ff18..cba0cb9 100644 --- a/test/test_matrix.cu +++ b/test/test_matrix.cu @@ -9,8 +9,6 @@ #include "../include/ecuda/matrix.hpp" #include "../include/ecuda/vector.hpp" -#include - #ifdef __CUDACC__ template __global__ void testIterators( const typename ecuda::matrix::kernel src, typename ecuda::matrix::kernel dest ) { @@ -89,11 +87,11 @@ void dummyFunction( typename ecuda::matrix::kernel_argument arg ) int main( int argc, char* argv[] ) { { - estd::matrix hostMatrix(5,5); + int hostMatrix[5][5]; ecuda::matrix deviceMatrix(5,5); - ecuda::copy( hostMatrix.begin(), hostMatrix.end(), deviceMatrix.begin() ); - ecuda::copy( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix.begin() ); - ecuda::copy( hostMatrix.begin(), hostMatrix.end(), hostMatrix.begin() ); + ecuda::copy( hostMatrix[0], hostMatrix[0]+5, deviceMatrix.begin() ); + ecuda::copy( deviceMatrix.begin(), deviceMatrix.end(), hostMatrix[0] ); + ecuda::copy( hostMatrix[0], hostMatrix[0]+5, hostMatrix[0] ); ecuda::copy( deviceMatrix.begin(), deviceMatrix.end(), deviceMatrix.begin() ); } {