ptx_parser/parser/instruction/
cp_async_wait_group.rs

1//! Original PTX specification:
2//!
3//! cp.async.wait_group N;
4//! cp.async.wait_all ;
5
6#![allow(unused)]
7
8use crate::lexer::PtxToken;
9use crate::parser::{PtxParseError, PtxParser, PtxTokenStream, Span};
10use crate::r#type::common::*;
11
12pub mod section_0 {
13    use super::*;
14    use crate::r#type::instruction::cp_async_wait_group::section_0::*;
15
16    impl PtxParser for CpAsyncWaitGroup {
17        fn parse(stream: &mut PtxTokenStream) -> Result<Self, PtxParseError> {
18            stream.expect_string("cp")?;
19            stream.expect_string(".async")?;
20            let async_ = ();
21            stream.expect_complete()?;
22            stream.expect_string(".wait_group")?;
23            let wait_group = ();
24            stream.expect_complete()?;
25            let n = GeneralOperand::parse(stream)?;
26            stream.expect_complete()?;
27            stream.expect_complete()?;
28            stream.expect(&PtxToken::Semicolon)?;
29            Ok(CpAsyncWaitGroup {
30                async_,
31                wait_group,
32                n,
33            })
34        }
35    }
36
37
38    impl PtxParser for CpAsyncWaitAll {
39        fn parse(stream: &mut PtxTokenStream) -> Result<Self, PtxParseError> {
40            stream.expect_string("cp")?;
41            stream.expect_string(".async")?;
42            let async_ = ();
43            stream.expect_complete()?;
44            stream.expect_string(".wait_all")?;
45            let wait_all = ();
46            stream.expect_complete()?;
47            stream.expect_complete()?;
48            stream.expect(&PtxToken::Semicolon)?;
49            Ok(CpAsyncWaitAll {
50                async_,
51                wait_all,
52            })
53        }
54    }
55
56
57}
58
ptx_parser/parser/instruction/cp_async_wait_group.rs

ptx_parser/parser/instruction/
cp_async_wait_group.rs
