|
1 |
| -//! fast , simple and cross-platform GPGPU parallel computing library |
2 |
| -//! NOTE : there are still some problems with vulkan backend on linux |
3 |
| -//! ##Example |
4 |
| -//! - this example is for v4.0.0 C ABI |
5 |
| -//! ```c |
6 |
| -//! #include <stdio.h> |
7 |
| -//! #include <stdint.h> |
8 |
| -//! #include <stdlib.h> |
9 |
| -//! #include "EMCompute.h" |
10 |
| -//! |
11 |
| -//! int main() { |
12 |
| -//! Define the kernel |
13 |
| -//! CKernel kernel; |
14 |
| -//! kernel.x = 60000; // Number of workgroups in the x dimension |
15 |
| -//! kernel.y = 1000; |
16 |
| -//! kernel.z = 100; |
17 |
| -//! |
18 |
| -//! // WGSL code to perform element-wise addition of example_data and example_data0 |
19 |
| -//! const char* code = |
20 |
| -//! "@group(0)@binding(0) var<storage, read_write> v_indices: array<u32>; " |
21 |
| -//! "@group(0)@binding(1) var<storage, read> v_indices0: array<u32>; " |
22 |
| -//! "@compute @workgroup_size(10 , 1 , 1)" |
23 |
| -//! "fn main(@builtin(global_invocation_id) global_id: vec3<u32>) { " |
24 |
| -//! " let idx = global_id.x % 60000; " |
25 |
| -//! " " |
26 |
| -//! "v_indices[idx] = v_indices[idx] + v_indices0[idx]; " |
27 |
| -//! " " |
28 |
| -//! "}"; |
29 |
| -//! |
30 |
| -//! uintptr_t index = set_kernel_default_config(&kernel); |
31 |
| -//! kernel.kernel_code_index = register_computing_kernel_code(index , code , "main"); |
32 |
| -//! |
33 |
| -//! |
34 |
| -//! |
35 |
| -//! // Initialize data |
36 |
| -//! uint32_t example_data[60000]; |
37 |
| -//! uint32_t example_data0[60000]; |
38 |
| -//! |
39 |
| -//! for (int i = 0; i < 60000; ++i) { |
40 |
| -//! example_data[i] = 1; |
41 |
| -//! example_data0[i] = 1; |
42 |
| -//! } |
43 |
| -//! |
44 |
| -//! // Bind data |
45 |
| -//! DataBinder data; |
46 |
| -//! data.bind = 0; |
47 |
| -//! data.data = (uint8_t *)example_data; |
48 |
| -//! data.data_len = sizeof(uint32_t)*60000/sizeof(uint8_t); |
49 |
| -//! |
50 |
| -//! DataBinder data0; |
51 |
| -//! data0.bind = 1; |
52 |
| -//! data0.data = (uint8_t *)example_data0; |
53 |
| -//! data0.data_len = sizeof(uint32_t)*60000/sizeof(uint8_t); |
54 |
| -//! |
55 |
| -//! DataBinder group0[] = {data, data0}; |
56 |
| -//! GroupOfBinders wrapper; |
57 |
| -//! wrapper.group = 0; |
58 |
| -//! wrapper.datas = group0; |
59 |
| -//! wrapper.datas_len = 2; |
60 |
| -//! |
61 |
| -//! GroupOfBinders groups[] = {wrapper}; |
62 |
| -//! |
63 |
| -//! compute(&kernel, groups, 1); |
64 |
| -//! |
65 |
| -//! |
66 |
| -//! // Check results |
67 |
| -//! printf("example_data[4]: %d\n", example_data[50000]); |
68 |
| -//! printf("example_data0[4]: %d\n", example_data0[4]); |
69 |
| -//! |
70 |
| -//! free_compute_cache(); |
71 |
| -//! |
72 |
| -//! return 0; |
73 |
| -//! } |
74 |
| -//! ``` |
| 1 | +/*! fast , simple and cross-platform GPGPU parallel computing library |
| 2 | +NOTE : there are still some problems with vulkan backend on linux |
| 3 | +##Example |
| 4 | +- this example is for v6.0.0 C ABI |
| 5 | + ```c |
| 6 | +#include <stdio.h> |
| 7 | +#include <stdint.h> |
| 8 | +#include <stdlib.h> |
| 9 | +#include "EMCompute.h" |
| 10 | +
|
| 11 | +int main() { |
| 12 | + GPUDevices infos = get_computing_gpu_infos(0); |
| 13 | + printf("%s\n" , infos.infos[0].name); |
| 14 | + free_gpu_devices_infos(&infos); |
| 15 | +
|
| 16 | + // Define the kernel |
| 17 | + CKernel kernel; |
| 18 | + kernel.x = 60000; // Number of workgroups in the x dimension |
| 19 | + kernel.y = 1000; |
| 20 | + kernel.z = 100; |
| 21 | +
|
| 22 | + // WGSL code to perform element-wise addition of example_data and example_data0 |
| 23 | + const char* code = |
| 24 | + "@group(0)@binding(0) var<storage, read_write> v_indices: array<u32>; " |
| 25 | + "@group(0)@binding(1) var<storage, read_write> v_indices0: array<u32>; " |
| 26 | + "@compute @workgroup_size(10 , 1 , 1)" |
| 27 | + "fn main(@builtin(global_invocation_id) global_id: vec3<u32>) { " |
| 28 | + " let idx = global_id.x % 60000; " |
| 29 | + " " |
| 30 | + "v_indices[idx] = v_indices[idx] + v_indices0[idx]; " |
| 31 | + " " |
| 32 | + "}"; |
| 33 | +
|
| 34 | + uintptr_t index = set_kernel_default_config(&kernel); |
| 35 | + kernel.kernel_code_index = register_computing_kernel_code(index , code , "main"); |
| 36 | +
|
| 37 | +
|
| 38 | +
|
| 39 | + // Initialize data |
| 40 | + uint32_t* example_data = (uint32_t*)malloc(sizeof(uint32_t)*60000); |
| 41 | + uint32_t* example_data0 = (uint32_t*)malloc(sizeof(uint32_t)*60000); |
| 42 | +
|
| 43 | + for (int i = 0; i < 60000; ++i) { |
| 44 | + example_data[i] = 1; |
| 45 | + example_data0[i] = 1; |
| 46 | + } |
| 47 | +
|
| 48 | + // Bind data |
| 49 | + DataBinder data; |
| 50 | + data.bind = 0; |
| 51 | + data.data = (uint8_t **)&example_data; |
| 52 | + data.data_len = sizeof(uint32_t)*60000/sizeof(uint8_t); |
| 53 | +
|
| 54 | + DataBinder data0; |
| 55 | + data0.bind = 1; |
| 56 | + data0.data = (uint8_t **)&example_data0; |
| 57 | + data0.data_len = sizeof(uint32_t)*60000/sizeof(uint8_t); |
| 58 | +
|
| 59 | + DataBinder group0[] = {data, data0}; |
| 60 | + GroupOfBinders wrapper; |
| 61 | + wrapper.group = 0; |
| 62 | + wrapper.datas = group0; |
| 63 | + wrapper.datas_len = 2; |
| 64 | +
|
| 65 | + GroupOfBinders groups[] = {wrapper}; |
| 66 | + //for (int i = 0 ; i < 10000 ;++i){ |
| 67 | + compute(&kernel, groups, 1); |
| 68 | + //} |
| 69 | +
|
| 70 | + |
| 71 | +
|
| 72 | + // Check results |
| 73 | + printf("example_data[4]: %d\n", example_data[50000]); |
| 74 | + printf("example_data0[4]: %d\n", example_data0[0]); |
| 75 | +
|
| 76 | + free(example_data0); |
| 77 | + free(example_data); |
| 78 | + free_compute_cache(); |
| 79 | +
|
| 80 | + return 0; |
| 81 | +} |
| 82 | +``` |
| 83 | +*/ |
75 | 84 |
|
76 | 85 |
|
77 | 86 | use std::os::raw::c_char;
|
|
0 commit comments