Describe the issue PR <a class="issue-link js-issue-link" data-err

CC: <a class="user-mention notranslate" data-hovercard-type="user" data-hovercard-url=

<a class="user-mention notranslate" data-hovercard-type="user" data-hovercard-url="/us

[Performance] Disable stream synchronization at the end of CUDA Graphs replay about onnxruntime HOT 2 OPEN

WolframRhodium commented on September 26, 2024

[Performance] Disable stream synchronization at the end of CUDA Graphs replay

from onnxruntime.

Comments (2)

hariharans29 commented on September 26, 2024

CC: @gedoensmax

from onnxruntime.

gedoensmax commented on September 26, 2024

@hariharans29 If i understand correctly we can simply remove the sync in the GraphManager. The sync should be done in the EP iteself here:

onnxruntime/onnxruntime/core/providers/cuda/cuda_execution_provider.cc

Lines 434 to 447 in c47a6ce

 if (IsGraphCaptureEnabled() && !GetPerThreadContext().IsGraphCaptured(cuda_graph_annotation_id)) { 

 if (GetPerThreadContext().IsGraphCaptureAllowed(cuda_graph_annotation_id)) { 

 GetPerThreadContext().CaptureEnd(cuda_graph_annotation_id); 

 // CUDA work issued to a capturing stream doesn’t actually run on the GPU, 

 // so run the captured graph here to actually execute the work. 

 ORT_RETURN_IF_ERROR(GetPerThreadContext().ReplayGraph(cuda_graph_annotation_id)); 

 } else { 

 GetPerThreadContext().IncrementRegularRunCountBeforeGraphCapture(cuda_graph_annotation_id); 

 } 

 } 

 if (sync_stream) { 

 CUDA_RETURN_IF_ERROR(cudaStreamSynchronize(static_cast<cudaStream_t>(stream_))); 

 }

from onnxruntime.

[Performance] Disable stream synchronization at the end of CUDA Graphs replay about onnxruntime HOT 2 OPEN

Comments (2)

Related Issues (20)

Recommend Projects

React

Vue.js

Typescript

TensorFlow

Django

Laravel

D3

Recommend Topics

javascript

web

server

Machine learning

Visualization

Game

Recommend Org

Facebook

Microsoft

Google

Alibaba

D3

Tencent

	if (IsGraphCaptureEnabled() && !GetPerThreadContext().IsGraphCaptured(cuda_graph_annotation_id)) {
	if (GetPerThreadContext().IsGraphCaptureAllowed(cuda_graph_annotation_id)) {
	GetPerThreadContext().CaptureEnd(cuda_graph_annotation_id);
	// CUDA work issued to a capturing stream doesn’t actually run on the GPU,
	// so run the captured graph here to actually execute the work.
	ORT_RETURN_IF_ERROR(GetPerThreadContext().ReplayGraph(cuda_graph_annotation_id));
	} else {
	GetPerThreadContext().IncrementRegularRunCountBeforeGraphCapture(cuda_graph_annotation_id);
	}
	}

	if (sync_stream) {
	CUDA_RETURN_IF_ERROR(cudaStreamSynchronize(static_cast<cudaStream_t>(stream_)));
	}