Mercurial > hg > Members > yutaka > OpenCL

comparison WordCount/main.cc @ 0:0e6e76dbdb0f

Find changesets by keywords (author, files, the commit message), revision number or hash, or revset expression.
add file
author Yutaka_Kinjyo
date Tue, 12 Jul 2011 11:12:51 +0900
parents
children 403e35dd9b6d
comparison
equal deleted inserted replaced
-1:000000000000 0:0e6e76dbdb0f
1 #include <stdio.h>
2 #include <stdlib.h>
3 #include <string.h>
4 #include <sys/stat.h>
5 #include <sys/mman.h>
6 #include <sys/types.h>
7 #include <fcntl.h>
8 #include <unistd.h>
9 #include <OpenCL/opencl.h>
10 #include <oclUtils.h>
11
12 #define OUT_PARAM_NUM 2
13
14 typedef struct {
15 caddr_t file_mmap;
16 off_t size;
17 } st_mmap_t;
18
19 /*与えられたsizeをfix_byte_sizeの倍数にする(丸め込むっていうのかな？)*/
20 static int
21 fix_byte(int size,int fix_byte_size)
22 {
23 size = (size/fix_byte_size)*fix_byte_size + ((size%fix_byte_size)!= 0)*fix_byte_size;
24
25 return size;
26 }
27
28
29 static st_mmap_t
30 my_mmap(char *filename)
31 {
32
33 /*マッピングだよ!*/
34 int fd = -1;
35 int map = MAP_PRIVATE;
36 st_mmap_t st_mmap;
37 struct stat sb;
38
39 if ((fd=open(filename,O_RDONLY,0666))==0) {
40 fprintf(stderr,"can't open %s\n",filename);
41 }
42
43 if (fstat(fd,&sb)) {
44 fprintf(stderr,"can't fstat %s\n",filename);
45 }
46
47 printf("file size %d\n",(int)sb.st_size);
48
49 /*sizeをページングサイズの倍数にあわせる*/
50 st_mmap.size = fix_byte(sb.st_size,4096);
51
52 printf("fix 4096byte file size %d\n",(int)st_mmap.size);
53
54 st_mmap.file_mmap = (char*)mmap(NULL,st_mmap.size,PROT_READ,map,fd,(off_t)0);
55 if (st_mmap.file_mmap == (caddr_t)-1) {
56 fprintf(stderr,"Can't mmap file\n");
57 perror(NULL);
58 exit(0);
59 }
60
61 return st_mmap;
62
63 }
64
65
66
67 int main(int args, char *argv[])
68 {
69
70 char *filename = 0;
71
72 for (int i = 1; argv[i]; ++i) {
73 if (strcmp(argv[i], "-file") == 0) {
74 filename = argv[i+1];
75 } else if (strcmp(argv[i], "-help")) {
76 printf("Usage: ./word_count [-file filename]\n");
77 }
78 }
79
80 if (filename == 0) {
81 printf("Usage: ./word_count [-file filename]\n");
82 return 0;
83 }
84
85 //指定されたファイルをメモリにmap
86 st_mmap_t st_mmap = my_mmap(filename);
87
88 //kernelファイルの大きさ取得して、メモリ確保
89 int fd = -1;
90 const char *kernel_filename = "./word_count.cl";
91 if ((fd=open(kernel_filename,O_RDONLY,0666))==0) {
92 fprintf(stderr,"can't open %s\n",kernel_filename);
93 }
94
95 struct stat sb;
96 if (fstat(fd,&sb)) {
97 fprintf(stderr,"can't fstat %s\n",filename);
98 }
99
100 size_t source_size = sb.st_size;
101 void *source_str = malloc(source_size);
102
103 int err = read(fd, source_str, source_size);
104 if (err == -1) {
105 fprintf(stderr,"can't read %s\n",filename);
106 }
107
108 close(fd);
109
110 cl_platform_id platform_id = NULL;
111 cl_uint ret_num_platforms = NULL;
112 cl_int ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
113
114 cl_device_id device_id = NULL;
115 cl_uint ret_num_devices = NULL;
116 // CL_DEVICE_TYPE_CPU, CL_DEVICE_TYPE_GPU, と指定できる
117 // CL_DEVICE_TYPE_DEFAULT はどうなるのか
118 ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_DEFAULT,
119 1, &device_id, &ret_num_devices);
120
121 oclCheckError(ret, CL_SUCCESS);
122
123 //OpenCLコンテキストの作成
124 cl_context context = clCreateContext( NULL, 1, &device_id, NULL, NULL, &ret);
125 oclCheckError(ret, CL_SUCCESS);
126 cl_command_queue command_queue = clCreateCommandQueue(context, device_id, 0, &ret);
127 oclCheckError(ret, CL_SUCCESS);
128 // カーネルプログラムを読み込む
129 cl_program program = clCreateProgramWithSource(context, 1, (const char **)&source_str,
130 (const size_t *)&source_size, &ret);
131 ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);
132
133 oclCheckError(ret, CL_SUCCESS);
134
135 //カーネルプログラムをビルド
136 //Task選択にあたる
137 cl_kernel kernel = clCreateKernel(program, "word_count", &ret);
138 oclCheckError(ret, CL_SUCCESS);
139
140 //カウントするテキストデータのメモリオブジェクト
141 cl_mem text_memobj = clCreateBuffer(context, CL_MEM_READ_ONLY,
142 st_mmap.size * sizeof(char), NULL, &ret);
143
144 ret = clEnqueueWriteBuffer(command_queue, text_memobj,
145 CL_TRUE, 0, st_mmap.size, (char*)st_mmap.file_mmap,
146 0, NULL, NULL);
147
148 oclCheckError(ret, CL_SUCCESS);
149
150 // 必要なパラメータのオブジェクト
151 cl_mem param_memobj = clCreateBuffer(context, CL_MEM_READ_ONLY,
152 sizeof(int), NULL, &ret);
153
154 ret = clEnqueueWriteBuffer(command_queue, param_memobj,
155 CL_TRUE, 0, sizeof(int), (int*)(&st_mmap.size),
156 0, NULL, NULL);
157
158 oclCheckError(ret, CL_SUCCESS);
159
160
161 /*
162 * 並列度の計算
163 * wcするファイルの大きさに合わせる
164 *
165 */
166
167 // このdivi_size はどうやって決めるよ
168 int divi_size = 1024;
169 st_mmap.size / 1024;
170
171
172 size_t global_work_size = 4;
173
174 // 行数、単語数を格納する2のint配列
175 int out_size = sizeof(int) * OUT_PARAM_NUM * global_work_size;
176 cl_mem out_memobj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
177 out_size * sizeof(char), NULL, &ret);
178
179 oclCheckError(ret, CL_SUCCESS);
180
181
182 // 引数のSet
183 // memory object にしなくてもできるsetできるかも
184 ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&text_memobj);
185 ret |= clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&param_memobj);
186 ret |= clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *)&out_memobj);
187
188 oclCheckError(ret, CL_SUCCESS);
189
190 /*
191 * kernel実行
192 　　 * 並列に処理せずに work-item ひとつで動かしたい場合は、clEnqueueNDRangeKernel の簡易版 clEnqueueTask が使える
193 *
194 */　　
195
196 //ret = clEnqueueTask(command_queue, kernel, 0, NULL, NULL);
197 // global_work_size は配列。並列動作させる時の次元数にあわせて、配列の次元数も決まるはず
198 ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, &global_work_size, NULL, 0, NULL, NULL);
199
200 oclCheckError(ret, CL_SUCCESS);
201
202 int *out_data = (int*)malloc(out_size);
203
204 // 演算結果の読み込み
205 ret = clEnqueueReadBuffer(command_queue, out_memobj, CL_TRUE, 0,
206 out_size * sizeof(char), out_data, 0, NULL, NULL);
207
208 oclCheckError(ret, CL_SUCCESS);
209
210 for (int i = 0; i < global_work_size; i++) {
211 for (int j = 0; j < OUT_PARAM_NUM; j++) {
212 printf("%d ", out_data[i*OUT_PARAM_NUM+j]);
213 }
214 printf("\n");
215 }
216
217 clFlush(command_queue);
218 clFinish(command_queue);
219 clReleaseKernel(kernel);
220 clReleaseProgram(program);
221 clReleaseMemObject(text_memobj);
222 clReleaseMemObject(param_memobj);
223 clReleaseMemObject(out_memobj);
224 clReleaseCommandQueue(command_queue);
225 clReleaseContext(context);
226
227 free(source_str);
228 free(out_data);
229
230 return 0;
231
232 }