diff --git a/clang/lib/DPCT/ASTTraversal.cpp b/clang/lib/DPCT/ASTTraversal.cpp index 6224f1183290..0ba542eeae3e 100644 --- a/clang/lib/DPCT/ASTTraversal.cpp +++ b/clang/lib/DPCT/ASTTraversal.cpp @@ -7983,7 +7983,8 @@ void StreamAPICallRule::runRule(const MatchFinder::MatchResult &Result) { } else if (FuncName == "cudaStreamWaitEvent" || FuncName == "cuStreamWaitEvent") { std::string ReplStr; - auto StmtStr1 = getStmtSpelling(CE->getArg(1)); + ExprAnalysis EA(CE->getArg(1)); + std::string StmtStr1 = EA.getReplacedString(); if (!DpctGlobalInfo::useEnqueueBarrier()) { // ext_oneapi_submit_barrier is specified in the value of option // --no-dpcpp-extensions. diff --git a/clang/test/dpct/cuda-stream-api.cu b/clang/test/dpct/cuda-stream-api.cu index 67d9e93022fe..a9e0cf9385eb 100644 --- a/clang/test/dpct/cuda-stream-api.cu +++ b/clang/test/dpct/cuda-stream-api.cu @@ -211,8 +211,8 @@ static void func() MY_ERROR_CHECKER(cudaStreamAttachMemAsync(s0, nullptr)); cudaEvent_t e; - // CHECK: s0->ext_oneapi_submit_barrier({*e}); - cudaStreamWaitEvent(s0, e, 0); + // CHECK: s0->ext_oneapi_submit_barrier({*(dpct::event_ptr)e}); + cudaStreamWaitEvent(s0, (cudaEvent_t)e, 0); // CHECK: /* // CHECK-NEXT: DPCT1026:{{[0-9]+}}: The call to cudaStreamQuery was removed because SYCL currently does not support query operations on queues.