vyre 0.4.0

GPU compute intermediate representation with a standard operation library
Documentation 
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97
98
99
100
101
102
103
104
105
106
107
108
109
110
111
112
113
114
115
116
117
118
119
120
121
122
123
124
125
126
127
128
129
130
131

//! Catalog entry for `detect_xss`.

use crate::ops::security_detection::detector_support::{attacks, DetectionError};

/// Embedded operation spec formerly stored in metadata/spec.toml.
pub const SPEC_TOML: &str = r#"schema_version = 1
id = "security_detection.detect_xss"
archetype = "rule-bytes-to-bool"
display_name = "Detect XSS"
summary = "Returns true for script tags, JavaScript URLs, or executable event handlers."
category = "C"

[intrinsic]
wgsl = "security_detection_detect_xss"

[signature]
inputs = ["Bytes"]
output = "Bool"

laws = []
equivalence_classes = ["script_tag", "event_handler", "plain_html", "t47_cap"]
workgroup_size = [64, 1, 1]
tags = ["security-detection", "xss", "owasp", "t47"]
fixtures_dir = "fixtures/"
"#;

/// Embedded reference vectors formerly stored in fixtures/reference-vectors.toml.
pub const REFERENCE_VECTORS_TOML: &str = r#"[[case]]
name = "positive_script"
input = "<script>alert(1)</script>"
expected = true

[[case]]
name = "negative_plain_html"
input = "<p>safe markup</p>"
expected = false
"#;

/// WGSL lowering source for this detector.
pub mod lowering {
    /// Return the detector-specific WGSL source.
    #[must_use]
    pub const fn source() -> &'static str {
        r#"struct Params {
    input_len: u32,
    _pad0: u32,
    _pad1: u32,
    _pad2: u32,
}

@group(0) @binding(0) var<storage, read> input: array<u32>;
@group(0) @binding(1) var<storage, read_write> output: array<u32>;
@group(0) @binding(2) var<uniform> params: Params;

fn lower_ascii(byte: u32) -> u32 {
    if (byte >= 65u && byte <= 90u) {
        return byte + 32u;
    }
    return byte;
}

fn matches_at(index: u32, n0: u32, n1: u32, n2: u32, n3: u32, n4: u32, n5: u32, n6: u32, n7: u32, n8: u32, n9: u32, n10: u32, len: u32) -> bool {
    if (index + len > params.input_len) {
        return false;
    }
    let needle = array<u32, 11>(n0, n1, n2, n3, n4, n5, n6, n7, n8, n9, n10);
    for (var offset = 0u; offset < len; offset = offset + 1u) {
        if (lower_ascii(input[index + offset]) != needle[offset]) {
            return false;
        }
    }
    return true;
}

fn contains_pattern(lane: u32, n0: u32, n1: u32, n2: u32, n3: u32, n4: u32, n5: u32, n6: u32, n7: u32, n8: u32, n9: u32, n10: u32, len: u32) -> bool {
    for (var index = lane; index < params.input_len; index = index + 64u) {
        if (matches_at(index, n0, n1, n2, n3, n4, n5, n6, n7, n8, n9, n10, len)) {
            return true;
        }
    }
    return false;
}

var<workgroup> lane_matches: array<u32, 64>;

@compute @workgroup_size(64)
fn security_detection_detect_xss(@builtin(local_invocation_id) lid: vec3<u32>) {
    let lane = lid.x;
    lane_matches[lane] = select(0u, 1u,
        contains_pattern(lane, 60u, 115u, 99u, 114u, 105u, 112u, 116u, 0u, 0u, 0u, 0u, 7u) ||
        contains_pattern(lane, 106u, 97u, 118u, 97u, 115u, 99u, 114u, 105u, 112u, 116u, 58u, 11u) ||
        contains_pattern(lane, 111u, 110u, 101u, 114u, 114u, 111u, 114u, 61u, 0u, 0u, 0u, 8u) ||
        contains_pattern(lane, 111u, 110u, 108u, 111u, 97u, 100u, 61u, 0u, 0u, 0u, 0u, 7u));
    workgroupBarrier();
    if (lane == 0u) {
        var found = 0u;
        for (var i = 0u; i < 64u; i = i + 1u) {
            found = found | lane_matches[i];
        }
        output[0] = found;
    }
}
"#
    }
}

/// Return true for script tags, JavaScript URLs, or executable event handlers.
///
/// # Errors
///
/// Returns `Fix: ...` when input exceeds 64 MiB.
pub fn detect_xss(input: &[u8]) -> Result<bool, DetectionError> {
    attacks::detect_xss(input)
}

/// Compatibility surface for the previous generated implementation module.
pub mod implementation {
    pub use super::detect_xss;
    /// Compatibility module for callers that used the generated kernel path.
    pub mod kernel {
        pub use super::super::detect_xss;
    }

    /// Compatibility module for callers that used the generated lowering path.
    pub mod lowering {
        /// Compatibility module for callers that used `implementation::lowering::wgsl`.
        pub mod wgsl {
            pub use super::super::super::lowering::source;
        }
    }
}