 	CUDA_CHECK_ERRORS();
 	stop.synchronize();
 
-	std::vector<double> hostVector( n*p );
-	cudaMemcpy2D( &hostVector.front(), sizeof(double)*p, AB, pitchAB, sizeof(double)*p, n, cudaMemcpyDeviceToHost );
-
-	for( std::size_t i = 0; i < n; ++i ) {
-		std::cout << "ROW[" << i << "]";
-		for( std::size_t j = 0; j < p; ++j ) std::cout << " " << std::fixed << hostVector[i*p+j];
-		std::cout << std::endl;
-	}
+	//std::vector<double> hostVector( n*p );
+	//cudaMemcpy2D( &hostVector.front(), sizeof(double)*p, AB, pitchAB, sizeof(double)*p, n, cudaMemcpyDeviceToHost );
+	//for( std::size_t i = 0; i < n; ++i ) {
+	//	std::cout << "ROW[" << i << "]";
+	//	for( std::size_t j = 0; j < p; ++j ) std::cout << " " << std::fixed << hostVector[i*p+j];
+	//	std::cout << std::endl;
+	//}
 
 	cudaFree( A );
 	cudaFree( B );

	std::size_t pitchA,
	const T* B,
	std::size_t pitchB,
	std::size_t n, std::size_t m, std::size_t p,
	T* AB,
	std::size_t pitchAB
);
template<typename T> __global__ void matrixMultiply( const ecuda::matrix<T> A, const ecuda::matrix<T> B, ecuda::matrix<T> AB );


__global__ void matrixMultiply(	const ecuda::matrix<T> A, const ecuda::matrix<T> B,	ecuda::matrix<T> AB ) {
	const int x = blockIdx.x*blockDim.x+threadIdx.x; // row
	const int y = blockIdx.y*blockDim.y+threadIdx.y; // column
	if( x < A.number_rows() and y < B.number_columns() ) {
		T result = 0;
		for( std::size_t i = 0; i < A.number_columns(); ++i ) result += A[x][i] * B[i][y];
		AB[x][y] = result;
	}
}
