Mercurial > hg > Game > Cerium

changeset 1836:56692133c5fb draft

Find changesets by keywords (author, files, the commit message), revision number or hash, or revset expression.
success run wordcount with gpu, but result is wrong
author Shohei KOKUBO <e105744@ie.u-ryukyu.ac.jp>
date Fri, 20 Dec 2013 04:36:27 +0900
parents 144e573b030b
children 093bdd5e940e
files example/word_count/gpu/Exec.cl example/word_count/main.cc example/word_count/ppe/Exec_Data_Parallel.cc example/word_count/task_init.cc
diffstat 4 files changed, 12 insertions(+), 12 deletions(-) [+]
line wrap: on
line diff
--- a/example/word_count/gpu/Exec.cl	Fri Dec 20 03:00:32 2013 +0900
+++ b/example/word_count/gpu/Exec.cl	Fri Dec 20 04:36:27 2013 +0900
@@ -1,12 +1,10 @@
 __kernel void
-wordcount(__global int *data_count,
-    __global void *rbuf,
-    __global void *wbuf)
+wordcount(__constant long *param,
+          __global char *i_data,
+          __global unsigned long long *o_data)
 {
-    __global char *i_data =  (__global char *)rbuf;
-    __global unsigned long long *o_data = (__global unsigned long long*)wbuf;
-    __global  unsigned long long *head_tail_flag = o_data +2;
-    int length = data_count[0];
+    __global unsigned long long *head_tail_flag = o_data +2;
+    long length = param[0];
     int word_flag = 0;
     int word_num = 0;
     int line_num = 0;
--- a/example/word_count/main.cc	Fri Dec 20 03:00:32 2013 +0900
+++ b/example/word_count/main.cc	Fri Dec 20 04:36:27 2013 +0900
@@ -102,7 +102,7 @@
         if (use_task_array) {
             int task_num = (w->size+size-1)/size;
             if (task_num>array_task_num) task_num = array_task_num;
-            task_array = manager->create_task_array(TASK_EXEC,task_num,0,1,1);
+            task_array = manager->create_task_array(TASK_EXEC,task_num,1,1,1);
             if (!all) { 
                 t_next->wait_for(task_array);
             } else {
@@ -120,6 +120,7 @@
                 t_exec = task_array->next_task_array(TASK_EXEC,t_exec);
                 t_exec->set_inData(0,w->file_mmap + i*w->division_size, size);
                 t_exec->set_outData(0,w->o_data + i*w->out_size, w->division_out_size);
+                t_exec->set_param(0,(long)size);
             } else if (use_compat) {
                 h_exec = manager->create_task(TASK_EXEC);
                 h_exec->set_inData(0,w->file_mmap + i*w->division_size, size);
@@ -139,7 +140,7 @@
                 h_exec->set_param(1,(long)w->division_size);
                 h_exec->set_param(2,(long)size);
                 h_exec->set_param(3,(long)w->out_size);
-
+                
                 t_next->wait_for(h_exec);
                 h_exec->set_cpu(spe_cpu);
                 h_exec->iterate(array_task_num);
--- a/example/word_count/ppe/Exec_Data_Parallel.cc	Fri Dec 20 03:00:32 2013 +0900
+++ b/example/word_count/ppe/Exec_Data_Parallel.cc	Fri Dec 20 04:36:27 2013 +0900
@@ -4,7 +4,7 @@
 #include "Func.h"
 
 /* これは必須 */
-SchedDefineTask1(Exec_Data_Prallel,wordcount);
+SchedDefineTask1(Exec_Data_Parallel,wordcount);
 
 static int
 wordcount(SchedTask *s, void *rbuf, void *wbuf)
--- a/example/word_count/task_init.cc	Fri Dec 20 03:00:32 2013 +0900
+++ b/example/word_count/task_init.cc	Fri Dec 20 04:36:27 2013 +0900
@@ -4,7 +4,7 @@
 
 /* 必ずこの位置に書いて */
 SchedExternTask(Exec);
-SchedExternTask(Exec_Data_Prallel);
+SchedExternTask(Exec_Data_Parallel);
 SchedExternTask(Print);
 SchedExternTask(RUN_TASK_BLOCKS);
 
@@ -18,9 +18,10 @@
 {
 #ifdef __CERIUM_GPU__
     GpuSchedRegister(TASK_EXEC, "gpu/Exec.cl", "wordcount");
+    GpuSchedRegister(TASK_EXEC_DATA_PARALLEL, "gpu/Exec_Data_Parallel.cl","wordcount_parallel");
 #endif
     SchedRegisterTask(TASK_EXEC, Exec);
-    SchedRegisterTask(TASK_EXEC_DATA_PARALLEL, Exec_Data_Prallel);
+    SchedRegisterTask(TASK_EXEC_DATA_PARALLEL, Exec_Data_Parallel);
 
     SchedRegisterTask(TASK_PRINT, Print);
     SchedRegister(RUN_TASK_BLOCKS);