custos_math/raw_ops/opencl/
str_op.rs

1use custos::{opencl::enqueue_kernel, prelude::CLBuffer, CDatatype, OpenCL};
2
3pub fn cl_str_op<'a, T>(
4    device: &'a OpenCL,
5    x: &CLBuffer<T>,
6    out: &CLBuffer<T>, // TODO: should be mutable
7    op: &str,
8) -> custos::Result<()>
9where
10    T: CDatatype,
11{
12    let src = format!(
13        "
14        __kernel void str_op(__global const {datatype}* lhs, __global {datatype}* out) {{
15            size_t id = get_global_id(0);
16            {datatype} x = lhs[id];
17            out[id] = {op};
18        }}
19    ",
20        datatype = T::as_c_type_str()
21    );
22
23    //let out: CLBuffer<T> = device.retrieve(x.len(), x.node.idx);
24    enqueue_kernel(device, &src, [x.len(), 0, 0], None, &[x, out])?;
25    Ok(())
26}
27
28#[inline]
29pub fn cl_str_op_mut<'a, T: CDatatype>(
30    device: &'a OpenCL,
31    x: &mut CLBuffer<T>,
32    op: &str,
33) -> custos::Result<()> {
34    cl_str_op(device, x, x, op)?;
35    Ok(())
36}