diff --git a/NeoMathEngine/CMakeLists.txt b/NeoMathEngine/CMakeLists.txt index 70b98ca8d..75e233446 100644 --- a/NeoMathEngine/CMakeLists.txt +++ b/NeoMathEngine/CMakeLists.txt @@ -47,4 +47,7 @@ add_subdirectory(src) if(NeoMathEngine_BUILD_TESTS AND NOT IOS AND NOT ANDROID) enable_testing() add_subdirectory(test/FullTestDesktop) + if(WIN32 AND NOT USE_FINE_OBJECTS AND CMAKE_SIZEOF_VOID_P EQUAL 8) + add_subdirectory(test/AvxTestDesktop) + endif() endif() diff --git a/NeoMathEngine/include/NeoMathEngine/SimdMathEngine.h b/NeoMathEngine/include/NeoMathEngine/SimdMathEngine.h index ce330305f..21e4a5731 100644 --- a/NeoMathEngine/include/NeoMathEngine/SimdMathEngine.h +++ b/NeoMathEngine/include/NeoMathEngine/SimdMathEngine.h @@ -41,6 +41,15 @@ class ISimdMathEngine : public CCrtAllocatedObject { virtual void BlobConvolution( const CConvolutionDesc& convDesc, const float* source, const float* filter, const float* freeTerm, float* result ) const = 0; + virtual CConvolutionDesc* InitBlockedConvolution( const CBlobDesc& source, int paddingHeight, int paddingWidth, + int strideHeight, int strideWidth, int dilationHeight, int dilationWidth, const CBlobDesc& filter, + const CBlobDesc& result ) const = 0; + virtual void PackBlockedData(const CBlobDesc& desc, const float* source, float* result) const = 0; + virtual void UnpackBlockedData( const CBlobDesc& desc, const float* source, float* result ) const = 0; + virtual void PackBlockedFilter( const CBlobDesc& desc, const float* source, float* result ) const = 0; + virtual void BlockedConvolution( const CConvolutionDesc& convDesc, const float* packedSource, + const float* packedFilter, const float* freeTerm, float* packedResult ) const = 0; + virtual SgemmFunc GetSgemmFunction() const = 0; virtual void Tanh( float* dst, const float* src, size_t dataSize, bool isMultithread = true ) = 0; diff --git a/NeoMathEngine/src/CPU/CpuMathEngineDnnConv.cpp b/NeoMathEngine/src/CPU/CpuMathEngineDnnConv.cpp index f134d24bb..6b7f42c60 100644 --- a/NeoMathEngine/src/CPU/CpuMathEngineDnnConv.cpp +++ b/NeoMathEngine/src/CPU/CpuMathEngineDnnConv.cpp @@ -45,13 +45,16 @@ struct CCpuConvolutionDesc : public CCommonConvolutionDesc { TConvAlgo ForwardAlgo; TConvAlgo BackwardAlgo; std::unique_ptr SimdConvolutionDesc; + std::unique_ptr BlockedConvolutionDesc; - CCpuConvolutionDesc( std::unique_ptr& simdConvolutionDesc, const CBlobDesc& source, const CBlobDesc& result, const CBlobDesc& filter, - int paddingHeight, int paddingWidth, int strideHeight, int strideWidth, int dilationHeight, int dilationWidth ) : + CCpuConvolutionDesc( std::unique_ptr& simdConvolutionDesc, std::unique_ptr& blockedConvolutionDesc, + const CBlobDesc& source, const CBlobDesc& result, const CBlobDesc& filter, int paddingHeight, int paddingWidth, + int strideHeight, int strideWidth, int dilationHeight, int dilationWidth ) : CCommonConvolutionDesc( source, result, filter, paddingHeight, paddingWidth, strideHeight, strideWidth, dilationHeight, dilationWidth ), ForwardAlgo( getActualForwardAlgo() ), BackwardAlgo( getActualBackwardAlgo() ), - SimdConvolutionDesc( std::move( simdConvolutionDesc ) ) + SimdConvolutionDesc( std::move( simdConvolutionDesc ) ), + BlockedConvolutionDesc( std::move( blockedConvolutionDesc ) ) { } @@ -131,14 +134,20 @@ CConvolutionDesc* CCpuMathEngine::InitBlobConvolution( const CBlobDesc& source, ASSERT_EXPR( result.Channels() == filter.BatchWidth() ); ASSERT_EXPR( result.Depth() == 1 ); - std::unique_ptr simdConvolutionDesc; + std::unique_ptr blockedConvolutionDesc; if( simdMathEngine != nullptr ) { + blockedConvolutionDesc.reset( simdMathEngine->InitBlockedConvolution( source, paddingHeight, paddingWidth, + strideHeight, strideWidth, dilationHeight, dilationWidth, filter, result ) ); + } + + std::unique_ptr simdConvolutionDesc; + if( simdMathEngine != nullptr && blockedConvolutionDesc == nullptr ) { simdConvolutionDesc = std::unique_ptr( simdMathEngine->InitBlobConvolution( source, paddingHeight, paddingWidth, strideHeight, strideWidth, dilationHeight, dilationWidth, filter, result ) ); } - CCpuConvolutionDesc* desc = new CCpuConvolutionDesc( simdConvolutionDesc, source, result, filter, - paddingHeight, paddingWidth, strideHeight, strideWidth, dilationHeight, dilationWidth ); + CCpuConvolutionDesc* desc = new CCpuConvolutionDesc( simdConvolutionDesc, blockedConvolutionDesc, source, result, + filter, paddingHeight, paddingWidth, strideHeight, strideWidth, dilationHeight, dilationWidth ); return desc; } @@ -517,6 +526,21 @@ void CCpuMathEngine::BlobConvolution( const CConvolutionDesc& convDesc, const CC const CCpuConvolutionDesc& desc = static_cast( convDesc ); + if( desc.BlockedConvolutionDesc != nullptr ) { + CFloatHandleStackVar packBuff( *this, + std::max( desc.Source.BlobSize(), desc.Result.BlobSize() ) + desc.Filter.BlobSize() ); + float* packedFilter = GetRaw( packBuff.GetHandle() ); + float* packedIO = packedFilter + desc.Filter.BlobSize(); + float* rawResult = GetRaw( result ); + simdMathEngine->PackBlockedData( desc.Source, GetRaw( source ), packedIO ); + simdMathEngine->PackBlockedFilter( desc.Filter, GetRaw( filter ), packedFilter ); + simdMathEngine->BlockedConvolution( *desc.BlockedConvolutionDesc, packedIO, packedFilter, + freeTerm != nullptr ? GetRaw( *freeTerm ) : nullptr, rawResult ); + simdMathEngine->UnpackBlockedData( desc.Result, rawResult, packedIO ); + dataCopy( rawResult, packedIO, desc.Result.BlobSize() ); + return; + } + if( desc.SimdConvolutionDesc != nullptr ) { simdMathEngine->BlobConvolution( *desc.SimdConvolutionDesc, sourceRaw, filterRaw, freeTermRaw, resultRaw ); return; diff --git a/NeoMathEngine/src/CPU/x86/avx/CMakeLists.txt b/NeoMathEngine/src/CPU/x86/avx/CMakeLists.txt index 9c69eb142..b8d2b7c94 100644 --- a/NeoMathEngine/src/CPU/x86/avx/CMakeLists.txt +++ b/NeoMathEngine/src/CPU/x86/avx/CMakeLists.txt @@ -25,8 +25,10 @@ target_sources(${PROJECT_NAME} ./src/BlobConvolution_jit_FltCnt_18.inl ./src/BlobConvolution_jit_FltCnt_24.inl ./src/BlobConvolution_jit_FltCnt_32.inl + ./src/BlobBlockedConvolution.cpp ./src/PrimitivesJit.h ./src/AvxCommon.h + ./src/AvxMathEngine.h ./src/JitCommon.h ./src/MatrixMultiplyingInterleaved/Interleavers/Interleavers.h ./src/MatrixMultiplyingInterleaved/MicroKernels/Kernel_AVX_6x16.h diff --git a/NeoMathEngine/src/CPU/x86/avx/src/AvxMathEngine.cpp b/NeoMathEngine/src/CPU/x86/avx/src/AvxMathEngine.cpp index 02b3f9d3a..a434f700a 100644 --- a/NeoMathEngine/src/CPU/x86/avx/src/AvxMathEngine.cpp +++ b/NeoMathEngine/src/CPU/x86/avx/src/AvxMathEngine.cpp @@ -19,6 +19,7 @@ limitations under the License. #include #include #include +#include #include namespace NeoML { @@ -48,32 +49,6 @@ CAvxConvolutionDesc::CAvxConvolutionDesc( IMathEngine* mathEngine, const CBlobDe { } -class CAvxMathEngine : public ISimdMathEngine { -public: - CAvxMathEngine( IMathEngine* _mathEngine, int _threadCount ) : - mathEngine( _mathEngine ), threadCount( _threadCount ), primitives( _mathEngine, _threadCount ) {} - - CConvolutionDesc* InitBlobConvolution( const CBlobDesc& source, int paddingHeight, int paddingWidth, - int strideHeight, int strideWidth, int dilationHeight, int dilationWidth, const CBlobDesc& filter, - const CBlobDesc& result ) const override; - - void BlobConvolution( const CConvolutionDesc& convDesc, const float* source, - const float* filter, const float* freeTerm, float* result ) const override; - - SgemmFunc GetSgemmFunction() const override; - - void Tanh( float* dst, const float* src, size_t dataSize, bool isMultithread ) override; - void Sigmoid( float* dst, const float* src, size_t dataSize, bool isMultithread ) override; - void Exp( float* dst, const float* src, size_t dataSize, bool isMultithread ) override; - void RunOnceRestOfLstm( CMathEngineLstmDesc* desc, const CConstFloatHandle& inputStateBackLink, - const CFloatHandle& outputStateBackLink, const CFloatHandle& outputMainBackLink, bool isMultithread ) override; - -private: - IMathEngine* mathEngine; - int threadCount; - CPrimitivesJit primitives; -}; - CConvolutionDesc* CAvxMathEngine::InitBlobConvolution( const CBlobDesc& source, int paddingHeight, int paddingWidth, int strideHeight, int strideWidth, int dilationHeight, int dilationWidth, const CBlobDesc& filter, const CBlobDesc& result ) const diff --git a/NeoMathEngine/src/CPU/x86/avx/src/AvxMathEngine.h b/NeoMathEngine/src/CPU/x86/avx/src/AvxMathEngine.h new file mode 100644 index 000000000..ee93542f9 --- /dev/null +++ b/NeoMathEngine/src/CPU/x86/avx/src/AvxMathEngine.h @@ -0,0 +1,59 @@ +/* Copyright © 2017-2022 ABBYY Production LLC + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. +--------------------------------------------------------------------------------------------------------------*/ + +#pragma once + +#include +#include + +namespace NeoML { + +class CAvxMathEngine : public ISimdMathEngine { +public: + CAvxMathEngine( IMathEngine* _mathEngine, int _threadCount ) : + mathEngine( _mathEngine ), threadCount( _threadCount ), primitives( _mathEngine, _threadCount ) {} + + CConvolutionDesc* InitBlobConvolution( const CBlobDesc& source, int paddingHeight, int paddingWidth, + int strideHeight, int strideWidth, int dilationHeight, int dilationWidth, const CBlobDesc& filter, + const CBlobDesc& result ) const override; + + void BlobConvolution( const CConvolutionDesc& convDesc, const float* source, + const float* filter, const float* freeTerm, float* result ) const override; + + virtual CConvolutionDesc* InitBlockedConvolution( const CBlobDesc& source, int paddingHeight, int paddingWidth, + int strideHeight, int strideWidth, int dilationHeight, int dilationWidth, const CBlobDesc& filter, + const CBlobDesc& result ) const override; + void PackBlockedData( const CBlobDesc& desc, const float* source, float* result ) const override; + void UnpackBlockedData( const CBlobDesc& desc, const float* source, float* result ) const override; + void PackBlockedFilter( const CBlobDesc& desc, const float* source, float* result ) const override; + void BlockedConvolution( const CConvolutionDesc& convDesc, const float* packedSource, + const float* packedFilter, const float* freeTerm, float* packedResult ) const override; + + SgemmFunc GetSgemmFunction() const override; + + void Tanh( float* dst, const float* src, size_t dataSize, bool isMultithread ) override; + void Sigmoid( float* dst, const float* src, size_t dataSize, bool isMultithread ) override; + void Exp( float* dst, const float* src, size_t dataSize, bool isMultithread ) override; + void RunOnceRestOfLstm( CMathEngineLstmDesc* desc, const CConstFloatHandle& inputStateBackLink, + const CFloatHandle& outputStateBackLink, const CFloatHandle& outputMainBackLink, bool isMultithread ) override; + +private: + IMathEngine* mathEngine; + int threadCount; + CPrimitivesJit primitives; +}; + +} + diff --git a/NeoMathEngine/src/CPU/x86/avx/src/BlobBlockedConvolution.cpp b/NeoMathEngine/src/CPU/x86/avx/src/BlobBlockedConvolution.cpp new file mode 100644 index 000000000..196ebf8d9 --- /dev/null +++ b/NeoMathEngine/src/CPU/x86/avx/src/BlobBlockedConvolution.cpp @@ -0,0 +1,842 @@ +/* Copyright © 2017-2020 ABBYY Production LLC + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. +--------------------------------------------------------------------------------------------------------------*/ + +#include +#pragma hdrstop + +#include +#include + +#include +#include +#include +#include + +namespace NeoML { + +// Blocked convolution is a special algo when input and output channels are divided into blocks of size, used in SIMD +// This algo requires data to be packed from +// N H W C_orig -> N C H W c +// where c == BlockSize and C == C_Orig / BlockSize +// This applies to both input and output +// As a result this algo requires both C_Input and C_Output to be a multiple of BlockSize + +// AVX works with 256-bit registers (8 floats) +static const int BlockSize = 8; + +// Descriptor for blocked convolution +struct CAvxBlockedConvDesc : public CConvolutionDesc { + CAvxBlockedConvDesc( const CBlobDesc& source, const CBlobDesc& result, const CBlobDesc& filter, int strideHeight, + int strideWidth, int paddingHeight, int paddingWidth, int dilationHeight, int dilationWidth ) : + Source( source ), + Result( result ), + Filter( filter ), + StrideHeight( strideHeight ), + StrideWidth( strideWidth ), + PaddingHeight( paddingHeight ), + PaddingWidth( paddingWidth ), + DilationHeight( dilationHeight ), + DilationWidth( dilationWidth ) + { + } + + + ~CAvxBlockedConvDesc() override = default; + + const CBlobDesc Source; // Input blob size + const CBlobDesc Result; // Output blob size + const CBlobDesc Filter; // Filter blob size + const int StrideHeight; // Step along height + const int StrideWidth; // Step along width + const int PaddingHeight; // Padding rows + const int PaddingWidth; // Paddinc columns + const int DilationHeight; // Dilation along height + const int DilationWidth; // Dilation along width +}; + +CConvolutionDesc* CAvxMathEngine::InitBlockedConvolution( const CBlobDesc& source, int paddingHeight, int paddingWidth, + int strideHeight, int strideWidth, int dilationHeight, int dilationWidth, const CBlobDesc& filter, + const CBlobDesc& result ) const +{ + const int filterCount = filter.ObjectCount(); + const int inputChannels = source.Depth() * source.Channels(); + if( filterCount % BlockSize != 0 || inputChannels % BlockSize != 0 ) { + // Algorithmic restrictions + return nullptr; + } + + if( CCPUInfo::IsAvx512Available() ) { + // This algo isn't tuned for AVX512 (yet) + return nullptr; + } + + if( filter.Height() == 1 && filter.Width() == 1 ) { + // We can't outperform matrix multiplication + return nullptr; + } + + // Heuristics tuned for better effectiveness + if( inputChannels < 32 || filterCount < 32 ) { + // All the packing doesn't improve performance when there's no enough channels to pack + return nullptr; + } + + if( dilationWidth > 1 ) { + // The cycle in JIT part is very sensitive to filters dilated over width + return nullptr; + } + + // Number of operations in convolution == outputBlobsSize * filterHeight * filterWidth * inputChannels + // Packing each data (input/output/filter) is linear + + // Ratio between number of operations in convolution and different packing operations + const size_t inputRatio = static_cast( result.ObjectSize() ) * filter.Height() * filter.Width() + / source.Height() / source.Width(); + const size_t outputRatio = static_cast( filter.Height() ) * filter.Width() * source.Depth() * source.Channels(); + const size_t filterRatio = static_cast( result.ObjectCount() * result.Height() * result.Width() ); + // If any of ratio is less than this value then packing takes too much time... + const size_t minRatio = 200; + if( inputRatio < minRatio || outputRatio < minRatio || filterRatio < minRatio ) { + return nullptr; + } + + // If both input and output ratios are slightly above min value the algo is still slow + const size_t minIOMult = 65536; + if( inputRatio * outputRatio < minIOMult ) { + return nullptr; + } + + return new CAvxBlockedConvDesc( source, result, filter, strideHeight, strideWidth, paddingHeight, paddingWidth, + dilationHeight, dilationWidth ); +} + +// Packs NHWC data from source into NCHWc result +// The desc must contain data dims in NHWC format +void CAvxMathEngine::PackBlockedData( const CBlobDesc& desc, const float* source, float* result ) const +{ + const int channels = desc.Depth() * desc.Channels(); + assert( channels % BlockSize == 0 ); + + const int height = desc.Height(); + const int width = desc.Width(); + const int geomSize = height * width; + const int batch = desc.ObjectCount(); + + for( int b = 0; b < batch; ++b ) { + for( int hw = 0; hw < geomSize; ++hw ) { + float* channelResult = result + hw * BlockSize; + for( int C = 0; C < channels / BlockSize; ++C ) { + _mm256_storeu_ps( channelResult, _mm256_loadu_ps( source ) ); + source += BlockSize; + channelResult += geomSize * BlockSize; + } + } + result += channels * geomSize; + } + _mm256_zeroupper(); +} + +// Unpacks NCHWc data from source into NHWC result +// The desc must contain data dims in NHWC format +void CAvxMathEngine::UnpackBlockedData( const CBlobDesc& desc, const float* source, float* result ) const +{ + const int channels = desc.Depth() * desc.Channels(); + assert( channels % BlockSize == 0 ); + + const int height = desc.Height(); + const int width = desc.Width(); + const int geomSize = height * width; + const int batch = desc.ObjectCount(); + + for( int b = 0; b < batch; ++b ) { + for( int C = 0; C < channels / BlockSize; ++C ) { + float* geomResult = result + C * BlockSize; + for( int hw = 0; hw < geomSize; ++hw ) { + _mm256_storeu_ps( geomResult, _mm256_loadu_ps( source ) ); + source += BlockSize; + geomResult += channels; + } + } + result += channels * geomSize; + } + _mm256_zeroupper(); +} + +// This algo requires special pack for filters from O_Orig x H x W x I_Orig into O x I x H x W x i x o +// where O_Orig and I_Orig are the output and input channels in convolution +// O = O_Orig / BlockSize, I = I_Orig / BlockSize, and i = o = BlockSize +// The desc must contain filter data dims in OHWI format +void CAvxMathEngine::PackBlockedFilter( const CBlobDesc& desc, const float* source, float* result ) const +{ + const int filterCount = desc.ObjectCount(); + const int channels = desc.Depth() * desc.Channels(); + const int height = desc.Height(); + const int width = desc.Width(); + assert( filterCount % BlockSize == 0 ); + assert( channels % BlockSize == 0 ); + + for( int O = 0; O < filterCount / BlockSize; ++O ) { + for( int o = 0; o < BlockSize; ++o ) { + for( int h = 0; h < height; ++h ) { + for( int w = 0; w < width; ++w ) { + for( int I = 0; I < channels / BlockSize; ++I ) { +#ifdef _MSC_VER + __m256 simdSource = _mm256_loadu_ps( source ); + for( int i = 0; i < BlockSize; ++i ) { + const int idx = o + BlockSize * ( i + BlockSize * ( w + width * ( h + height * ( I + channels / BlockSize * O ) ) ) ); + result[idx] = simdSource.m256_f32[i]; + } +#else + for( int i = 0; i < BlockSize; ++i ) { + const int idx = o + BlockSize * ( i + BlockSize * ( w + width * ( h + height * ( I + channels / BlockSize * O ) ) ) ); + result[idx] = source[i]; + } +#endif + source += BlockSize; + } + } + } + } + } +} + +// -------------------------------------------------------------------------------------------------------------------- + +// Main part with convolution itself + +using namespace Xbyak::util; +using namespace Xbyak; + +// Algorithm splits filter blocks (groups of 8 filters) into sets +// One filter sets contains up to 4 filter blocks +// This restriction is caused by the number of Ymm register +static const int MaxFilterSetSize = 4; + +// One jit call calculates convolution of one row of output over BlockSize of input channels +// for one filter set (up to MaxFilterSetSize * BlockSize filters) + +// The idea is to call it over all blocks of input channels for each filter set +// for all of the output rows + +class CBlockedConvGen : public Xbyak::CodeGenerator { +public: + // Flags which can be added to Flags parameter in order to + // control convolution behavior + + // Add values from memory to the result + // Used when memory contains previous calculations + // (e.g. processied input channel block isn't first) + static const size_t AccumulateOutputFlag = 1; + // Add values from free terms to result + // Used when processing last input channel block + static const size_t AddFreeTermFlag = 2; + + // JIT call parameters and their order + struct CParams { + // Current position in input + // May be outside of physical data (when pointing at padding) + const float* Input; + // First position in input which point to physical data + // Used for padding detection + const float* InputAfterLeftPad; + // Distance between elements in the same column of input image + size_t InputWidthBytes; + // Distance to the input element covered by the next filter column + size_t DilationWidthBytes; + // Distance to the input element covered by the next filter row + size_t DilatedInputWidthBytes; + // Distance to the next input element covered by the same filter + size_t StrideWidthBytes; + // Distance between the input element covered by the last element of this filter row + // and the first element covered by the next filter row + size_t InputStrideBytes; + // Filter pointer + const float* Filter; + // Number of blocks in the current set + size_t FilterSetSize; + // Number of filter rows to be processed (!rows in padding are excluded!) + size_t FilterHeight; + // Number of filter columns to be processed + size_t FilterWidth; + // Distance to the next filter block + size_t FilterStrideBytes; + // Free term pointer + const float* FreeTerm; + // Output pointer + float* Output; + // Distance between different channel blocks belonging to the same "original" output channel + size_t OutputStrideBytes; + // Number of output elements in row affected by left padding + size_t OutputColCountLeftPad; + // Number of output elements not affected by any padding + size_t OutputColCountNoPad; + // Number of output elements in row affected by right padding + size_t OutputColCountRightPad; + // Special flags for the correct processing of first/last input channel block + size_t Flags; + // Special field used to store stack pointer during JIT call + float* PrevRsp; + }; + + CBlockedConvGen(); + + // Runs generated code with the given parameters + void Run( CParams& params ); + +private: + // Prologue + void genPrologue(); + + void genComputeOutputColumns( int filterSetSize, int outputColCount ); + void genComputeOutputColumnsPrologue( int outputColCount ); + void genClearAccumulators( int filterSetSize, int outputColCount ); + void genFilterGeometryLoops( int filterSetSize, int outputColCount ); + void genProcessSingleInputChannel( int filterSetSize, int outputColCount, int inBlockPos ); + void genComputeOutputColumnsEpilogue( int filterSetSize, int outputColCount ); + void genPaddingProcessing( int filterSetSize ); + void genProcessFilterSet( int filterSetSize ); + void genConvKernel(); + + // Epilogue + void genEpilogue(); + + const reg64_t regInput = rdi; + const reg64_t regRemOutputColCount = r10; + + const reg64_t regBlockInput = rcx; + const reg64_t regStrideWidth = r9; + const reg64_t regFilter = rdx; + const reg64_t regFilterStride = rsi; + const reg64_t regDilationWidth = rbp; + const reg64_t regInputStride = r15; + const reg64_t regNegInputAfterLeftPad = r13; + const reg64_t regInputWidth = r14; + const reg64_t regOutput = r8; + const reg64_t regFilterHeight = r11; + const reg64_t regFilterWidth = r12; + const reg64_t regShiftedInput = r14; + const reg64_t regShiftedFilter = rbx; + + const Ymm acc[3][MaxFilterSetSize] = { + { Ymm( 0 ), Ymm( 1 ), Ymm( 2 ), Ymm( 3 ) }, + { Ymm( 4 ), Ymm( 5 ), Ymm( 6 ), Ymm( 7 ) }, + { Ymm( 8 ), Ymm( 9 ), Ymm( 10 ), Ymm( 11 ) } + }; +}; + +CBlockedConvGen::CBlockedConvGen() : CodeGenerator( 16 * 1024 ) +{ + genPrologue(); + genConvKernel(); + genEpilogue(); +} + +void CBlockedConvGen::Run( CParams& params ) +{ + typedef void (*TComputeBlockJitFunc)( CParams* ); + getCode()( ¶ms ); +} + +void CBlockedConvGen::genPrologue() +{ + push( rbp ); + mov( rbp, rsp ); + + // Store all the YMM registers + sub( rsp, static_cast( 16 * SizeOfYmm ) ); + for( int i = 0; i < 16; i++ ) { + vmovdqu( ptr[rsp + i * SizeOfYmm], Ymm( i ) ); + } + + // Store all the general purpose registers + reg64Vec_t gprs = { rax, rbx, rbp, rcx, rdx, rsi, rdi, r8, r9, r10, r11, r12, r13, r14, r15 }; + for( int i = 0; i < gprs.size(); i++ ) { + push( gprs[i] ); + } + + // At this moment Param1 register contains a pointer to the CBlockedConvGen::CParams + // The idea is to + // 1. Write current stack pointer (rsp) to a special field (CParams::PrevRsp) + // 2. Move stack pointer to the beginning of the CParams + // 3. Exract params via rsp + offsetof(CParams, ParamName) + // 4. Later (in epilogue) restore original rsp from this special field (CParams::PrevRsp) + mov( ptr[Param1 + offsetof( CParams, PrevRsp )], rsp ); + mov( rsp, Param1 ); + + mov( regInput, ptr[rsp + offsetof( CParams, Input )] ); + mov( regFilterStride, ptr[rsp + offsetof( CParams, FilterStrideBytes )] ); + mov( regDilationWidth, ptr[rsp + offsetof( CParams, DilationWidthBytes )] ); + mov( regOutput, ptr[rsp + offsetof( CParams, Output )] ); + mov( regStrideWidth, ptr[rsp + offsetof( CParams, StrideWidthBytes )] ); + mov( regInputStride, ptr[rsp + offsetof( CParams, InputStrideBytes )] ); +} + +// Generates code which computes outputColCount neighboring output columns +// for the given filter set over current input channels block +void CBlockedConvGen::genComputeOutputColumns( int filterSetSize, int outputColCount ) +{ + genComputeOutputColumnsPrologue( outputColCount ); + genClearAccumulators( filterSetSize, outputColCount ); + genFilterGeometryLoops( filterSetSize, outputColCount ); + genComputeOutputColumnsEpilogue( filterSetSize, outputColCount ); +} + +// Generates code which fills registers used during computing output columns +void CBlockedConvGen::genComputeOutputColumnsPrologue( int outputColCount ) +{ + mov( regBlockInput, regInput ); + mov( regFilter, ptr[rsp + offsetof( CParams, Filter )] ); + mov( regFilterHeight, ptr[rsp + offsetof( CParams, FilterHeight )] ); + mov( regFilterWidth, ptr[rsp + offsetof( CParams, FilterWidth )] ); + if( outputColCount == 1 ) { + mov( regNegInputAfterLeftPad, ptr[rsp + offsetof( CParams, InputAfterLeftPad )] ); + neg( regNegInputAfterLeftPad ); + mov( regInputWidth, ptr[rsp + offsetof( CParams, InputWidthBytes )] ); + } +} + + // Generates code which zeroes ymm accumulators used during computing output columns +void CBlockedConvGen::genClearAccumulators( int filterSetSize, int outputColCount ) +{ + for( int filterBlockIdx = 0; filterBlockIdx < filterSetSize; ++filterBlockIdx ) { + for( int outputColIdx = 0; outputColIdx < outputColCount; ++outputColIdx ) { + vxorps( acc[outputColIdx][filterBlockIdx], acc[outputColIdx][filterBlockIdx] ); + } + } +} + +// Generates loops processing whole filter geometry (filterHeight x filterWidth) in current call +void CBlockedConvGen::genFilterGeometryLoops( int filterSetSize, int outputColCount ) +{ + // Cycle over filter rows + // for( regFilterHeight = filterHeight; regFilterHeight > 0; --regFilterHeight ) + mov( regFilterHeight, ptr[rsp + offsetof( CParams, FilterHeight )] ); + + // Corner case: whole filter is in padding + Label skilpWholeFilterInPadding; + test( regFilterHeight, regFilterHeight ); + jz( skilpWholeFilterInPadding, T_NEAR ); + + Label rowCycleStart; + L( rowCycleStart ); + + // Inner cycle over filter columns + // for( regRemCols = filterWidth; regRemCols > 0; --regRemCols ) + const reg64_t regRemCols = rax; + mov( regRemCols, regFilterWidth ); + + Label colCycleStart; + L( colCycleStart ); + + // The padding check is performed only while process 1 output column at a time + Label skipPadding; + if( outputColCount == 1 ) { + // regNegInputAfterLeftPad contains negative value of the pointer to first element after padding + // regBlockInput contains current pointer + // Due to the fact that used types are unsigned we can check both left and right padding by + // checking the value of (currInput - inputAfterLeftPad) + // In case of right padding the value will be >= inputWidth + // In case of left padding the value will be very large (overflow of unsigned type) + lea( rbx, ptr[regBlockInput + regNegInputAfterLeftPad] ); + cmp( rbx, regInputWidth ); + jae( skipPadding, CodeGenerator::T_NEAR ); + } + + if( outputColCount >= 3 ) { + lea( regShiftedInput, ptr[regBlockInput + 2 * regStrideWidth] ); + } + + if( filterSetSize >= 3 ) { + lea( regShiftedFilter, ptr[regFilter + 2 * regFilterStride] ); + } + + // (Reminder: this code is inside cycles over filterHeight and filterWidth) + // Generate code for processing each input channel in this block + for( int inBlockPos = 0; inBlockPos < BlockSize; ++inBlockPos ) { + genProcessSingleInputChannel( filterSetSize, outputColCount, inBlockPos ); + } + + L( skipPadding ); + add( regBlockInput, regDilationWidth ); + // Move filter pointer to the next columns + // Due to the OIHWio packing it moves the filter pointer to the beginning of the next row + // when current filter column is the last one + add( regFilter, BlockSize * BlockSize * sizeof( float ) ); + dec( regRemCols ); + jnz( colCycleStart, CodeGenerator::T_NEAR ); + // Cycle over filter columns ends + + add( regBlockInput, regInputStride ); + if( outputColCount == 1 ) { + // Move pointer used for padding detection to the next input line covered by this filter set + sub( regNegInputAfterLeftPad, ptr[rsp + offsetof( CParams, DilatedInputWidthBytes )] ); + } + + dec( regFilterHeight ); + jnz( rowCycleStart, CodeGenerator::T_NEAR ); + + L( skilpWholeFilterInPadding ); +} + +// Generates code which mutilplies one column of one filter set by outputColCount input elements +// inBlockPos is an index of the used input channel inside of a block +// +// It takes outputColCount input elements belonging to the inBlockPos'th input channel inside the block of input channels +// and affected by the same filter elements when filter moves along image width +// The results of applying current filter set to this channel are added to neighboring output columns +void CBlockedConvGen::genProcessSingleInputChannel( int filterSetSize, int outputColCount, int inBlockPos ) +{ + const int inBlockInputOffset = inBlockPos * sizeof( float ); + const int inBlockFilerOffset = inBlockPos * BlockSize * sizeof( float ); + + Ymm inputYmm[3] = { Ymm( 13 ), Ymm( 14 ), Ymm( 15 ) }; + // Calculate address of input elements for current filter position, + // and for (outputColCount - 1) neighboring filter position + Address inputAddr[3] = { ptr[regBlockInput + inBlockInputOffset], + ptr[regBlockInput + regStrideWidth + inBlockInputOffset], ptr[regShiftedInput + inBlockInputOffset] }; + + // Fill input ymms with the values from input from the current channel + // corresponding to given output columns + for( int outputColIdx = 0; outputColIdx < outputColCount; ++outputColIdx ) { + vbroadcastss( inputYmm[outputColIdx], inputAddr[outputColIdx] ); + } + + Address filterAddr[MaxFilterSetSize] = { ptr[regFilter + inBlockFilerOffset], + ptr[regFilter + regFilterStride + inBlockFilerOffset], ptr[regShiftedFilter + inBlockFilerOffset], + ptr[regShiftedFilter + regFilterStride + inBlockFilerOffset] }; + + if( outputColCount == 1 ) { + for( int filterBlockIdx = 0; filterBlockIdx < filterSetSize; ++filterBlockIdx ) { + vfmadd231ps( acc[0][filterBlockIdx], inputYmm[0], filterAddr[filterBlockIdx] ); + } + } else { + for( int filterBlockIdx = 0; filterBlockIdx < filterSetSize; ++filterBlockIdx ) { + Ymm filterYmm( 12 ); + vmovups( filterYmm, filterAddr[filterBlockIdx] ); + for( int outputColIdx = 0; outputColIdx < outputColCount; ++outputColIdx ) { + vfmadd231ps( acc[outputColIdx][filterBlockIdx], inputYmm[outputColIdx], filterYmm); + } + } + } +} + +void CBlockedConvGen::genComputeOutputColumnsEpilogue( int filterSetSize, int outputColCount ) +{ + const reg64_t regOutputStride = rax; + mov( regOutputStride, ptr[rsp + offsetof( CParams, OutputStrideBytes )] ); + const reg64_t regShiftedOutput = rbx; + if( filterSetSize >= 3 ) { + lea( regShiftedOutput, ptr[regOutput + 2 * regOutputStride] ); + } + const reg64_t regFlags = rdx; + mov( regFlags, ptr[rsp + offsetof( CParams, Flags )] ); + + Label skipAccumulateOutput; + test( regFlags, AccumulateOutputFlag ); + jz( skipAccumulateOutput, CodeGenerator::T_NEAR ); + + RegExp outputRegExp[4] = { regOutput, regOutput + regOutputStride, + regShiftedOutput, regShiftedOutput + regOutputStride }; + for( int filterBlockIdx = 0; filterBlockIdx < filterSetSize; ++filterBlockIdx ) { + for( int outputColIdx = 0; outputColIdx < outputColCount; ++outputColIdx ) { + vaddps( acc[outputColIdx][filterBlockIdx], acc[outputColIdx][filterBlockIdx], + ptr[outputRegExp[filterBlockIdx] + outputColIdx * SizeOfYmm] ); + } + } + + L( skipAccumulateOutput ); + + Label skipFreeTerm; + test( regFlags, AddFreeTermFlag ); + jz( skipFreeTerm, CodeGenerator::T_NEAR ); + + const reg64_t regFreeTerm = rcx; + mov( regFreeTerm, ptr[rsp + offsetof( CParams, FreeTerm )] ); + + if( outputColCount == 1 ) { + for( int filterBlockIdx = 0; filterBlockIdx < filterSetSize; ++filterBlockIdx ) { + vaddps( acc[0][filterBlockIdx], acc[0][filterBlockIdx], ptr[regFreeTerm + filterBlockIdx * SizeOfYmm]); + } + } else { + const Ymm freeTermYmm[4] = { Ymm( 12 ), Ymm( 13 ), Ymm( 14 ), Ymm( 15 ) }; + for( int filterBlockIdx = 0; filterBlockIdx < filterSetSize; ++filterBlockIdx ) { + vmovups( freeTermYmm[filterBlockIdx], ptr[regFreeTerm + filterBlockIdx * SizeOfYmm]); + } + for( int outputColIdx = 0; outputColIdx < outputColCount; ++outputColIdx ) { + for( int filterBlockIdx = 0; filterBlockIdx < filterSetSize; ++filterBlockIdx ) { + vaddps( acc[outputColIdx][filterBlockIdx], acc[outputColIdx][filterBlockIdx], freeTermYmm[filterBlockIdx] ); + } + } + } + + L( skipFreeTerm ); + + for( int filterBlockIdx = 0; filterBlockIdx < filterSetSize; ++filterBlockIdx ) { + for( int outputColIdx = 0; outputColIdx < outputColCount; ++outputColIdx ) { + vmovups( ptr[outputRegExp[filterBlockIdx] + outputColIdx * SizeOfYmm], + acc[outputColIdx][filterBlockIdx] ); + } + } + + add( regOutput, outputColCount * BlockSize * sizeof( float ) ); +} + +// Generates code which processes output columns affected by padding +void CBlockedConvGen::genPaddingProcessing( int filterSetSize ) +{ + Label processNextOutput; + + L( processNextOutput ); + // Current algo generates padding processing only when 1 output column is processed + genComputeOutputColumns( filterSetSize, 1 ); + add( regInput, regStrideWidth ); + dec( regRemOutputColCount ); + jnz( processNextOutput, T_NEAR ); +} + +// Generate processing of filter set of given size +void CBlockedConvGen::genProcessFilterSet( int filterSetSize ) +{ + assert( filterSetSize >= 1 && filterSetSize <= MaxFilterSetSize ); + Label skipLeftPad; + mov( regRemOutputColCount, ptr[rsp + offsetof( CParams, OutputColCountLeftPad )] ); + // Process left padding (if needed) + test( regRemOutputColCount, regRemOutputColCount ); + jz( skipLeftPad, T_NEAR ); + genPaddingProcessing( filterSetSize ); + L( skipLeftPad ); + + Label processThreeOutputs; + Label processRemainingOutputs; + Label processRemainingWithRightPad; + + mov( regRemOutputColCount, ptr[rsp + offsetof( CParams, OutputColCountNoPad )] ); + // Initially subtract 3 for jumping via comparison with 0 (jz, jb, jae, etc.) + sub( regRemOutputColCount, 3 ); + jb( processRemainingOutputs, T_NEAR ); + + L( processThreeOutputs ); + genComputeOutputColumns( filterSetSize, 3 ); + const reg64_t regTemp = rax; + lea( regTemp, ptr[regStrideWidth + 2 * regStrideWidth] ); + add( regInput, regTemp ); + sub( regRemOutputColCount, 3 ); + jae( processThreeOutputs, T_NEAR ); + + L( processRemainingOutputs ); + // Compensate subtraction above + add( regRemOutputColCount, 3 ); + jz( processRemainingWithRightPad, T_NEAR ); + cmp( regRemOutputColCount, 2 ); + jb( processRemainingWithRightPad, T_NEAR ); + genComputeOutputColumns( filterSetSize, 2 ); + lea( regInput, ptr[regInput + 2 * regStrideWidth] ); + sub( regRemOutputColCount, 2 ); + + // Process irhgt padding (if needed) + Label skipRightPad; + L( processRemainingWithRightPad ); + add( regRemOutputColCount, ptr[rsp + offsetof( CParams, OutputColCountRightPad )] ); + jz( skipRightPad, T_NEAR ); + genPaddingProcessing( filterSetSize ); + L( skipRightPad ); +} + +void CBlockedConvGen::genConvKernel() +{ + const reg64_t regFilterSetSize = r11; + mov( regFilterSetSize, ptr[rsp + offsetof( CParams, FilterSetSize )] ); + + Label processThreeFilters; + Label processLessThanThreeFilters; + Label processOneFilter; + Label convKernelEnd; + + cmp( regFilterSetSize, 3 ); + je( processThreeFilters, T_NEAR ); + jb( processLessThanThreeFilters, T_NEAR ); + genProcessFilterSet( 4 ); + jmp( convKernelEnd, T_NEAR ); + + L( processThreeFilters ); + genProcessFilterSet( 3 ); + jmp( convKernelEnd, T_NEAR ); + + L( processLessThanThreeFilters ); + cmp( regFilterSetSize, 2 ); + jb( processOneFilter, T_NEAR ); + genProcessFilterSet( 2 ); + jmp( convKernelEnd, T_NEAR ); + + L( processOneFilter ); + genProcessFilterSet( 1 ); + L( convKernelEnd ); +} + +void CBlockedConvGen::genEpilogue() +{ + // Restore stack pointer from the special field + mov( rsp, ptr[rsp + offsetof( CParams, PrevRsp )] ); + + reg64Vec_t gprs = { rax, rbx, rbp, rcx, rdx, rsi, rdi, r8, r9, r10, r11, r12, r13, r14, r15 }; + for( int i = static_cast( gprs.size() - 1 ); i >= 0; i-- ) { + pop( gprs[i] ); + } + + for( int i = 0; i < 16; i++ ) { + vmovdqu( Ymm( i ), ptr[rsp + static_cast( i * SizeOfYmm )]); + } + + leave(); + ret(); +} + +static CBlockedConvGen blockedConvGen; + +// Calculates the number of output elements affected by front or back padding +static void calcOutputPad( int inputSize, int filterSize, int outputSize, int stride, int padding, int dilation, + size_t& outputFrontPad, size_t& outputNoPad, size_t& outputBackPad ) +{ + const int filterCoverage = ( filterSize - 1 ) * dilation + 1; + const int unaffectedByBackPad = ( inputSize + padding >= filterCoverage ) + ? ( inputSize + padding - filterCoverage ) / stride + 1 : 0; + outputFrontPad = static_cast( min( unaffectedByBackPad, ( padding + stride - 1 ) / stride ) ); + outputNoPad = static_cast( unaffectedByBackPad ) - outputFrontPad; + outputBackPad = static_cast( outputSize - unaffectedByBackPad ); +} + +void CAvxMathEngine::BlockedConvolution( const CConvolutionDesc& convDesc, const float* packedSource, + const float* packedFilter, const float* freeTerm, float* packedResult ) const +{ + const CAvxBlockedConvDesc& desc = static_cast( convDesc ); + + const int inputHeight = desc.Source.Height(); + const int inputWidth = desc.Source.Width(); + const int inputChannels = desc.Source.Depth() * desc.Source.Channels(); + const int outputHeight = desc.Result.Height(); + const int outputWidth = desc.Result.Width(); + const int filterCount = desc.Filter.ObjectCount(); + const int filterHeight = desc.Filter.Height(); + const int filterWidth = desc.Filter.Width(); + const int dilationHeight = desc.DilationHeight; + + size_t outputColLeftPad = 0, outputColNoPad = 0, outputColRightPad = 0; + calcOutputPad( inputWidth, filterWidth, outputWidth, desc.StrideWidth, desc.PaddingWidth, desc.DilationWidth, + outputColLeftPad, outputColNoPad, outputColRightPad ); + + const int filterNumInSet = MaxFilterSetSize * BlockSize; + const int filterSetCount = ( filterCount + filterNumInSet - 1 ) / filterNumInSet; + const int totalWork = desc.Source.ObjectCount() * filterSetCount * outputHeight; + const int curThreadCount = IsOmpRelevant( totalWork, + static_cast( desc.Result.BlobSize() ) * desc.Filter.ObjectSize() ) ? threadCount : 1; + + NEOML_OMP_NUM_THREADS( curThreadCount ) + { + int workIndex, workRem; + if( OmpGetTaskIndexAndCount( totalWork, workIndex, workRem ) ) { + int outputRowIndex = workIndex % outputHeight; + int filterSetIndex = ( workIndex / outputHeight ) % filterSetCount; + const int batchIndex = ( workIndex / outputHeight ) / filterSetCount; + + const float* inputImage = packedSource + batchIndex * desc.Source.ObjectSize(); + const float* filter = packedFilter + filterSetIndex * filterNumInSet * desc.Filter.ObjectSize(); + float* output = packedResult + batchIndex * desc.Result.ObjectSize() + + filterSetIndex * filterNumInSet * outputHeight * outputWidth; + + CBlockedConvGen::CParams callParams; + + callParams.FreeTerm = freeTerm != nullptr ? freeTerm + filterSetIndex * filterNumInSet : freeTerm; + callParams.StrideWidthBytes = sizeof( float ) * BlockSize * desc.StrideWidth; + callParams.DilationWidthBytes = sizeof( float ) * BlockSize * desc.DilationWidth; + callParams.FilterStrideBytes = sizeof( float ) * BlockSize * desc.Filter.ObjectSize(); + callParams.OutputStrideBytes = sizeof( float ) * BlockSize * outputHeight * outputWidth; + callParams.InputWidthBytes = sizeof( float ) * BlockSize * inputWidth; + callParams.DilatedInputWidthBytes = sizeof( float ) * BlockSize * dilationHeight * inputWidth; + callParams.InputStrideBytes = callParams.DilatedInputWidthBytes - filterWidth * callParams.DilationWidthBytes; + callParams.FilterWidth = filterWidth; + callParams.OutputColCountLeftPad = outputColLeftPad; + callParams.OutputColCountNoPad = outputColNoPad; + callParams.OutputColCountRightPad = outputColRightPad; + callParams.FilterSetSize = min( MaxFilterSetSize, ( filterCount / BlockSize ) - filterSetIndex * MaxFilterSetSize ); + + const int blockedOutputWidth = BlockSize * outputWidth; + + while( workRem > 0 ) { + const int workThisIter = min( workRem, outputHeight - outputRowIndex ); + for( int ic = 0; ic < inputChannels; ic += BlockSize ) { + callParams.Flags = ic == 0 ? 0 : CBlockedConvGen::AccumulateOutputFlag; + + if( ic + BlockSize == inputChannels && callParams.FreeTerm != nullptr ) { + callParams.Flags |= CBlockedConvGen::AddFreeTermFlag; + } + + const float* input = inputImage + ic * inputHeight * inputWidth; + callParams.Output = output + outputRowIndex * blockedOutputWidth; + + for( int work = 0; work < workThisIter; ++work ) { + callParams.Filter = filter + BlockSize * ic * filterHeight * filterWidth; + callParams.FilterHeight = static_cast( filterHeight ); + + // Calculate filter intersection with top and bottom padding + const int firstInputRowIndex = ( outputRowIndex + work ) * desc.StrideHeight - desc.PaddingHeight; + const int topPaddingFilterRows = min( filterHeight, firstInputRowIndex < 0 + ? ( -firstInputRowIndex + dilationHeight - 1 ) / dilationHeight : 0 ); + const int lastInputRowIndex = firstInputRowIndex + ( filterHeight - 1 ) * dilationHeight; + const int bottomPaddingFilterRows = min( filterHeight, lastInputRowIndex >= inputHeight + ? ( lastInputRowIndex - inputHeight ) / dilationHeight + 1 : 0 ); + + // Skip filter rows which are covering top padding + callParams.Filter += topPaddingFilterRows * BlockSize * BlockSize * filterWidth; + // Don't process paddings + callParams.FilterHeight -= topPaddingFilterRows + bottomPaddingFilterRows; + + // Calculate input pointers accordingly (skip top padding) + const int inputRowOffset = firstInputRowIndex + topPaddingFilterRows * dilationHeight; + callParams.Input = input + BlockSize * ( inputRowOffset * inputWidth - desc.PaddingWidth ); + callParams.InputAfterLeftPad = input + BlockSize * ( inputRowOffset * inputWidth ); + + blockedConvGen.Run( callParams ); + + callParams.Output += blockedOutputWidth; + } + } + + workRem -= workThisIter; + if( outputRowIndex + workThisIter == outputHeight ) { + size_t blockedFilterCount = BlockSize * callParams.FilterSetSize; + output += blockedFilterCount * outputHeight * outputWidth; + filter += blockedFilterCount * inputChannels * filterHeight * filterWidth; + if( callParams.FreeTerm != nullptr ) { + callParams.FreeTerm += blockedFilterCount; + } + + if( ++filterSetIndex == filterSetCount ) { + inputImage += desc.Source.ObjectSize(); + filter = packedFilter; + callParams.FreeTerm = freeTerm; + filterSetIndex = 0; + } + + callParams.FilterSetSize = min( MaxFilterSetSize, ( filterCount / BlockSize ) - filterSetIndex * MaxFilterSetSize ); + outputRowIndex = 0; + } + } + + _mm256_zeroupper(); + } + } +} + +} // namespace NeoML diff --git a/NeoMathEngine/test/AvxTestDesktop/BlockedConvTest.cpp b/NeoMathEngine/test/AvxTestDesktop/BlockedConvTest.cpp new file mode 100644 index 000000000..e4ed2a500 --- /dev/null +++ b/NeoMathEngine/test/AvxTestDesktop/BlockedConvTest.cpp @@ -0,0 +1,5936 @@ +/* Copyright © 2017-2022 ABBYY Production LLC + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. +--------------------------------------------------------------------------------------------------------------*/ + +#include "common.h" +#pragma hdrstop + +#include +#include + +#include + +using namespace NeoML; + +namespace NeoMLTest { + +// The class to generate random values +// It uses the complementary-multiply-with-carry algorithm +// C lag-1024, multiplier(a) = 108798, initial carry(c) = 12345678 +class CRandom { +public: + explicit CRandom( unsigned int seed = 0xBADF00D ) { srand(seed); } + + // Returns the next random value + unsigned int Next() { return (rand() << 16) + rand(); } + + // Returns a double value from a uniform distribution in [ min, max ) range + // If min == max, min is returned + double Uniform( double min, double max ) { return min + (max - min) * Next() / 4294967296.; } + + // Returns an int value from a uniform distribution in [ min, max ] range. Note that max return value is possible! + // If min == max, min is returned + int UniformInt( int min, int max ) { return (int)(min + (unsigned int)(((static_cast(max) - min + 1) * Next()) >> 32)); } +}; + +template +class CBufferWrapper { +public: + CBufferWrapper( IMathEngine& _mathEngine, T* _data, int _size) : mathEngine( _mathEngine ), isCopyBack(false), size(_size), data(_data) + { + mathData = CTypedMemoryHandle( mathEngine.HeapAlloc( size * sizeof(T) ) ); + mathEngine.DataExchangeTyped(CTypedMemoryHandle(mathData), data, size); + } + + ~CBufferWrapper() + { + if(isCopyBack) { + mathEngine.DataExchangeTyped(data, CTypedMemoryHandle(mathData), size); + } + mathEngine.HeapFree(mathData); + } + + operator CTypedMemoryHandle() const { isCopyBack = true; return CTypedMemoryHandle(mathData); } + operator CTypedMemoryHandle() const { return CTypedMemoryHandle(mathData); } + +private: + IMathEngine& mathEngine; + mutable bool isCopyBack; + int size; + T* data; + CTypedMemoryHandle mathData; +}; + +typedef CBufferWrapper CFloatWrapper; +typedef CBufferWrapper CIntWrapper; + +#define CARRAY_WRAPPER(TYPE, arr) CBufferWrapper( MathEngine(), ( arr.data() ), ( static_cast( arr.size() ) ) ) +#define CARRAY_FLOAT_WRAPPER(arr) CARRAY_WRAPPER(float, arr) +#define CARRAY_INT_WRAPPER(arr) CARRAY_WRAPPER(int, arr) + +#define ARR_WRAPPER(TYPE, arr) CBufferWrapper( MathEngine(), (arr), (int)sizeof(arr) / sizeof(TYPE) ) +#define FLOAT_WRAPPER(arr) CFloatWrapper( MathEngine(), (arr), (int)sizeof(arr) / sizeof(float) ) +#define FLOAT_WRAPPER_MATHENGINE(mathEngine, arr) CFloatWrapper( mathEngine, (arr), (int)sizeof(arr) / sizeof(float) ) +#define INT_WRAPPER(arr) CIntWrapper( MathEngine(), (arr), (int)sizeof(arr) / sizeof(int) ) +#define INT_WRAPPER_MATHENGINE(mathEngine, arr) CIntWrapper( mathEngine, (arr), (int)sizeof(arr) / sizeof(int) ) + +#define CREATE_FILL_ARRAY(TYPE, arr, min, max, size, random) \ + std::vector arr; \ + arr.resize( size ); \ + for(int i = 0; i < size; ++i) { \ + arr[i] = static_cast( random.Uniform( min, max ) ); \ + } + +#define CREATE_FILL_FLOAT_ARRAY(arr, min, max, size, random) \ + CREATE_FILL_ARRAY(float, arr, min, max, size, random) + +// Leaving CREATE_FILL_INT_ARRAY as it was before CREATE_FILL_ARRAY +// for the sake of backward compatibility +#define CREATE_FILL_INT_ARRAY(arr, min, max, size, random) \ + std::vector arr; \ + arr.resize( size ); \ + for(int i = 0; i < size; ++i) { \ + arr[i] = random.UniformInt( min, max ); \ + } + +static inline int calcConvOutputSize( int input, int padding, int filter, int dilation, int stride ) +{ + return 1 + ( input - ( filter - 1 ) * dilation + 2 * padding - 1 ) / stride; +} + +} // namespace NeoMLTest + +using namespace NeoMLTest; + +class CBlockedConvTest : public ::testing::Test, public ::testing::WithParamInterface { +}; + +static const int RUN_COUNT = 100; + +TEST_P( CBlockedConvTest, Run ) +{ + const CTestParams& params = GetParam(); + CRandom random( params.GetValue( "Seed" ) ); + const int batch = params.GetValue( "Batch" ); + const int height = params.GetValue( "Height" ); + const int width = params.GetValue( "Width" ); + const int channels = params.GetValue( "Channels" ); + const int filterCount = params.GetValue( "FilterCount" ); + const int filterHeight = params.GetValue( "FilterHeight" ); + const int filterWidth = params.GetValue( "FilterWidth" ); + const int strideHeight = params.GetValue( "StrideHeight" ); + const int strideWidth = params.GetValue( "StrideWidth" ); + const int paddingHeight = params.GetValue( "PaddingHeight" ); + const int paddingWidth = params.GetValue( "PaddingWidth" ); + const int dilationHeight = params.GetValue( "DilationHeight" ); + const int dilationWidth = params.GetValue( "DilationWidth" ); + + const int outputHeight = calcConvOutputSize( height, paddingHeight, filterHeight, dilationHeight, strideHeight ); + const int outputWidth = calcConvOutputSize( width, paddingWidth, filterWidth, dilationWidth, strideWidth ); + + CBlobDesc inputDesc( { 1, batch, 1, height, width, 1, channels } ); + CBlobDesc filterDesc( { 1, filterCount, 1, filterHeight, filterWidth, 1, channels } ); + CBlobDesc outputDesc( { 1, batch, 1, outputHeight, outputWidth, 1, filterCount } ); + + std::unique_ptr blockedConvDesc( SimdMathEngine().InitBlockedConvolution( inputDesc, paddingHeight, paddingWidth, + strideHeight, strideWidth, dilationHeight, dilationWidth, filterDesc, outputDesc ) ); + if( blockedConvDesc == nullptr ) { + GTEST_SKIP() << "Config is not supported by SIMD"; + } + + CREATE_FILL_FLOAT_ARRAY( neomlInput, -2.f, 2.f, inputDesc.BlobSize(), random ); + CREATE_FILL_FLOAT_ARRAY( neomlFilter, -2.f, 2.f, filterDesc.BlobSize(), random ); + CREATE_FILL_FLOAT_ARRAY( bias, -5.f, 5.f, filterCount, random ); + std::vector expectedOutput( outputDesc.BlobSize() ); + + std::unique_ptr counters( MathEngine().CreatePerformanceCounters() ); + + using CPerfStat = std::vector; + + auto update = []( const IPerformanceCounters& counters, CPerfStat& stat ) + { + for( size_t i = 0; i < counters.size(); ++i ) { + stat[i] += counters[i].Value; + } + }; + + CPerfStat neomlPerf( counters->size() ); + CPerfStat allocPerf( counters->size() ); + CPerfStat inputConversionPerf( counters->size() ); + CPerfStat filterConversionPerf( counters->size() ); + CPerfStat blockedConvPerf( counters->size() ); + CPerfStat outputConversionPerf( counters->size() ); + CPerfStat outputCopyPerf( counters->size() ); + CPerfStat freePerf( counters->size() ); + + { + CConvolutionDesc* convDesc = MathEngine().InitBlobConvolution( inputDesc, paddingHeight, paddingWidth, strideHeight, + strideWidth, dilationHeight, dilationWidth, filterDesc, outputDesc ); + auto biasHandle = CARRAY_FLOAT_WRAPPER( bias ); + auto inputHandle = CARRAY_FLOAT_WRAPPER( neomlInput ); + auto filterHandle = CARRAY_FLOAT_WRAPPER( neomlFilter ); + auto outputHandle = CARRAY_FLOAT_WRAPPER( expectedOutput ); + for( int run = 0; run <= RUN_COUNT; ++run ) { + counters->Synchronise(); + MathEngine().BlobConvolution( *convDesc, inputHandle, filterHandle, &static_cast( biasHandle ), outputHandle ); + counters->Synchronise(); + if( run != 0 ) { + update( *counters, neomlPerf ); + } + } + delete convDesc; + } + + for( int run = 0; run <= RUN_COUNT; ++run ) { + CFloatHandleStackVar* stackVar; + counters->Synchronise(); + stackVar = new CFloatHandleStackVar( MathEngine(), neomlFilter.size() + std::max( neomlInput.size(), expectedOutput.size() ) ); + float* blockedFilter = static_cast( MathEngine().GetBuffer( stackVar->GetHandle(), 0, stackVar->Size(), false ) ); + float* ioBuff = blockedFilter + neomlFilter.size(); + counters->Synchronise(); + if( run != 0 ) update( *counters, allocPerf ); + + CFloatHandleStackVar* actualOutputVar = new CFloatHandleStackVar( MathEngine(), expectedOutput.size() ); + float* actualOutputBuff = static_cast( MathEngine().GetBuffer( actualOutputVar->GetHandle(), 0, actualOutputVar->Size(), false ) ); + + counters->Synchronise(); + SimdMathEngine().PackBlockedData( inputDesc, neomlInput.data(), ioBuff ); + counters->Synchronise(); + if( run != 0 ) update( *counters, inputConversionPerf ); + + counters->Synchronise(); + SimdMathEngine().PackBlockedFilter( filterDesc, neomlFilter.data(), blockedFilter ); + counters->Synchronise(); + if( run != 0 ) update( *counters, filterConversionPerf ); + + counters->Synchronise(); + SimdMathEngine().BlockedConvolution( *blockedConvDesc, ioBuff, blockedFilter, bias.data(), actualOutputBuff ); + counters->Synchronise(); + if( run != 0 ) update( *counters, blockedConvPerf ); + + counters->Synchronise(); + SimdMathEngine().UnpackBlockedData( outputDesc, actualOutputBuff, ioBuff ); + counters->Synchronise(); + if( run != 0 ) update( *counters, outputConversionPerf ); + + MathEngine().ReleaseBuffer( actualOutputVar->GetHandle(), actualOutputBuff, true ); + MathEngine().ReleaseBuffer( stackVar->GetHandle(), blockedFilter, true ); + + counters->Synchronise(); + MathEngine().VectorCopy( actualOutputVar->GetHandle(), stackVar->GetHandle() + neomlFilter.size(), + static_cast( expectedOutput.size() ) ); + counters->Synchronise(); + if( run != 0 ) update( *counters, outputCopyPerf ); + + std::vector actualOutput( expectedOutput.size() ); + MathEngine().DataExchangeTyped( actualOutput.data(), actualOutputVar->GetHandle(), actualOutput.size() ); + delete actualOutputVar; + + counters->Synchronise(); + delete stackVar; + counters->Synchronise(); + if( run != 0 ) update( *counters, freePerf ); + + if( run == 0 ) { + for( size_t i = 0; i < actualOutput.size(); ++i ) { + ASSERT_NEAR( actualOutput[i], expectedOutput[i], 1e-2f ) << "at #" << i; + } + } + } + + std::cout << "NeoML\tAlloc\tInput\tFilter\tBlocked\tOutput\tCopy\tFree\tInputSize\tDilationWidth\tInputChannels\tOutputChannels\tCoeffI\tCoeffO\tCoeffF\n" + << neomlPerf[0] / 1e9 << '\t' + << allocPerf[0] / 1e9 << '\t' + << inputConversionPerf[0] / 1e9 << '\t' + << filterConversionPerf[0] / 1e9 << '\t' + << blockedConvPerf[0] / 1e9 << '\t' + << outputConversionPerf[0] / 1e9 << '\t' + << outputCopyPerf[0] / 1e9 << '\t' + << freePerf[0] / 1e9 << '\t' + << inputDesc.BlobSize() << '\t' + << dilationWidth << '\t' + << channels << '\t' + << filterCount << '\t' + << outputDesc.Height() * outputDesc.Width() * filterCount * filterHeight * filterWidth / ( height * width ) << '\t' + << channels * filterHeight * filterWidth << '\t' + << batch * outputDesc.Height() * outputDesc.Width() + << '\n'; +} + +struct BlockedConvTestNameGenerator { + std::string operator()(const testing::TestParamInfo& paramInfo) + { + const CTestParams& params = paramInfo.param; + if( !params.Name().empty() ) { + return params.Name(); + } + + const int batch = params.GetValue( "Batch" ); + const int height = params.GetValue( "Height" ); + const int width = params.GetValue( "Width" ); + const int channels = params.GetValue( "Channels" ); + const int filterCount = params.GetValue( "FilterCount" ); + const int filterHeight = params.GetValue( "FilterHeight" ); + const int filterWidth = params.GetValue( "FilterWidth" ); + const int strideHeight = params.GetValue( "StrideHeight" ); + const int strideWidth = params.GetValue( "StrideWidth" ); + const int paddingHeight = params.GetValue( "PaddingHeight" ); + const int paddingWidth = params.GetValue( "PaddingWidth" ); + const int dilationHeight = params.GetValue( "DilationHeight" ); + const int dilationWidth = params.GetValue( "DilationWidth" ); + + + const int bufferSize = 1024; + int currLen = 0; + std::vector buffer; + buffer.resize( 1024 ); + + currLen += ::sprintf_s( buffer.data() + currLen, bufferSize - currLen, "I_%dx%dx%dx%d__", batch, height, width, channels ); + currLen += ::sprintf_s( buffer.data() + currLen, bufferSize - currLen, "F_%dx%dx%d__", filterCount, filterHeight, filterWidth ); + currLen += ::sprintf_s( buffer.data() + currLen, bufferSize - currLen, "St_%dx%d__", strideHeight, strideWidth ); + currLen += ::sprintf_s( buffer.data() + currLen, bufferSize - currLen, "P_%dx%d__", paddingHeight, paddingWidth ); + currLen += ::sprintf_s( buffer.data() + currLen, bufferSize - currLen, "D_%dx%d", dilationHeight, dilationWidth ); + + return std::string( buffer.data(), currLen ); + } +}; + +INSTANTIATE_TEST_SUITE_P( Trivial, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Minimal" + ), + CTestParams( + "Batch = 2;" + "Height = 1;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Batch" + ), + CTestParams( + "Batch = 1;" + "Height = 2;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "InputHeight" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 2;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "InputWidth" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 1;" + "Channels = 16;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "InputChannels" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "FilterCount" + ), + CTestParams( + "Batch = 1;" + "Height = 2;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 2;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "FilterHeight" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 2;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 2;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "FilterWidth" + ), + CTestParams( + "Batch = 1;" + "Height = 2;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 3;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "PaddingHeight" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 3;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "TrickyPaddingHeight" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 2;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "PaddingWidth" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "TrickyPaddingWidth" + ), + CTestParams( + "Batch = 1;" + "Height = 3;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 2;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "StrideHeight" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 3;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 2;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "StrideWidth" + ), + CTestParams( + "Batch = 1;" + "Height = 3;" + "Width = 1;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 2;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 2;" + "DilationWidth = 1;" + "Seed = 348;", + "DilationHeight" + ), + CTestParams( + "Batch = 1;" + "Height = 1;" + "Width = 3;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 1;" + "FilterWidth = 2;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 2;" + "Seed = 348;", + "DilationWidth" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P(Yolox0, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 320;" + "Width = 640;" + "Channels = 32;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_16_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 160;" + "Width = 320;" + "Channels = 64;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_19_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 160;" + "Width = 320;" + "Channels = 32;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_25_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 160;" + "Width = 320;" + "Channels = 32;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_28_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 160;" + "Width = 320;" + "Channels = 64;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_22_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 160;" + "Width = 320;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_33_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 160;" + "Width = 320;" + "Channels = 64;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_36_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_39_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_45_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_48_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_52_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_55_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_59_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_62_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_42_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_67_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_70_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_73_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_79_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_82_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_86_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_89_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_93_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_96_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_76_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_101_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_104_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_107_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 1024;" + "FilterCount = 512;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_114_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_117_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_123_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_126_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_120_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 512;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_130_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_133_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 512;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_138_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_144_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_147_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 512;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_141_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_151_Op" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P(Yolox1, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_154_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 256;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_159_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_165_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_168_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 256;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_162_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_172_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_215_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_218_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_221_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_175_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_179_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_185_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_188_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_182_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_192_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_233_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_236_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_239_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_195_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_199_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_205_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_208_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_202_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 512;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_212_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 512;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_251_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_254_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_257_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_224_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 80;" + "Width = 160;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_227_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_242_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 40;" + "Width = 80;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_245_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_260_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 20;" + "Width = 40;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 348;", + "Conv_263_Op" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P(DenseNetOnnx, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 56;" + "Width = 56;" + "Channels = 128;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "conv2_1_x2_1_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 28;" + "Width = 28;" + "Channels = 128;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "conv3_1_x2_1_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 14;" + "Width = 14;" + "Channels = 128;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "conv4_1_x2_1_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 7;" + "Width = 7;" + "Channels = 128;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "conv5_1_x2_1_Op" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P(Vgg16Onnx, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 224;" + "Width = 224;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "vgg0_conv1_fwd_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 112;" + "Width = 112;" + "Channels = 64;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "vgg0_conv2_fwd_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 112;" + "Width = 112;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "vgg0_conv3_fwd_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 56;" + "Width = 56;" + "Channels = 128;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "vgg0_conv4_fwd_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 56;" + "Width = 56;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "vgg0_conv5_fwd_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 28;" + "Width = 28;" + "Channels = 256;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "vgg0_conv7_fwd_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 28;" + "Width = 28;" + "Channels = 512;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "vgg0_conv8_fwd_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 14;" + "Width = 14;" + "Channels = 512;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "vgg0_conv10_fwd_Op" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P(YoloV2CocoOnnx, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 208;" + "Width = 208;" + "Channels = 32;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "139_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 104;" + "Width = 104;" + "Channels = 64;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "143_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 52;" + "Width = 52;" + "Channels = 128;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "153_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 26;" + "Width = 26;" + "Channels = 256;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "163_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 13;" + "Width = 13;" + "Channels = 512;" + "FilterCount = 1024;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "179_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 13;" + "Width = 13;" + "Channels = 1024;" + "FilterCount = 1024;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "194_Op" + ), + CTestParams( + "Batch = 1;" + "Height = 13;" + "Width = 13;" + "Channels = 1280;" + "FilterCount = 1024;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 123456;", + "215_Op" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P(DANet, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 36;" + "Width = 24;" + "Channels = 32;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "26" + ), + CTestParams( + "Batch = 1;" + "Height = 72;" + "Width = 48;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 9;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 4;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "16" + ), + CTestParams( + "Batch = 1;" + "Height = 72;" + "Width = 48;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 9;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "14" + ), + CTestParams( + "Batch = 1;" + "Height = 288;" + "Width = 192;" + "Channels = 8;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "8" + ), + CTestParams( + "Batch = 1;" + "Height = 144;" + "Width = 96;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "11" + ), + CTestParams( + "Batch = 1;" + "Height = 72;" + "Width = 48;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "18" + ), + CTestParams( + "Batch = 1;" + "Height = 36;" + "Width = 24;" + "Channels = 32;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 9;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 4;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "24" + ), + CTestParams( + "Batch = 1;" + "Height = 36;" + "Width = 24;" + "Channels = 32;" + "FilterCount = 32;" + "FilterHeight = 9;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "22" + ), + CTestParams( + "Batch = 1;" + "Height = 288;" + "Width = 192;" + "Channels = 16;" + "FilterCount = 8;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "49" + ), + CTestParams( + "Batch = 1;" + "Height = 18;" + "Width = 12;" + "Channels = 64;" + "FilterCount = 32;" + "FilterHeight = 9;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "30" + ), + CTestParams( + "Batch = 1;" + "Height = 18;" + "Width = 12;" + "Channels = 32;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 9;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 4;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "32" + ), + CTestParams( + "Batch = 1;" + "Height = 18;" + "Width = 12;" + "Channels = 64;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "34" + ), + CTestParams( + "Batch = 1;" + "Height = 288;" + "Width = 192;" + "Channels = 8;" + "FilterCount = 8;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "46" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( BusinessObjects, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 288;" + "Width = 192;" + "Channels = 8;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "8" + ), + CTestParams( + "Batch = 1;" + "Height = 144;" + "Width = 96;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 8;" + "PaddingWidth = 8;" + "DilationHeight = 8;" + "DilationWidth = 8;" + "Seed = 1984;", + "conv_10" + ), + CTestParams( + "Batch = 1;" + "Height = 288;" + "Width = 192;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_6" + ), + CTestParams( + "Batch = 1;" + "Height = 144;" + "Width = 96;" + "Channels = 16;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_7" + ), + CTestParams( + "Batch = 1;" + "Height = 144;" + "Width = 96;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 2;" + "PaddingWidth = 2;" + "DilationHeight = 2;" + "DilationWidth = 2;" + "Seed = 1984;", + "conv_8" + ), + CTestParams( + "Batch = 1;" + "Height = 144;" + "Width = 96;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 4;" + "DilationHeight = 4;" + "DilationWidth = 4;" + "Seed = 1984;", + "conv_9" + ), + CTestParams( + "Batch = 1;" + "Height = 144;" + "Width = 96;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 16;" + "PaddingWidth = 16;" + "DilationHeight = 16;" + "DilationWidth = 16;" + "Seed = 1984;", + "conv11" + ), + CTestParams( + "Batch = 1;" + "Height = 144;" + "Width = 96;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv_final" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( KoreanTwoStageRleSecondLevel, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 12;" + "Channels = 32;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv0" + ), + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 12;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 10;" + "Width = 10;" + "Channels = 16;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv2" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( EndToEndArabicFast, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 64;" + "Channels = 160;" + "FilterCount = 80;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( EndToEndArabic, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 32;" + "Channels = 160;" + "FilterCount = 80;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv0" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 32;" + "Channels = 80;" + "FilterCount = 160;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv1" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( JapaneseTwoStageGrayFirstLevel, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 12;" + "Channels = 16;" + "FilterCount = 8;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 3;" + "StrideWidth = 3;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv0" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 8;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv1" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( JapaneseTwoStageGraySecondLevel, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 15;" + "Width = 15;" + "Channels = 8;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv0" + ), + CTestParams( + "Batch = 1;" + "Height = 15;" + "Width = 15;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 3;" + "StrideWidth = 3;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 5;" + "Width = 5;" + "Channels = 16;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv2" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( SegLinks, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block1_conv2" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 64;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block2_conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 128;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv_9_2" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block2_conv2" + ), + CTestParams( + "Batch = 1;" + "Height = 128;" + "Width = 128;" + "Channels = 128;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block3_conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 128;" + "Width = 128;" + "Channels = 256;" + "FilterCount = 256;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block3_conv2" + ), + CTestParams( + "Batch = 1;" + "Height = 64;" + "Width = 64;" + "Channels = 256;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block4_conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 64;" + "Width = 64;" + "Channels = 512;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block4_conv2" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 512;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block5_conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 512;" + "FilterCount = 1024;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv_6" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 1024;" + "FilterCount = 1024;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv_7" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 1024;" + "FilterCount = 256;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv_8_1" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 256;" + "FilterCount = 512;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv_8_2" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 512;" + "FilterCount = 128;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv_9_1" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( SegLinksDepthwise, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 64;" + "Width = 64;" + "Channels = 24;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_4_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "expanded_conv_project" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 16;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_1_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 128;" + "Width = 128;" + "Channels = 96;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_1_project" + ), + CTestParams( + "Batch = 1;" + "Height = 64;" + "Width = 64;" + "Channels = 144;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_3_project" + ), + CTestParams( + "Batch = 1;" + "Height = 128;" + "Width = 128;" + "Channels = 24;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_2_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 128;" + "Width = 128;" + "Channels = 144;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_2_project" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 144;" + "FilterCount = 48;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_6_project" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 48;" + "FilterCount = 288;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_7_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 288;" + "FilterCount = 48;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_7_project" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 288;" + "FilterCount = 72;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_10_project" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 72;" + "FilterCount = 432;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_11_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 432;" + "FilterCount = 72;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_11_project" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 432;" + "FilterCount = 104;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_13_project" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 104;" + "FilterCount = 624;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_29_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 104;" + "FilterCount = 624;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_14_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 624;" + "FilterCount = 104;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_28_project" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 624;" + "FilterCount = 104;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_14_project" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 624;" + "FilterCount = 104;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_25_project" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 104;" + "FilterCount = 624;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_26_expand" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( SegLinksDepthwise2, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 32;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "expanded_conv_project" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 16;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_1_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 128;" + "Width = 128;" + "Channels = 96;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_1_project" + ), + CTestParams( + "Batch = 1;" + "Height = 128;" + "Width = 128;" + "Channels = 24;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_2_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 128;" + "Width = 128;" + "Channels = 144;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_2_project" + ), + CTestParams( + "Batch = 1;" + "Height = 64;" + "Width = 64;" + "Channels = 144;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_3_project" + ), + CTestParams( + "Batch = 1;" + "Height = 64;" + "Width = 64;" + "Channels = 32;" + "FilterCount = 192;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_4_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 64;" + "Width = 64;" + "Channels = 192;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_4_project" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 192;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_6_project" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 64;" + "FilterCount = 384;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_7_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 384;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_7_project" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 384;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_10_project" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 96;" + "FilterCount = 576;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_11_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 576;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_11_project" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 864;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_28_project" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 864;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_14_project" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 576;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_13_project" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 144;" + "FilterCount = 864;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_29_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 144;" + "FilterCount = 864;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_14_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 864;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_25_project" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 144;" + "FilterCount = 864;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_26_expand" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( PassportScanSegmentation, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_14" + ), + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 16;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_15" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 32;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_16_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 4;" + "DilationHeight = 4;" + "DilationWidth = 4;" + "Seed = 1984;", + "conv2d_18_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 2;" + "PaddingWidth = 2;" + "DilationHeight = 2;" + "DilationWidth = 2;" + "Seed = 1984;", + "conv2d_17_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 8;" + "PaddingWidth = 8;" + "DilationHeight = 8;" + "DilationWidth = 8;" + "Seed = 1984;", + "conv2d_19_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 64;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 16;" + "PaddingWidth = 16;" + "DilationHeight = 16;" + "DilationWidth = 16;" + "Seed = 1984;", + "conv2d_20_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 4;" + "DilationHeight = 4;" + "DilationWidth = 4;" + "Seed = 1984;", + "conv2d_21_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 2;" + "PaddingWidth = 2;" + "DilationHeight = 2;" + "DilationWidth = 2;" + "Seed = 1984;", + "conv2d_22_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 128;" + "FilterCount = 64;" + "FilterHeight = 7;" + "FilterWidth = 7;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 3;" + "PaddingWidth = 3;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_23_" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( PassportPhotoSegmentation, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_2" + ), + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 16;" + "FilterCount = 32;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_3" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 32;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_4_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 4;" + "DilationHeight = 4;" + "DilationWidth = 4;" + "Seed = 1984;", + "conv2d_6_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 2;" + "PaddingWidth = 2;" + "DilationHeight = 2;" + "DilationWidth = 2;" + "Seed = 1984;", + "conv2d_5_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 64;" + "FilterCount = 64;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 8;" + "PaddingWidth = 8;" + "DilationHeight = 8;" + "DilationWidth = 8;" + "Seed = 1984;", + "conv2d_7_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 96;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_8_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 128;" + "FilterCount = 128;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_9_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 128;" + "FilterCount = 64;" + "FilterHeight = 7;" + "FilterWidth = 7;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 3;" + "PaddingWidth = 3;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_11_" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( SolitaireIDSegmentation, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_2_" + ), + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_3_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_4_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 48;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 2;" + "PaddingWidth = 2;" + "DilationHeight = 2;" + "DilationWidth = 2;" + "Seed = 1984;", + "conv2d_5_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 48;" + "FilterCount = 96;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 4;" + "DilationHeight = 4;" + "DilationWidth = 4;" + "Seed = 1984;", + "conv2d_6_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 96;" + "FilterCount = 144;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 8;" + "PaddingWidth = 8;" + "DilationHeight = 8;" + "DilationWidth = 8;" + "Seed = 1984;", + "conv2d_7_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 144;" + "FilterCount = 192;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 16;" + "PaddingWidth = 16;" + "DilationHeight = 16;" + "DilationWidth = 16;" + "Seed = 1984;", + "conv2d_8_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 216;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_9_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 192;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_14_" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( BarcodesPostcodes, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "separable_conv2d_2_pointwise" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "separable_conv2d_3_pointwise" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_1_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 4;" + "DilationHeight = 4;" + "DilationWidth = 4;" + "Seed = 1984;", + "conv2d_3_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 2;" + "PaddingWidth = 2;" + "DilationHeight = 2;" + "DilationWidth = 2;" + "Seed = 1984;", + "conv2d_2_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 8;" + "PaddingWidth = 8;" + "DilationHeight = 8;" + "DilationWidth = 8;" + "Seed = 1984;", + "conv2d_4_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 16;" + "PaddingWidth = 16;" + "DilationHeight = 16;" + "DilationWidth = 16;" + "Seed = 1984;", + "conv2d_5_" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( BarcodesNoPostcodes, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_11_" + ), + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_12_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_13_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 4;" + "DilationHeight = 4;" + "DilationWidth = 4;" + "Seed = 1984;", + "conv2d_15_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 2;" + "PaddingWidth = 2;" + "DilationHeight = 2;" + "DilationWidth = 2;" + "Seed = 1984;", + "conv2d_14_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 8;" + "PaddingWidth = 8;" + "DilationHeight = 8;" + "DilationWidth = 8;" + "Seed = 1984;", + "conv2d_16_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 16;" + "PaddingWidth = 16;" + "DilationHeight = 16;" + "DilationWidth = 16;" + "Seed = 1984;", + "conv2d_17_" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( FindText, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 512;" + "Width = 512;" + "Channels = 16;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 2;" + "StrideWidth = 2;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_9_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_10_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 24;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 2;" + "PaddingWidth = 2;" + "DilationHeight = 2;" + "DilationWidth = 2;" + "Seed = 1984;", + "conv2d_11_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 24;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 4;" + "PaddingWidth = 4;" + "DilationHeight = 4;" + "DilationWidth = 4;" + "Seed = 1984;", + "conv2d_12_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 8;" + "PaddingWidth = 8;" + "DilationHeight = 8;" + "DilationWidth = 8;" + "Seed = 1984;", + "conv2d_13_" + ), + CTestParams( + "Batch = 1;" + "Height = 256;" + "Width = 256;" + "Channels = 16;" + "FilterCount = 16;" + "FilterHeight = 3;" + "FilterWidth = 3;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 1;" + "PaddingWidth = 1;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2d_14_" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( MobileNetV2Cifar10, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 32;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block0convShortcut" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 144;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block20conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 32;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block0conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 16;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block10convShortcut" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 16;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block10conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 384;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block31conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 96;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block10conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 24;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block11conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 32;" + "Width = 32;" + "Channels = 144;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block11conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 32;" + "FilterCount = 192;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block21conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 96;" + "FilterCount = 576;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block41conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 16;" + "Width = 16;" + "Channels = 192;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block21conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 320;" + "FilterCount = 1280;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "conv2" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 192;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block30conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 64;" + "FilterCount = 384;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block31conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 64;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block40convShortcut" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 384;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block40conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 8;" + "Width = 8;" + "Channels = 576;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block41conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 576;" + "FilterCount = 160;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block50conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 160;" + "FilterCount = 960;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block51conv1" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 960;" + "FilterCount = 160;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block51conv3" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 160;" + "FilterCount = 320;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block6convShortcut" + ), + CTestParams( + "Batch = 1;" + "Height = 4;" + "Width = 4;" + "Channels = 960;" + "FilterCount = 320;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block6conv3" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( archmobilenetv2_accurate, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 384;" + "Width = 256;" + "Channels = 32;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "expanded_conv_project" + ), + CTestParams( + "Batch = 1;" + "Height = 384;" + "Width = 256;" + "Channels = 16;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_1_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 192;" + "Width = 128;" + "Channels = 96;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_1_project" + ), + CTestParams( + "Batch = 1;" + "Height = 192;" + "Width = 128;" + "Channels = 24;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_2_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 192;" + "Width = 128;" + "Channels = 144;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_2_project" + ), + CTestParams( + "Batch = 1;" + "Height = 96;" + "Width = 64;" + "Channels = 144;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_3_project" + ), + CTestParams( + "Batch = 1;" + "Height = 96;" + "Width = 64;" + "Channels = 32;" + "FilterCount = 192;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_4_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 96;" + "Width = 64;" + "Channels = 192;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_4_project" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 32;" + "Channels = 192;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_6_project" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 32;" + "Channels = 64;" + "FilterCount = 384;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_7_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 32;" + "Channels = 384;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_7_project" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 32;" + "Channels = 384;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_10_project" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 16;" + "Channels = 864;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_14_project" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 32;" + "Channels = 96;" + "FilterCount = 576;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_11_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 32;" + "Channels = 576;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_11_project" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 16;" + "Channels = 576;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_13_project" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 16;" + "Channels = 144;" + "FilterCount = 864;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_14_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 8;" + "Channels = 864;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_25_project" + ), + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 8;" + "Channels = 144;" + "FilterCount = 864;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_26_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 6;" + "Width = 4;" + "Channels = 864;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_28_project" + ), + CTestParams( + "Batch = 1;" + "Height = 6;" + "Width = 4;" + "Channels = 144;" + "FilterCount = 864;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_29_expand" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( TorchVisionMobileNetV2, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 112;" + "Width = 112;" + "Channels = 32;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer7" + ), + CTestParams( + "Batch = 1;" + "Height = 112;" + "Width = 112;" + "Channels = 16;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer9" + ), + CTestParams( + "Batch = 1;" + "Height = 56;" + "Width = 56;" + "Channels = 96;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer15" + ), + CTestParams( + "Batch = 1;" + "Height = 56;" + "Width = 56;" + "Channels = 24;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer17" + ), + CTestParams( + "Batch = 1;" + "Height = 56;" + "Width = 56;" + "Channels = 144;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer23" + ), + CTestParams( + "Batch = 1;" + "Height = 28;" + "Width = 28;" + "Channels = 144;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer32" + ), + CTestParams( + "Batch = 1;" + "Height = 28;" + "Width = 28;" + "Channels = 32;" + "FilterCount = 192;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer34" + ), + CTestParams( + "Batch = 1;" + "Height = 28;" + "Width = 28;" + "Channels = 192;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer40" + ), + CTestParams( + "Batch = 1;" + "Height = 14;" + "Width = 14;" + "Channels = 192;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer58" + ), + CTestParams( + "Batch = 1;" + "Height = 14;" + "Width = 14;" + "Channels = 64;" + "FilterCount = 384;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer60" + ), + CTestParams( + "Batch = 1;" + "Height = 14;" + "Width = 14;" + "Channels = 384;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer66" + ), + CTestParams( + "Batch = 1;" + "Height = 14;" + "Width = 14;" + "Channels = 384;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer93" + ), + CTestParams( + "Batch = 1;" + "Height = 14;" + "Width = 14;" + "Channels = 96;" + "FilterCount = 576;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer95" + ), + CTestParams( + "Batch = 1;" + "Height = 14;" + "Width = 14;" + "Channels = 576;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer101" + ), + CTestParams( + "Batch = 1;" + "Height = 7;" + "Width = 7;" + "Channels = 576;" + "FilterCount = 160;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer119" + ), + CTestParams( + "Batch = 1;" + "Height = 7;" + "Width = 7;" + "Channels = 320;" + "FilterCount = 1280;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer147" + ), + CTestParams( + "Batch = 1;" + "Height = 7;" + "Width = 7;" + "Channels = 160;" + "FilterCount = 960;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer121" + ), + CTestParams( + "Batch = 1;" + "Height = 7;" + "Width = 7;" + "Channels = 960;" + "FilterCount = 160;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer127" + ), + CTestParams( + "Batch = 1;" + "Height = 7;" + "Width = 7;" + "Channels = 960;" + "FilterCount = 320;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "FineAILayer145" + ) + ), BlockedConvTestNameGenerator() ); + +INSTANTIATE_TEST_SUITE_P( ScreenshotFeatureExtractor, CBlockedConvTest, + ::testing::Values( + CTestParams( + "Batch = 1;" + "Height = 192;" + "Width = 192;" + "Channels = 32;" + "FilterCount = 16;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "expanded_conv_project" + ), + CTestParams( + "Batch = 1;" + "Height = 192;" + "Width = 192;" + "Channels = 16;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_1_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 96;" + "Width = 96;" + "Channels = 96;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_1_project" + ), + CTestParams( + "Batch = 1;" + "Height = 96;" + "Width = 96;" + "Channels = 24;" + "FilterCount = 144;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_2_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 96;" + "Width = 96;" + "Channels = 144;" + "FilterCount = 24;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_2_project" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 48;" + "Channels = 144;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_3_project" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 48;" + "Channels = 32;" + "FilterCount = 192;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_4_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 48;" + "Width = 48;" + "Channels = 192;" + "FilterCount = 32;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_4_project" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 24;" + "Channels = 192;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_6_project" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 24;" + "Channels = 64;" + "FilterCount = 384;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_7_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 24;" + "Channels = 384;" + "FilterCount = 64;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_7_project" + ), + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 12;" + "Channels = 320;" + "FilterCount = 1280;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "Conv_1" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 24;" + "Channels = 384;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_10_project" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 24;" + "Channels = 96;" + "FilterCount = 576;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_11_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 24;" + "Width = 24;" + "Channels = 576;" + "FilterCount = 96;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_11_project" + ), + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 12;" + "Channels = 576;" + "FilterCount = 160;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_13_project" + ), + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 12;" + "Channels = 160;" + "FilterCount = 960;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_14_expand" + ), + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 12;" + "Channels = 960;" + "FilterCount = 160;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_14_project" + ), + CTestParams( + "Batch = 1;" + "Height = 12;" + "Width = 12;" + "Channels = 960;" + "FilterCount = 320;" + "FilterHeight = 1;" + "FilterWidth = 1;" + "StrideHeight = 1;" + "StrideWidth = 1;" + "PaddingHeight = 0;" + "PaddingWidth = 0;" + "DilationHeight = 1;" + "DilationWidth = 1;" + "Seed = 1984;", + "block_16_project" + ) + ), BlockedConvTestNameGenerator() ); diff --git a/NeoMathEngine/test/AvxTestDesktop/CMakeLists.txt b/NeoMathEngine/test/AvxTestDesktop/CMakeLists.txt new file mode 100644 index 000000000..3b3c98ec9 --- /dev/null +++ b/NeoMathEngine/test/AvxTestDesktop/CMakeLists.txt @@ -0,0 +1,50 @@ +project(NeoAvxTestDesktop) + +include(Utils) + +add_executable(${PROJECT_NAME} + BlockedConvTest.cpp + common.cpp + main.cpp + + common.h +) + +set_target_properties( ${PROJECT_NAME} PROPERTIES + UNITY_BUILD_MODE BATCH + UNITY_BUILD_BATCH_SIZE ${NeoML_UNITY_BUILD_BATCH_SIZE} +) + +if(MSVC) + target_link_options(${PROJECT_NAME} PRIVATE "/SUBSYSTEM:Console") +endif() + +target_link_libraries(${PROJECT_NAME} PRIVATE + CommonTestTools + NeoMathEngine + gtest +) + +configure_target(${PROJECT_NAME}) + +if(USE_FINE_OBJECTS AND (IOS OR DARWIN)) + #TODO: fix FineGTest + target_compile_options(${PROJECT_NAME} PRIVATE + $<$:-Wno-deprecated-copy> + $<$:-Wno-unknown-warning-option>) +endif() + +add_gtest_for_target(${PROJECT_NAME} CPU ${CMAKE_CURRENT_SOURCE_DIR}) +if(WIN32 AND CMAKE_CUDA_COMPILER) + add_gtest_for_target(${PROJECT_NAME} GPU ${CMAKE_CURRENT_SOURCE_DIR}) +endif() + +if(USE_FINE_OBJECTS) + fine_install(TARGETS ${PROJECT_NAME}) +else() + install( + TARGETS ${PROJECT_NAME} + LIBRARY DESTINATION ${CMAKE_INSTALL_LIBDIR} + RUNTIME DESTINATION ${CMAKE_INSTALL_BINDIR} + ) +endif() diff --git a/NeoMathEngine/test/AvxTestDesktop/common.cpp b/NeoMathEngine/test/AvxTestDesktop/common.cpp new file mode 100644 index 000000000..e2e7c32bf --- /dev/null +++ b/NeoMathEngine/test/AvxTestDesktop/common.cpp @@ -0,0 +1,91 @@ +/* Copyright © 2017-2022 ABBYY Production LLC + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. +--------------------------------------------------------------------------------------------------------------*/ + +#include "common.h" + +#include +#include +#include +#include "windows.h" + +using namespace NeoML; + +class CMathEngineInstance { +public: + explicit CMathEngineInstance( int threadCount ); + ~CMathEngineInstance(); + + CMathEngineInstance( const CMathEngineInstance& ) = delete; + CMathEngineInstance( const CMathEngineInstance&& ) = delete; + CMathEngineInstance& operator=( const CMathEngineInstance& ) = delete; + CMathEngineInstance& operator=( const CMathEngineInstance&& ) = delete; + + IMathEngine& GetMathEngine() { return *mathEngine; } + ISimdMathEngine& GetSimdMathEngine() { return *simdMathEngine; } + +private: + std::mutex mutex; + HMODULE avxModule; + IMathEngine* mathEngine; + ISimdMathEngine* simdMathEngine; +}; + +typedef ISimdMathEngine* ( TCreateSimdMathEngine )( IMathEngine*, int ); + +CMathEngineInstance::CMathEngineInstance( int threadCount ) : + avxModule( NULL ), + mathEngine( nullptr ), + simdMathEngine( nullptr ) +{ + avxModule = ::LoadLibrary( "NeoMathEngineAvx.dll" ); + assert( avxModule != nullptr ); + + FARPROC createSimdME = ::GetProcAddress( avxModule, "CreateSimdMathEngine" ); + assert( createSimdME != nullptr ); + + mathEngine = CreateCpuMathEngine( threadCount, 0 ); + assert( mathEngine != nullptr ); + + simdMathEngine = reinterpret_cast( createSimdME )( mathEngine, threadCount ); + assert( simdMathEngine != nullptr ); +} + +CMathEngineInstance::~CMathEngineInstance() +{ + if( simdMathEngine != nullptr ) { + delete simdMathEngine; + } + + if( mathEngine != nullptr ) { + delete mathEngine; + } + + if( avxModule != nullptr ) { + ::FreeModule( avxModule ); + } +} + +static constexpr int meThreadCount = 1; +static CMathEngineInstance meInstance( meThreadCount ); + +IMathEngine& MathEngine() +{ + return meInstance.GetMathEngine(); +} + +ISimdMathEngine& SimdMathEngine() +{ + return meInstance.GetSimdMathEngine(); +} diff --git a/NeoMathEngine/test/AvxTestDesktop/common.h b/NeoMathEngine/test/AvxTestDesktop/common.h new file mode 100644 index 000000000..21d760e9d --- /dev/null +++ b/NeoMathEngine/test/AvxTestDesktop/common.h @@ -0,0 +1,22 @@ +/* Copyright © 2017-2022 ABBYY Production LLC + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. +--------------------------------------------------------------------------------------------------------------*/ + +#pragma once + +#include +#include + +NeoML::IMathEngine& MathEngine(); +NeoML::ISimdMathEngine& SimdMathEngine(); diff --git a/NeoMathEngine/test/AvxTestDesktop/main.cpp b/NeoMathEngine/test/AvxTestDesktop/main.cpp new file mode 100644 index 000000000..0cd08ec35 --- /dev/null +++ b/NeoMathEngine/test/AvxTestDesktop/main.cpp @@ -0,0 +1,27 @@ +/* Copyright © 2017-2022 ABBYY Production LLC + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. +--------------------------------------------------------------------------------------------------------------*/ + +#include + +#include "common.h" + +int main( int argc, char* argv[] ) +{ + ::testing::InitGoogleTest( &argc, argv ); + + int result = RUN_ALL_TESTS(); + + return result; +} diff --git a/NeoMathEngine/test/src/common/TestParams.cpp b/NeoMathEngine/test/src/common/TestParams.cpp index 31acf3eb1..1bd189862 100644 --- a/NeoMathEngine/test/src/common/TestParams.cpp +++ b/NeoMathEngine/test/src/common/TestParams.cpp @@ -147,10 +147,11 @@ static std::string trim( const std::string& str ) return std::string( res + leftSize, static_cast( len - leftSize - rightSize ) ); } -CTestParams::CTestParams( const std::string& string ) +CTestParams::CTestParams( const std::string& paramText, const std::string& testName ) : + testName( testName ) { std::vector keyValue; - splitStringsByDelimiter( keyValue, string, ";" ); + splitStringsByDelimiter( keyValue, paramText, ";" ); for( size_t i = 0; i < keyValue.size(); ++i ) { const int equalSign = find( keyValue[i].c_str(), '=' ); @@ -163,22 +164,11 @@ CTestParams::CTestParams( const std::string& string ) } } -CTestParams::CTestParams( const CTestParams& other ) : - flags( other.flags ) -{ -} - std::string CTestParams::GetStrValue( const std::string& key ) const { return flags.find( key )->second; } -CTestParams& CTestParams::operator=( const CTestParams& other ) -{ - flags = other.flags; - return *this; -} - CInterval CTestParams::GetInterval( const std::string& key ) const { return parseInterval( flags.find( key )->second ); diff --git a/NeoMathEngine/test/src/common/TestParams.h b/NeoMathEngine/test/src/common/TestParams.h index 708ffedb5..9cb8ffe30 100644 --- a/NeoMathEngine/test/src/common/TestParams.h +++ b/NeoMathEngine/test/src/common/TestParams.h @@ -31,10 +31,12 @@ struct CInterval { class CTestParams { public: - explicit CTestParams( const std::string& str ); - CTestParams( const CTestParams& other ); + explicit CTestParams( const std::string& str, const std::string& testName = ""); + CTestParams( const CTestParams& other ) = default; - CTestParams& operator=( const CTestParams& other ); + CTestParams& operator=( const CTestParams& other ) = default; + + const std::string& Name() const { return testName; } std::string GetStrValue( const std::string& key ) const; @@ -49,6 +51,7 @@ class CTestParams { friend ::std::ostream& operator<<( ::std::ostream& os, const CTestParams& params ); private: + std::string testName; std::unordered_map flags; CInterval parseInterval( const std::string& stringValue ) const;