diff --git a/include/onnxruntime/core/session/onnxruntime_c_api.h b/include/onnxruntime/core/session/onnxruntime_c_api.h index c02f6cb300..9e89923c58 100644 --- a/include/onnxruntime/core/session/onnxruntime_c_api.h +++ b/include/onnxruntime/core/session/onnxruntime_c_api.h @@ -7,7 +7,6 @@ #include #include "onnxruntime_session_options_config_keys.h" - // This value is used in structures passed to ORT so that a newer version of ORT will still work with them #define ORT_API_VERSION 5 @@ -261,16 +260,17 @@ typedef enum OrtMemType { } OrtMemType; typedef enum OrtCudnnConvAlgoSearch { - EXHAUSTIVE, // expensive exhaustive benchmarking using cudnnFindConvolutionForwardAlgorithmEx - HEURISTIC, // lightweight heuristic based search using cudnnGetConvolutionForwardAlgorithm_v7 - DEFAULT, // default algorithm using CUDNN_CONVOLUTION_FWD_ALGO_IMPLICIT_PRECOMP_GEMM + EXHAUSTIVE, // expensive exhaustive benchmarking using cudnnFindConvolutionForwardAlgorithmEx + HEURISTIC, // lightweight heuristic based search using cudnnGetConvolutionForwardAlgorithm_v7 + DEFAULT, // default algorithm using CUDNN_CONVOLUTION_FWD_ALGO_IMPLICIT_PRECOMP_GEMM } OrtCudnnConvAlgoSearch; typedef struct OrtCUDAProviderOptions { - int device_id; // cuda device with id=0 as default device. + int device_id; // cuda device with id=0 as default device. OrtCudnnConvAlgoSearch cudnn_conv_algo_search; // cudnn conv algo search option - size_t cuda_mem_limit; // default cuda memory limitation to maximum finite value of size_t. - int arena_extend_strategy; // default area extend strategy to KNextPowerOfTwo. + size_t cuda_mem_limit; // default cuda memory limitation to maximum finite value of size_t. + int arena_extend_strategy; // default area extend strategy to KNextPowerOfTwo. + int do_copy_in_default_stream; } OrtCUDAProviderOptions; struct OrtApi; @@ -1046,8 +1046,7 @@ struct OrtApi { */ ORT_API2_STATUS(SessionGetProfilingStartTimeNs, _In_ const OrtSession* sess, _Outptr_ uint64_t* out); - -/** + /** * Use this API to configure the global thread pool options to be used in the call to CreateEnvWithGlobalThreadPools. * A value of 0 means ORT will pick the default. * A value of 1 means the invoking thread will be used; no threads will be created in the thread pool. @@ -1077,7 +1076,7 @@ struct OrtApi { */ ORT_API2_STATUS(AddInitializer, _Inout_ OrtSessionOptions* options, _In_z_ const char* name, _In_ const OrtValue* val); - + /** * Creates a custom environment with global threadpools and logger that will be shared across sessions. * Use this in conjunction with DisablePerSessionThreads API or else the session will use @@ -1088,8 +1087,8 @@ struct OrtApi { ORT_API2_STATUS(CreateEnvWithCustomLoggerAndGlobalThreadPools, OrtLoggingFunction logging_function, _In_opt_ void* logger_param, OrtLoggingLevel logging_level, _In_ const char* logid, _In_ const struct OrtThreadingOptions* tp_options, _Outptr_ OrtEnv** out); - #ifdef USE_CUDA - /** +#ifdef USE_CUDA + /** * Append CUDA execution provider */ ORT_API2_STATUS(OrtSessionOptionsAppendExecutionProvider_CUDA, diff --git a/onnxruntime/core/providers/cuda/cuda_execution_provider.cc b/onnxruntime/core/providers/cuda/cuda_execution_provider.cc index 2a8a92b9be..8c9d554984 100644 --- a/onnxruntime/core/providers/cuda/cuda_execution_provider.cc +++ b/onnxruntime/core/providers/cuda/cuda_execution_provider.cc @@ -121,7 +121,8 @@ CUDAExecutionProvider::CUDAExecutionProvider(const CUDAExecutionProviderInfo& in device_id_(info.device_id), cuda_mem_limit_(info.cuda_mem_limit), arena_extend_strategy_(info.arena_extend_strategy), - cudnn_conv_algo_(info.cudnn_conv_algo) { + cudnn_conv_algo_(info.cudnn_conv_algo), + do_copy_in_default_stream_(info.do_copy_in_default_stream) { CUDA_CALL_THROW(cudaSetDevice(device_id_)); // must wait GPU idle, otherwise cudaGetDeviceProperties might fail @@ -1811,7 +1812,7 @@ static bool CastNeedFallbackToCPU(const onnxruntime::Node& node) { } std::unique_ptr CUDAExecutionProvider::GetDataTransfer() const { - return onnxruntime::make_unique(); + return onnxruntime::make_unique(do_copy_in_default_stream_); } std::vector> diff --git a/onnxruntime/core/providers/cuda/cuda_execution_provider.h b/onnxruntime/core/providers/cuda/cuda_execution_provider.h index 139902c731..2f915477af 100644 --- a/onnxruntime/core/providers/cuda/cuda_execution_provider.h +++ b/onnxruntime/core/providers/cuda/cuda_execution_provider.h @@ -25,6 +25,7 @@ struct CUDAExecutionProviderInfo { size_t cuda_mem_limit{std::numeric_limits::max()}; ArenaExtendStrategy arena_extend_strategy{ArenaExtendStrategy::kNextPowerOfTwo}; OrtCudnnConvAlgoSearch cudnn_conv_algo{OrtCudnnConvAlgoSearch::EXHAUSTIVE}; + bool do_copy_in_default_stream{true}; }; // Logical device representation. @@ -87,6 +88,7 @@ private: size_t cuda_mem_limit_; ArenaExtendStrategy arena_extend_strategy_; int cudnn_conv_algo_; + bool do_copy_in_default_stream_; struct DeferredReleaseCPUPtrs { bool recorded = false; diff --git a/onnxruntime/core/providers/cuda/cuda_provider_factory.cc b/onnxruntime/core/providers/cuda/cuda_provider_factory.cc index 7518ed09dd..c12ae89cf0 100644 --- a/onnxruntime/core/providers/cuda/cuda_provider_factory.cc +++ b/onnxruntime/core/providers/cuda/cuda_provider_factory.cc @@ -18,11 +18,13 @@ struct CUDAProviderFactory : IExecutionProviderFactory { CUDAProviderFactory(OrtDevice::DeviceId device_id, size_t cuda_mem_limit = std::numeric_limits::max(), ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo, - OrtCudnnConvAlgoSearch cudnn_conv_algo_search = OrtCudnnConvAlgoSearch::EXHAUSTIVE) + OrtCudnnConvAlgoSearch cudnn_conv_algo_search = OrtCudnnConvAlgoSearch::EXHAUSTIVE, + bool do_copy_in_default_stream = true) : device_id_(device_id), cuda_mem_limit_(cuda_mem_limit), arena_extend_strategy_(arena_extend_strategy), - cudnn_conv_algo_search_(cudnn_conv_algo_search) {} + cudnn_conv_algo_search_(cudnn_conv_algo_search), + do_copy_in_default_stream_(do_copy_in_default_stream) {} ~CUDAProviderFactory() override {} std::unique_ptr CreateProvider() override; @@ -32,6 +34,7 @@ struct CUDAProviderFactory : IExecutionProviderFactory { size_t cuda_mem_limit_; ArenaExtendStrategy arena_extend_strategy_; OrtCudnnConvAlgoSearch cudnn_conv_algo_search_; + bool do_copy_in_default_stream_; }; std::unique_ptr CUDAProviderFactory::CreateProvider() { @@ -40,14 +43,16 @@ std::unique_ptr CUDAProviderFactory::CreateProvider() { info.cuda_mem_limit = cuda_mem_limit_; info.arena_extend_strategy = arena_extend_strategy_; info.cudnn_conv_algo = cudnn_conv_algo_search_; + info.do_copy_in_default_stream = do_copy_in_default_stream_; return onnxruntime::make_unique(info); } std::shared_ptr CreateExecutionProviderFactory_CUDA(OrtDevice::DeviceId device_id, OrtCudnnConvAlgoSearch cudnn_conv_algo_search = OrtCudnnConvAlgoSearch::EXHAUSTIVE, size_t cuda_mem_limit = std::numeric_limits::max(), - ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo) { - return std::make_shared(device_id, cuda_mem_limit, arena_extend_strategy, cudnn_conv_algo_search); + ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo, + bool do_copy_in_default_stream = true) { + return std::make_shared(device_id, cuda_mem_limit, arena_extend_strategy, cudnn_conv_algo_search, do_copy_in_default_stream); } } // namespace onnxruntime @@ -61,6 +66,7 @@ ORT_API_STATUS_IMPL(OrtApis::OrtSessionOptionsAppendExecutionProvider_CUDA, _In_ OrtSessionOptions* options, _In_ OrtCUDAProviderOptions* cuda_options) { options->provider_factories.push_back(onnxruntime::CreateExecutionProviderFactory_CUDA(static_cast(cuda_options->device_id), cuda_options->cudnn_conv_algo_search, cuda_options->cuda_mem_limit, - static_cast(cuda_options->arena_extend_strategy))); + static_cast(cuda_options->arena_extend_strategy), + cuda_options->do_copy_in_default_stream)); return nullptr; } diff --git a/onnxruntime/core/providers/cuda/gpu_data_transfer.cc b/onnxruntime/core/providers/cuda/gpu_data_transfer.cc index 5b7603277c..6618688087 100644 --- a/onnxruntime/core/providers/cuda/gpu_data_transfer.cc +++ b/onnxruntime/core/providers/cuda/gpu_data_transfer.cc @@ -4,22 +4,35 @@ #include "core/providers/cuda/gpu_data_transfer.h" #include "cuda_common.h" +// use default stream for copy for now, to avoid racing in BFC arena as in issue #4829 +// note this may cause some models to run slower if there are ops running on CPU +// so we leave it as optional, in case user need the previous behavior +// a full fix to BFC arena is being looked at, and once it's in, we can revert this change namespace onnxruntime { -GPUDataTransfer::GPUDataTransfer() { +GPUDataTransfer::GPUDataTransfer(bool do_copy_in_default_stream) { // create streams, default is nullptr streams_[kCudaStreamDefault] = nullptr; - CUDA_CALL_THROW(cudaStreamCreateWithFlags(&streams_[kCudaStreamCopyIn], cudaStreamNonBlocking)); - CUDA_CALL_THROW(cudaStreamCreateWithFlags(&streams_[kCudaStreamCopyOut], cudaStreamNonBlocking)); + if (do_copy_in_default_stream) { + streams_[kCudaStreamCopyIn] = nullptr; + streams_[kCudaStreamCopyOut] = nullptr; + } else { + CUDA_CALL_THROW(cudaStreamCreateWithFlags(&streams_[kCudaStreamCopyIn], cudaStreamNonBlocking)); + CUDA_CALL_THROW(cudaStreamCreateWithFlags(&streams_[kCudaStreamCopyOut], cudaStreamNonBlocking)); + } } GPUDataTransfer::~GPUDataTransfer() { - CUDA_CALL(cudaStreamDestroy(streams_[kCudaStreamCopyIn])); - CUDA_CALL(cudaStreamDestroy(streams_[kCudaStreamCopyOut])); + if (streams_[kCudaStreamCopyIn] != nullptr) { + CUDA_CALL(cudaStreamDestroy(streams_[kCudaStreamCopyIn])); + } + if (streams_[kCudaStreamCopyOut] != nullptr) { + CUDA_CALL(cudaStreamDestroy(streams_[kCudaStreamCopyOut])); + } } bool GPUDataTransfer::CanCopy(const OrtDevice& src_device, const OrtDevice& dst_device) const { - return src_device.Type() == OrtDevice::GPU || src_device.MemType() == OrtDevice::MemType::CUDA_PINNED - || dst_device.Type() == OrtDevice::GPU || dst_device.MemType() == OrtDevice::MemType::CUDA_PINNED; + return src_device.Type() == OrtDevice::GPU || src_device.MemType() == OrtDevice::MemType::CUDA_PINNED || + dst_device.Type() == OrtDevice::GPU || dst_device.MemType() == OrtDevice::MemType::CUDA_PINNED; } common::Status GPUDataTransfer::CopyTensor(const Tensor& src, Tensor& dst, int exec_queue_id) const { diff --git a/onnxruntime/core/providers/cuda/gpu_data_transfer.h b/onnxruntime/core/providers/cuda/gpu_data_transfer.h index 50ba91c441..055e2a90fd 100644 --- a/onnxruntime/core/providers/cuda/gpu_data_transfer.h +++ b/onnxruntime/core/providers/cuda/gpu_data_transfer.h @@ -17,7 +17,7 @@ enum CUDAStreamType : int { class GPUDataTransfer : public IDataTransfer { public: - GPUDataTransfer(); + GPUDataTransfer(bool do_copy_in_default_stream = true); ~GPUDataTransfer(); bool CanCopy(const OrtDevice& src_device, const OrtDevice& dst_device) const override; diff --git a/onnxruntime/python/onnxruntime_pybind_state.cc b/onnxruntime/python/onnxruntime_pybind_state.cc index 6c8de4b33e..6499389eee 100644 --- a/onnxruntime/python/onnxruntime_pybind_state.cc +++ b/onnxruntime/python/onnxruntime_pybind_state.cc @@ -145,6 +145,8 @@ OrtDevice::DeviceId cuda_device_id = 0; OrtCudnnConvAlgoSearch cudnn_conv_algo_search = OrtCudnnConvAlgoSearch::EXHAUSTIVE; size_t cuda_mem_limit = std::numeric_limits::max(); onnxruntime::ArenaExtendStrategy arena_extend_strategy = onnxruntime::ArenaExtendStrategy::kNextPowerOfTwo; +bool do_copy_in_default_stream = true; + #endif #ifdef USE_TENSORRT #include "core/providers/tensorrt/tensorrt_provider_factory.h" @@ -189,7 +191,8 @@ std::shared_ptr CreateExecutionProviderFactory_CPU(in std::shared_ptr CreateExecutionProviderFactory_CUDA(OrtDevice::DeviceId device_id, OrtCudnnConvAlgoSearch cudnn_conv_algo_search, size_t cuda_mem_limit, - onnxruntime::ArenaExtendStrategy arena_extend_strategy); + onnxruntime::ArenaExtendStrategy arena_extend_strategy, + bool do_copy_in_default_stream); std::shared_ptr CreateExecutionProviderFactory_Tensorrt(int device_id); std::shared_ptr CreateExecutionProviderFactory_MIGraphX(int device_id); std::shared_ptr CreateExecutionProviderFactory_Dnnl(int use_arena); @@ -626,13 +629,15 @@ static void RegisterExecutionProviders(InferenceSession* sess, const std::vector sess, *onnxruntime::CreateExecutionProviderFactory_CUDA(cuda_provider_options.device_id, cuda_provider_options.cudnn_conv_algo, cuda_provider_options.cuda_mem_limit, - cuda_provider_options.arena_extend_strategy)); + cuda_provider_options.arena_extend_strategy, + cuda_provider_options.do_copy_in_default_stream)); } else { RegisterExecutionProvider( sess, *onnxruntime::CreateExecutionProviderFactory_CUDA(cuda_device_id, cudnn_conv_algo_search, cuda_mem_limit, - arena_extend_strategy)); + arena_extend_strategy, + do_copy_in_default_stream)); } #endif } else if (type == kDnnlExecutionProvider) { @@ -853,7 +858,7 @@ void addGlobalMethods(py::module& m, const Environment& env) { std::vector> factories = { onnxruntime::CreateExecutionProviderFactory_CPU(0), #ifdef USE_CUDA - onnxruntime::CreateExecutionProviderFactory_CUDA(cuda_device_id, cudnn_conv_algo_search, cuda_mem_limit, arena_extend_strategy), + onnxruntime::CreateExecutionProviderFactory_CUDA(cuda_device_id, cudnn_conv_algo_search, cuda_mem_limit, arena_extend_strategy, do_copy_in_default_stream), #endif #ifdef USE_DNNL onnxruntime::CreateExecutionProviderFactory_Dnnl(1), @@ -909,6 +914,7 @@ void addGlobalMethods(py::module& m, const Environment& env) { m.def("set_cudnn_conv_algo_search", [](const OrtCudnnConvAlgoSearch algo) { cudnn_conv_algo_search = algo; }); m.def("set_cuda_mem_limit", [](const int64_t limit) { cuda_mem_limit = static_cast(limit); }); m.def("set_arena_extend_strategy", [](const onnxruntime::ArenaExtendStrategy strategy) { arena_extend_strategy = strategy; }); + m.def("set_do_copy_in_default_stream", [](const bool use_single_stream) { do_copy_in_default_stream = use_single_stream; }); #endif } diff --git a/onnxruntime/test/onnx/main.cc b/onnxruntime/test/onnx/main.cc index 07a8436c13..43f253974d 100644 --- a/onnxruntime/test/onnx/main.cc +++ b/onnxruntime/test/onnx/main.cc @@ -336,6 +336,7 @@ int real_main(int argc, char* argv[], Ort::Env& env) { OrtCudnnConvAlgoSearch::EXHAUSTIVE, std::numeric_limits::max(), 0, + true }; Ort::ThrowOnError(sf.OrtSessionOptionsAppendExecutionProvider_CUDA(sf, &cuda_options)); #else diff --git a/onnxruntime/test/perftest/command_args_parser.cc b/onnxruntime/test/perftest/command_args_parser.cc index cc8daff820..50f833a9e2 100644 --- a/onnxruntime/test/perftest/command_args_parser.cc +++ b/onnxruntime/test/perftest/command_args_parser.cc @@ -49,12 +49,13 @@ namespace perftest { "\t\tPlease see onnxruntime_c_api.h (enum GraphOptimizationLevel) for the full list of all optimization levels. \n" "\t-u [optimized_model_path]: Specify the optimized model path for saving.\n" "\t-d [cudnn_conv_algorithm]: Specify CUDNN convolution algothrithms: 0(benchmark), 1(heuristic), 2(default). \n" + "\t-q: [CUDA only] use separate stream for copy. \n" "\t-h: help\n"); } /*static*/ bool CommandLineParser::ParseArguments(PerformanceTestConfig& test_config, int argc, ORTCHAR_T* argv[]) { int ch; - while ((ch = getopt(argc, argv, ORT_TSTR("b:m:e:r:t:p:x:y:c:d:o:u:AMPIvhs"))) != -1) { + while ((ch = getopt(argc, argv, ORT_TSTR("b:m:e:r:t:p:x:y:c:d:o:u:AMPIvhsq"))) != -1) { switch (ch) { case 'm': if (!CompareCString(optarg, ORT_TSTR("duration"))) { @@ -181,6 +182,9 @@ namespace perftest { case 'd': test_config.run_config.cudnn_conv_algo = static_cast(OrtStrtol(optarg, nullptr)); break; + case 'q': + test_config.run_config.do_cuda_copy_in_separate_stream = true; + break; case '?': case 'h': default: diff --git a/onnxruntime/test/perftest/ort_test_session.cc b/onnxruntime/test/perftest/ort_test_session.cc index e399bc359f..73ba1c30c4 100644 --- a/onnxruntime/test/perftest/ort_test_session.cc +++ b/onnxruntime/test/perftest/ort_test_session.cc @@ -48,9 +48,10 @@ OnnxRuntimeTestSession::OnnxRuntimeTestSession(Ort::Env& env, std::random_device #ifdef USE_CUDA OrtCUDAProviderOptions cuda_options{ 0, - OrtCudnnConvAlgoSearch::EXHAUSTIVE, + static_cast(performance_test_config.run_config.cudnn_conv_algo), std::numeric_limits::max(), 0, + !performance_test_config.run_config.do_cuda_copy_in_separate_stream }; Ort::ThrowOnError(session_options.OrtSessionOptionsAppendExecutionProvider_CUDA(session_options, &cuda_options)); #else diff --git a/onnxruntime/test/perftest/test_configuration.h b/onnxruntime/test/perftest/test_configuration.h index cd5b242962..2fb65dcf80 100644 --- a/onnxruntime/test/perftest/test_configuration.h +++ b/onnxruntime/test/perftest/test_configuration.h @@ -51,6 +51,7 @@ struct RunConfig { GraphOptimizationLevel optimization_level{ORT_ENABLE_ALL}; std::basic_string optimized_model_path; int cudnn_conv_algo{0}; + bool do_cuda_copy_in_separate_stream{false}; }; struct PerformanceTestConfig { diff --git a/onnxruntime/test/python/onnxruntime_test_python.py b/onnxruntime/test/python/onnxruntime_test_python.py index f1526a7e70..1d81da6b56 100644 --- a/onnxruntime/test/python/onnxruntime_test_python.py +++ b/onnxruntime/test/python/onnxruntime_test_python.py @@ -765,6 +765,24 @@ class TestInferenceSession(unittest.TestCase): # The constructed OrtValue should still be valid after being used in a session self.assertTrue(np.array_equal(ortvalue2.numpy(), numpy_arr_input)) + def testRunModelWithCudaCopyStream(self): + available_providers = onnxrt.get_available_providers() + + if (not 'CUDAExecutionProvider' in available_providers): + print("Skipping testRunModelWithCudaCopyStream when CUDA is not available") + else: + # adapted from issue #4829 for a race condition when copy is not on default stream + # note: + # 1. if there are intermittent failure in this test, something is wrong + # 2. it's easier to repro on slower GPU (like M60, Geforce 1070) + + # to repro #4829, uncomment the line below to run copy in a separate stream + #onnxrt.capi._pybind_state.set_do_copy_in_default_stream(False) + + session = onnxrt.InferenceSession(get_name("issue4829.onnx")) + shape = np.array([2,2], dtype=np.int64) + for iteration in range(100000): + result = session.run(output_names=['output'], input_feed={'shape': shape}) if __name__ == '__main__': unittest.main() diff --git a/onnxruntime/test/testdata/issue4829.onnx b/onnxruntime/test/testdata/issue4829.onnx new file mode 100755 index 0000000000..93090cbffb Binary files /dev/null and b/onnxruntime/test/testdata/issue4829.onnx differ diff --git a/onnxruntime/test/util/default_providers.cc b/onnxruntime/test/util/default_providers.cc index 5e61fa6a40..324c3dcd48 100644 --- a/onnxruntime/test/util/default_providers.cc +++ b/onnxruntime/test/util/default_providers.cc @@ -13,7 +13,8 @@ std::shared_ptr CreateExecutionProviderFactory_CPU(in std::shared_ptr CreateExecutionProviderFactory_CUDA(OrtDevice::DeviceId device_id, OrtCudnnConvAlgoSearch cudnn_conv_algo = OrtCudnnConvAlgoSearch::EXHAUSTIVE, size_t cuda_mem_limit = std::numeric_limits::max(), - ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo); + ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo, + bool do_copy_in_default_stream = true); std::shared_ptr CreateExecutionProviderFactory_Dnnl(int use_arena); std::shared_ptr CreateExecutionProviderFactory_NGraph(const char* ng_backend_type); std::shared_ptr CreateExecutionProviderFactory_OpenVINO(const char* device_type, bool enable_vpu_fast_compile, const char* device_id); diff --git a/orttraining/orttraining/models/bert/main.cc b/orttraining/orttraining/models/bert/main.cc index a1c46e244c..a31e613178 100644 --- a/orttraining/orttraining/models/bert/main.cc +++ b/orttraining/orttraining/models/bert/main.cc @@ -22,7 +22,8 @@ namespace onnxruntime { std::shared_ptr CreateExecutionProviderFactory_CUDA(OrtDevice::DeviceId device_id, OrtCudnnConvAlgoSearch cudnn_conv_algo_search = OrtCudnnConvAlgoSearch::EXHAUSTIVE, size_t cuda_mem_limit = std::numeric_limits::max(), - onnxruntime::ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo); + onnxruntime::ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo, + bool do_copy_in_default_stream = true); } using namespace onnxruntime; diff --git a/orttraining/orttraining/models/gpt2/main.cc b/orttraining/orttraining/models/gpt2/main.cc index c415bf9578..70610cf1d3 100644 --- a/orttraining/orttraining/models/gpt2/main.cc +++ b/orttraining/orttraining/models/gpt2/main.cc @@ -21,7 +21,8 @@ namespace onnxruntime { std::shared_ptr CreateExecutionProviderFactory_CUDA(OrtDevice::DeviceId device_id, OrtCudnnConvAlgoSearch cudnn_conv_algo_search = OrtCudnnConvAlgoSearch::EXHAUSTIVE, size_t cuda_mem_limit = std::numeric_limits::max(), - onnxruntime::ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo); + onnxruntime::ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo, + bool do_copy_in_default_stream = true); } using namespace onnxruntime; diff --git a/orttraining/orttraining/models/mnist/main.cc b/orttraining/orttraining/models/mnist/main.cc index 52e02df82d..5e9c0502cb 100644 --- a/orttraining/orttraining/models/mnist/main.cc +++ b/orttraining/orttraining/models/mnist/main.cc @@ -21,7 +21,8 @@ namespace onnxruntime { std::shared_ptr CreateExecutionProviderFactory_CUDA(OrtDevice::DeviceId device_id, OrtCudnnConvAlgoSearch cudnn_conv_algo_search = OrtCudnnConvAlgoSearch::EXHAUSTIVE, size_t cuda_mem_limit = std::numeric_limits::max(), - onnxruntime::ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo); + onnxruntime::ArenaExtendStrategy arena_extend_strategy = ArenaExtendStrategy::kNextPowerOfTwo, + bool do_copy_in_default_stream = true); } using namespace onnxruntime;