From ad01b6e3144843ccb16ac99bd12fd1ab1fc7a7b5 Mon Sep 17 00:00:00 2001
From: =?UTF-8?q?Tom=C3=A1=C5=A1=20Oberhuber?= <oberhuber.tomas@gmail.com>
Date: Thu, 7 Mar 2019 21:41:06 +0100
Subject: [PATCH] Added type conversion to File read and write.
---
src/TNL/File.hpp | 145 +++++++++++++++++++++++++++------------
src/UnitTests/FileTest.h | 85 +++++++++++++++++++++++
2 files changed, 186 insertions(+), 44 deletions(-)
diff --git a/src/TNL/File.hpp b/src/TNL/File.hpp
index bbff0ab960..37ca5ef2db 100644
--- a/src/TNL/File.hpp
+++ b/src/TNL/File.hpp
@@ -88,7 +88,7 @@ bool File::read( Type* buffer, std::streamsize elements )
if( ! elements )
return true;
- return read_impl< Type, Device >( buffer, elements );
+ return read_impl< Type, Device, SourceType >( buffer, elements );
}
// Host
@@ -150,7 +150,7 @@ bool File::read_impl( Type* buffer, std::streamsize elements )
else
{
const std::streamsize cast_buffer_size = std::min( TransferBufferSize / (std::streamsize) sizeof(SourceType), elements );
- using BaseType = typename std::remove_cv< SorceType >::type;
+ using BaseType = typename std::remove_cv< SourceType >::type;
std::unique_ptr< BaseType[] > cast_buffer{ new BaseType[ cast_buffer_size ] };
while( readElements < elements )
@@ -186,32 +186,40 @@ bool File::read_impl( Type* buffer, std::streamsize elements )
std::unique_ptr< BaseType[] > host_buffer{ new BaseType[ host_buffer_size ] };
std::streamsize readElements = 0;
- while( readElements < elements )
+ if( std::is_same< Type, SourceType >::value )
{
- const std::streamsize transfer = std::min( elements - readElements, host_buffer_size );
- file.read( reinterpret_cast<char*>(host_buffer.get()), sizeof(Type) * transfer );
-
- Devices::MICHider<Type> device_buff;
- device_buff.pointer=buffer;
- #pragma offload target(mic) in(device_buff,readElements) in(host_buffer:length(transfer))
+ while( readElements < elements )
{
- /*
- for(int i=0;i<transfer;i++)
- device_buff.pointer[readElements+i]=host_buffer[i];
- */
- memcpy(&(device_buff.pointer[readElements]), host_buffer.get(), transfer*sizeof(Type) );
- }
+ const std::streamsize transfer = std::min( elements - readElements, host_buffer_size );
+ file.read( reinterpret_cast<char*>(host_buffer.get()), sizeof(Type) * transfer );
+
+ Devices::MICHider<Type> device_buff;
+ device_buff.pointer=buffer;
+ #pragma offload target(mic) in(device_buff,readElements) in(host_buffer:length(transfer))
+ {
+ /*
+ for(int i=0;i<transfer;i++)
+ device_buff.pointer[readElements+i]=host_buffer[i];
+ */
+ memcpy(&(device_buff.pointer[readElements]), host_buffer.get(), transfer*sizeof(Type) );
+ }
- readElements += transfer;
+ readElements += transfer;
+ }
+ free( host_buffer );
+ }
+ else
+ {
+ std::cerr << "Type conversion during loading is not implemented for MIC." << std::endl;
+ abort();
}
- free( host_buffer );
return true;
#else
throw Exceptions::MICSupportMissing();
#endif
}
-template< class Type, typename Device, typename TargeType >
+template< class Type, typename Device, typename TargetType >
bool File::write( const Type* buffer, std::streamsize elements )
{
TNL_ASSERT_GE( elements, 0, "Number of elements to write must be non-negative." );
@@ -219,7 +227,7 @@ bool File::write( const Type* buffer, std::streamsize elements )
if( ! elements )
return true;
- return write_impl< Type, Device >( buffer, elements );
+ return write_impl< Type, Device, TargetType >( buffer, elements );
}
// Host
@@ -229,7 +237,24 @@ template< typename Type,
typename >
bool File::write_impl( const Type* buffer, std::streamsize elements )
{
- file.write( reinterpret_cast<const char*>(buffer), sizeof(Type) * elements );
+ if( std::is_same< Type, TargetType >::value )
+ file.write( reinterpret_cast<const char*>(buffer), sizeof(Type) * elements );
+ else
+ {
+ const std::streamsize cast_buffer_size = std::min( TransferBufferSize / (std::streamsize) sizeof(TargetType), elements );
+ using BaseType = typename std::remove_cv< TargetType >::type;
+ std::unique_ptr< BaseType[] > cast_buffer{ new BaseType[ cast_buffer_size ] };
+ std::streamsize writtenElements = 0;
+ while( writtenElements < elements )
+ {
+ const std::streamsize transfer = std::min( elements - writtenElements, cast_buffer_size );
+ for( std::streamsize i = 0; i < transfer; i++ )
+ cast_buffer[ i ] = static_cast< TargetType >( buffer[ writtenElements ++ ] );
+ file.write( reinterpret_cast<char*>(cast_buffer.get()), sizeof(TargetType) * transfer );
+ writtenElements += transfer;
+ }
+
+ }
return true;
}
@@ -246,16 +271,40 @@ bool File::write_impl( const Type* buffer, std::streamsize elements )
std::unique_ptr< BaseType[] > host_buffer{ new BaseType[ host_buffer_size ] };
std::streamsize writtenElements = 0;
- while( writtenElements < elements )
+ if( std::is_same< Type, TargetType >::value )
+ {
+ while( writtenElements < elements )
+ {
+ const std::streamsize transfer = std::min( elements - writtenElements, host_buffer_size );
+ cudaMemcpy( (void*) host_buffer.get(),
+ (void*) &buffer[ writtenElements ],
+ transfer * sizeof(Type),
+ cudaMemcpyDeviceToHost );
+ TNL_CHECK_CUDA_DEVICE;
+ file.write( reinterpret_cast<const char*>(host_buffer.get()), sizeof(Type) * transfer );
+ writtenElements += transfer;
+ }
+ }
+ else
{
- const std::streamsize transfer = std::min( elements - writtenElements, host_buffer_size );
- cudaMemcpy( (void*) host_buffer.get(),
- (void*) &buffer[ writtenElements ],
- transfer * sizeof(Type),
- cudaMemcpyDeviceToHost );
- TNL_CHECK_CUDA_DEVICE;
- file.write( reinterpret_cast<const char*>(host_buffer.get()), sizeof(Type) * transfer );
- writtenElements += transfer;
+ const std::streamsize cast_buffer_size = std::min( TransferBufferSize / (std::streamsize) sizeof(TargetType), elements );
+ using BaseType = typename std::remove_cv< TargetType >::type;
+ std::unique_ptr< BaseType[] > cast_buffer{ new BaseType[ cast_buffer_size ] };
+
+ while( writtenElements < elements )
+ {
+ const std::streamsize transfer = std::min( elements - writtenElements, host_buffer_size );
+ cudaMemcpy( (void*) host_buffer.get(),
+ (void*) &buffer[ writtenElements ],
+ transfer * sizeof(Type),
+ cudaMemcpyDeviceToHost );
+ TNL_CHECK_CUDA_DEVICE;
+ for( std::streamsize i = 0; i < transfer; i++ )
+ cast_buffer[ i ] = static_cast< TargetType >( host_buffer[ i ] );
+
+ file.write( reinterpret_cast<const char*>(cast_buffer.get()), sizeof(TargetType) * transfer );
+ writtenElements += transfer;
+ }
}
return true;
#else
@@ -276,24 +325,32 @@ bool File::write_impl( const Type* buffer, std::streamsize elements )
std::unique_ptr< BaseType[] > host_buffer{ new BaseType[ host_buffer_size ] };
std::streamsize writtenElements = 0;
- while( this->writtenElements < elements )
+ if( std::is_same< Type, TargetType >::value )
{
- const std::streamsize transfer = std::min( elements - writtenElements, host_buffer_size );
-
- Devices::MICHider<const Type> device_buff;
- device_buff.pointer=buffer;
- #pragma offload target(mic) in(device_buff,writtenElements) out(host_buffer:length(transfer))
+ while( this->writtenElements < elements )
{
- //THIS SHOULD WORK... BUT NOT WHY?
- /*for(int i=0;i<transfer;i++)
- host_buffer[i]=device_buff.pointer[writtenElements+i];
- */
-
- memcpy(host_buffer.get(), &(device_buff.pointer[writtenElements]), transfer*sizeof(Type) );
+ const std::streamsize transfer = std::min( elements - writtenElements, host_buffer_size );
+
+ Devices::MICHider<const Type> device_buff;
+ device_buff.pointer=buffer;
+ #pragma offload target(mic) in(device_buff,writtenElements) out(host_buffer:length(transfer))
+ {
+ //THIS SHOULD WORK... BUT NOT WHY?
+ /*for(int i=0;i<transfer;i++)
+ host_buffer[i]=device_buff.pointer[writtenElements+i];
+ */
+
+ memcpy(host_buffer.get(), &(device_buff.pointer[writtenElements]), transfer*sizeof(Type) );
+ }
+
+ file.write( reinterpret_cast<const char*>(host_buffer.get()), sizeof(Type) * transfer );
+ writtenElements += transfer;
}
-
- file.write( reinterpret_cast<const char*>(host_buffer.get()), sizeof(Type) * transfer );
- writtenElements += transfer;
+ }
+ else
+ {
+ std::cerr << "Type conversion during saving is not implemented for MIC." << std::endl;
+ abort();
}
return true;
#else
diff --git a/src/UnitTests/FileTest.h b/src/UnitTests/FileTest.h
index 4f15d6ac71..93616bb17a 100644
--- a/src/UnitTests/FileTest.h
+++ b/src/UnitTests/FileTest.h
@@ -50,6 +50,37 @@ TEST( FileTest, WriteAndRead )
EXPECT_EQ( std::remove( "test-file.tnl" ), 0 );
};
+TEST( FileTest, WriteAndReadWithConversion )
+{
+ double doubleData[ 3 ] = { 3.1415926535897932384626433,
+ 2.7182818284590452353602874,
+ 1.6180339887498948482045868 };
+ float floatData[ 3 ];
+ int intData[ 3 ];
+ File file;
+ file.open( "test-file.tnl", File::Mode::Out | File::Mode::Truncate );
+ file.write< double, Devices::Host, float >( doubleData, 3 );
+ file.close();
+
+ file.open( "test-file.tnl", File::Mode::In );
+ file.read< float, Devices::Host, float >( floatData, 3 );
+ file.close();
+
+ file.open( "test-file.tnl", File::Mode::In );
+ file.read< int, Devices::Host, float >( intData, 3 );
+ file.close();
+
+ EXPECT_NEAR( floatData[ 0 ], 3.14159, 0.0001 );
+ EXPECT_NEAR( floatData[ 1 ], 2.71828, 0.0001 );
+ EXPECT_NEAR( floatData[ 2 ], 1.61803, 0.0001 );
+
+ EXPECT_EQ( intData[ 0 ], 3 );
+ EXPECT_EQ( intData[ 1 ], 2 );
+ EXPECT_EQ( intData[ 2 ], 1 );
+
+ EXPECT_EQ( std::remove( "test-file.tnl" ), 0 );
+}
+
#ifdef HAVE_CUDA
TEST( FileTest, WriteAndReadCUDA )
{
@@ -123,6 +154,60 @@ TEST( FileTest, WriteAndReadCUDA )
EXPECT_EQ( std::remove( "test-file.tnl" ), 0 );
};
+
+TEST( FileTest, WriteAndReadCUDAWithConversion )
+{
+ const double constDoubleData[ 3 ] = { 3.1415926535897932384626433,
+ 2.7182818284590452353602874,
+ 1.6180339887498948482045868 };
+ float floatData[ 3 ];
+ int intData[ 3 ];
+
+ int* cudaIntData;
+ float* cudaFloatData;
+ const double* cudaConstDoubleData;
+ cudaMalloc( ( void** ) &cudaIntData, 3 * sizeof( int ) );
+ cudaMalloc( ( void** ) &cudaFloatData, 3 * sizeof( float ) );
+ cudaMalloc( ( void** ) &cudaConstDoubleData, 3 * sizeof( double ) );
+ cudaMemcpy( (void*) cudaConstDoubleData,
+ &constDoubleData,
+ 3 * sizeof( double ),
+ cudaMemcpyHostToDevice );
+
+ File file;
+ file.open( String( "cuda-test-file.tnl" ), File::Mode::Out | File::Mode::Truncate );
+ file.write< double, Devices::Cuda, float >( cudaConstDoubleData, 3 );
+ file.close();
+
+ file.open( String( "cuda-test-file.tnl" ), File::Mode::In );
+ file.read< float, Devices::Cuda, float >( cudaFloatData, 3 );
+ file.close();
+
+ file.open( String( "cuda-test-file.tnl" ), File::Mode::In );
+ file.read< int, Devices::Cuda, float >( cudaIntData, 3 );
+ file.close();
+
+ cudaMemcpy( floatData,
+ cudaFloatData,
+ 3 * sizeof( float ),
+ cudaMemcpyDeviceToHost );
+ cudaMemcpy( &intData,
+ cudaIntData,
+ 3* sizeof( int ),
+ cudaMemcpyDeviceToHost );
+
+
+ EXPECT_NEAR( floatData[ 0 ], 3.14159, 0.0001 );
+ EXPECT_NEAR( floatData[ 1 ], 2.71828, 0.0001 );
+ EXPECT_NEAR( floatData[ 2 ], 1.61803, 0.0001 );
+
+ EXPECT_EQ( intData[ 0 ], 3 );
+ EXPECT_EQ( intData[ 1 ], 2 );
+ EXPECT_EQ( intData[ 2 ], 1 );
+
+ EXPECT_EQ( std::remove( "cuda-test-file.tnl" ), 0 );
+};
+
#endif
#endif
--
GitLab