source: CIVL/examples/translation/openclversion2/template.cvl@ 764d472

1.23 2.0 acw/focus-triggers main test-branch
Last change on this file since 764d472 was 1c47cb2, checked in by Jacob Trieu <fuufusuu@…>, 12 years ago

Slight changes to next example, planning for 2.1

git-svn-id: svn://vsl.cis.udel.edu/civl/trunk@1198 fb995dde-84ed-4084-dfe6-e5aef3e2452c

  • Property mode set to 100644
File size: 3.4 KB
Line 
1#include <stdio.h>
2#include <stdlib.h>
3#include <string.h>
4
5//RANDOM IS UNINTERPRETED
6//THIS ASSUMES THE PROGRAMMER HAS ALREADY RAN THE ORIGINAL OPENCL CODE AND IS NOT RESPONSIBLE FOR CHECKING CORRECT USAGE OF METHODS AS OF NOW
7
8
9$input int NUM_DEVICES;
10$input int MAX_NUM_DEVICES;
11$input int CL_DEVICE_MAX_WORK_GROUP_SIZE;
12$input int LOCAL;
13$assume 0 < NUM_DEVICES && NUM_DEVICES < MAX_NUM_DEVICES;
14$gbarrier gbarrier = $gbarrier_create($here, NUM_DEVICES);
15//struct goes here
16
17
18typedef struct
19{
20 int device_id;
21 int workgroup;
22 int global_id;
23 int local_id;
24
25 //kernel variables
26}kernel;
27
28//kernel goes here
29
30void workfunc(size_t local, size_t global, kernel param)
31{
32 for(int i = local * param.workgroup; i < local * param.workgroup + local; i++)
33 {
34 param.local_id = i % local;
35 param.global_id = i;
36 //printf("My workgroup id is %d, my global id is %d, my local id is %d\n", param.workgroup, param.global_id, param.local_id);
37 func(param.workgroup, param.global_id, param.local_id, );
38 }
39}
40
41/*
42Note that the original lines were "__kernel void square( \n" \
43" __global int* input, \n" \
44" __global int* output, \n" \
45" const unsigned int count) \n" \
46
47Any parser must take note of and don't input \n, "", or \ as is
48__global int * input, __global int * output, int count;
49*/
50
51
52int main(int argc, char** argv)
53{
54 //get the number from clGetDeviceIDs 3rd parameter
55 $assert (LOCAL < CL_DEVICE_MAX_WORK_GROUP_SIZE);
56 //variables from __kernel come here
57
58
59 //from the code before
60
61 //handle the definitions being put in different places
62
63
64 //comes from clCreateBuffer
65
66
67 //Possibly keep a list of variables, with a flag for whether they are init or not
68 //Not init, malloc one from what is found in
69 //output = clCreateBuffer(context, CL_MEM_WRITE_ONLY, sizeof(int) * count, NULL, NULL);
70
71
72
73 //came from clEnqueueWriteBuffer rather than the start of code
74
75
76 //Phase after this is the definitions
77 int device_id[NUM_DEVICES];
78 //put device_ids
79 for(int i = 0; i < NUM_DEVICES; i++)
80 {
81 device_id[i] = i;
82 }
83
84 //from clCreateContext, uses the device input
85 //but there may be a loop, take it into account in the next example
86
87 //
88 //"Get" local size from clEnqueueNDRangeKernel, but is really an input
89
90 //Creates an array of the struct according to clEnqueueNDRangeKernel
91 //Have to split array into parts using local and global, and those are a workgroup
92 //For now, assume local is 1, or else inputting the arrays will be odd, for now
93 kernel param[global/local];
94 for(int i = 0; i < global/local; i++)
95 {
96 //Also picks the device to be used
97 param[i].device_id = device_id[0];
98 //other parts of the struct
99 }
100
101
102 //spawns processes according to parameters in clEnqueueNDRangeKernel
103//spawns processes according to parameters in clEnqueueNDRangeKernel
104 $proc procs[global/local];
105 for(int i = 0; i < global/local; i++)
106 {
107 param[i].workgroup = i;
108 procs[i] = $spawn workfunc(local, global, param[i]);
109 }
110
111 for(int i = 0; i < global/local; i++)
112 {
113 $wait(procs[i]);
114 }
115
116 //$barrier barrier = $barrier_create($here, gbarrier, now[i].device_id);
117 //$barrier_call(barrier);
118 //$barrier_destroy(barrier);
119
120 $gbarrier_destroy(gbarrier);
121
122 //use the information from clEnqueueReadBuffer
123 //may have to alter later
124
125 return 0;
126}
Note: See TracBrowser for help on using the repository browser.