[SYCLomatic] Fix migration of cudaStreamDefault and cudaStreamNonBlocking#475
[SYCLomatic] Fix migration of cudaStreamDefault and cudaStreamNonBlocking#4750x12CC merged 5 commits intooneapi-src:SYCLomaticfrom
Conversation
| Begin, SM, DpctGlobalInfo::getContext().getLangOpts()); | ||
| emplaceTransformation(new ReplaceText(Begin, Length, "0")); | ||
| } | ||
| } |
There was a problem hiding this comment.
cudaStreamDefault is migrated to 0 if there has not been a replacement migrating it to &dpct::get_default_queue().
| } | ||
|
|
||
| REGISTER_RULE(PreDefinedStreamHandleRule, PassKind::PK_Migration) | ||
|
|
There was a problem hiding this comment.
This migration rule was migrating the macros cudaStreamDefault, cudaStreamLegacy, and cudaStreamPerThread to dpct::get_default_queue(). This is correct for cudaStreamLegacy and cudaStreamPerThread, as these macros' definition expand to a value of cudaStream_t. This is incorrect for cudaStreamDefault in a edge case where you assign cudaStreamDefault to an int. (See the old macro_test.cu - the checked output is not valid C++ because you cannot assign a sycl::queue * to an int). For the migration of cudaStreamLegacy and cudaStreamPerThread, this migration behavior is capture by CudaStreamCastRule, so it can be removed.
(Note: the intended purpose of the cudaStreamDefault macro is as a flag used with cudaStreamCreateWithFlags . The name of cudaStreamDefault makes it seem like you can use it as a the default cudaStream_t, which perhaps coincidently, you can, because cudaStreamDefault expands to 0, which is the same value used to denote the default stream with the default stream semantics of the cuda api.)
jzc
force-pushed
the
cudastream-flags
branch
2 times, most recently
from
23de87f to
c1efdae
Compare
ShengchenJ
left a comment
ShengchenJ
left a comment
There was a problem hiding this comment.
The comment is out of date
// If PredefinedStreamName is used with concatenated macro token,
// detect the previous macro expansion
please also clean it.
…api-src#475) Signed-off-by: Tang, Jiajun jiajun.tang@intel.com
5 participants
Signed-off-by: Cai, Justin justin.cai@intel.com