Mercurial > hg > Game > Cerium
changeset 1561:e8c9a7099bcc draft
add set NDRange param
line wrap: on
line diff
--- a/TaskManager/Cell/CellTaskManagerImpl.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Cell/CellTaskManagerImpl.cc Tue Mar 12 16:52:49 2013 +0900 @@ -18,10 +18,8 @@ Threads *speThreads); CellTaskManagerImpl::~CellTaskManagerImpl() { - delete speThreads; delete[] speTaskList; - delete ppeManager; } @@ -365,7 +363,12 @@ { TaskListPtr tl = taskListInfo[0]->create(); bzero(tl->tasks,sizeof(Task)*TASK_MAX_SIZE); - return tl; + return tl; +} + +void +CellTaskManagerImpl::set_NDRange(void *ndr) { + speThreads->set_NDRange(ndr); } #if defined (__CERIUM_CELL__)||defined (__CERIUM_GPU__) @@ -375,6 +378,7 @@ Threads *cpus = new SpeThreads(num); #elif __CERIUM_GPU__ int num_gpu = 1; + Threads *cpus = new CpuThreads(num, useRefDma,num_gpu); num += num_gpu; // for GPU #else
--- a/TaskManager/Cell/CellTaskManagerImpl.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Cell/CellTaskManagerImpl.h Tue Mar 12 16:52:49 2013 +0900 @@ -23,7 +23,6 @@ FifoTaskManagerImpl *ppeManager; int spe_running; int spuIdle; - /* functions */ // system void init(int spuIdle,int useRefDma, int export_task_log); @@ -35,6 +34,7 @@ TaskListPtr createTaskList(); //void set_runTaskList(*QueueInfo<HTask>); void set_runTaskList(QueueInfo<HTask>* activeTaskQueue); + void set_NDRange(void* ndr); void sendTaskList(); void append_activeTask(HTaskPtr); void show_profile() ; @@ -43,6 +43,7 @@ void polling(); void debug_check_spe_idle(QueueInfo<HTask> * activeTaskQueue, int spe_running_); void print_arch(); + private: void send_taskList(int id); void show_dead_lock_info();
--- a/TaskManager/Cell/spe/SpeTaskManagerImpl.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Cell/spe/SpeTaskManagerImpl.cc Tue Mar 12 16:52:49 2013 +0900 @@ -21,7 +21,7 @@ void SpeTaskManagerImpl::export_task_log() {} void SpeTaskManagerImpl::print_arch() { printf("SpeTaskManagerImpl\n"); } - +//void SpeTaskManagerImpl::set_NDRange(void* ndr){} // Odd #ifndef __CERIUM_FIFO__
--- a/TaskManager/Cell/spe/SpeTaskManagerImpl.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Cell/spe/SpeTaskManagerImpl.h Tue Mar 12 16:52:49 2013 +0900 @@ -32,7 +32,7 @@ void polling() {} void free_htask(HTaskPtr htask) {} void print_arch(); - + void set_NDRange(void* ndr){} #ifdef __CERIUM_GPU__ SpeTaskManagerImpl(int i);
--- a/TaskManager/Fifo/FifoTaskManagerImpl.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Fifo/FifoTaskManagerImpl.cc Tue Mar 12 16:52:49 2013 +0900 @@ -325,6 +325,9 @@ printf("FifoTaskManagerImpl\n"); } +void +FifoTaskManagerImpl::set_NDRange(void* ndr) {} + TaskListPtr FifoTaskManagerImpl::createTaskList() { TaskListPtr tl = taskListInfo->create();
--- a/TaskManager/Fifo/FifoTaskManagerImpl.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Fifo/FifoTaskManagerImpl.h Tue Mar 12 16:52:49 2013 +0900 @@ -40,6 +40,7 @@ void sendTaskList(); void print_arch(); + void set_NDRange(void* ndr); // call by user private:
--- a/TaskManager/Gpu/GpuScheduler.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Gpu/GpuScheduler.cc Tue Mar 12 16:52:49 2013 +0900 @@ -12,6 +12,7 @@ GpuScheduler::GpuScheduler() { + ndr= NULL; init_impl(0); init_gpu(); } @@ -74,7 +75,7 @@ cl_kernel *kernel = new cl_kernel[2]; cl_event *event = new cl_event[2]; event[0]=NULL;event[1]=NULL; - ND_RANGE_T_PTR ndr[2]; + cl_mem *memin[2]; cl_mem *memout[2]; HTask::htask_flag flag; @@ -114,14 +115,9 @@ const char *msg=convert_error_status(ret); error(msg); } + int param = 0; - - if (flag.nd_range) { - ndr[cur] = (ND_RANGE_T_PTR)nextTask->param(0); - param++; - } - // set arg count cl_mem memparam = clCreateBuffer(context, CL_MEM_READ_ONLY, sizeof(memaddr)*nextTask->param_count, NULL, &ret); @@ -211,8 +207,8 @@ } if (flag.nd_range){ - ret = clEnqueueNDRangeKernel(command_queue,kernel[cur],ndr[cur]->dimension, - NULL,ndr[cur]->gws,ndr[cur]->lws,0,NULL, NULL); + ret = clEnqueueNDRangeKernel(command_queue,kernel[cur],ndr->dimension, + NULL,ndr->gws,ndr->lws,0,NULL, NULL); } else { ret = clEnqueueTask(command_queue, kernel[cur], 0, NULL, NULL); } @@ -334,6 +330,11 @@ } +void +GpuScheduler::set_NDRange(void* ndr_) { + ndr=(ND_RANGE_T_PTR)ndr_; +} + // regist kernel file name void gpu_register_task(int cmd, const char* filename, const char* functionname)
--- a/TaskManager/Gpu/GpuScheduler.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Gpu/GpuScheduler.h Tue Mar 12 16:52:49 2013 +0900 @@ -5,6 +5,7 @@ #include "FifoDmaManager.h" #include "GpuThreads.h" #include "HTask.h" +#include "TaskManager.h" #ifdef __APPLE__ #include <OpenCL/opencl.h> @@ -12,20 +13,15 @@ #include <CL/cl.h> #endif -typedef struct nd_range { - cl_uint dimension; - size_t gws[3]; - size_t lws[3]; -} ND_RANGE_T, *ND_RANGE_T_PTR; - class GpuScheduler : public Scheduler { -public: + public: GpuScheduler(); virtual ~GpuScheduler(); void init_impl(int useRefDma); void init_gpu(); void run(); - + void set_NDRange(void* ndr_); + void mail_write_from_host(memaddr data) { fifoDmaManager->mail_write_from_host(data); } @@ -45,6 +41,7 @@ cl_context context; cl_command_queue command_queue; cl_int ret; + ND_RANGE_T_PTR ndr; private: FifoDmaManager *fifoDmaManager; void load_kernel(int cmd);
--- a/TaskManager/Gpu/GpuThreads.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Gpu/GpuThreads.cc Tue Mar 12 16:52:49 2013 +0900 @@ -28,6 +28,11 @@ } void +GpuThreads::set_NDRange(void* ndr) { + args->scheduler->set_NDRange(ndr); +} + +void GpuThreads::init() { args->scheduler = new GpuScheduler();
--- a/TaskManager/Gpu/GpuThreads.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/Gpu/GpuThreads.h Tue Mar 12 16:52:49 2013 +0900 @@ -41,7 +41,8 @@ void send_mail(int speid, int num, memaddr *data); void add_output_tasklist(int command, memaddr buff, int alloc_size); void set_wait(SemPtr); - + void set_NDRange(void* ndr); + private: gpu_thread_arg_t *args; pthread_t *threads;
--- a/TaskManager/kernel/ppe/CpuThreads.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/kernel/ppe/CpuThreads.cc Tue Mar 12 16:52:49 2013 +0900 @@ -14,9 +14,7 @@ SchedExternTask(ShowTime); SchedExternTask(StartProfile); - -CpuThreads::CpuThreads(int num, int useRefDma, int start_id) : cpu_num(num), use_refdma(useRefDma), id_offset(start_id) { - +CpuThreads::CpuThreads(int num, int useRefDma, int start_id) : cpu_num(num), use_refdma(useRefDma), id_offset(start_id){ #ifdef __CERIUM_GPU__ gpu = new GpuThreads(useRefDma); #endif @@ -29,7 +27,7 @@ CpuThreads::~CpuThreads() { memaddr mail = (memaddr)MY_SPE_COMMAND_EXIT; - + for (int i = 0; i < cpu_num; i++) { send_mail(i+id_offset, 1, &mail); } @@ -60,7 +58,6 @@ c_scheduler->id = (int)argt->cpuid; manager->set_scheduler(c_scheduler); - SchedRegister(ShowTime); SchedRegister(StartProfile); @@ -72,8 +69,8 @@ return NULL; } + void -//CpuThreads::init() CpuThreads::init() { #ifdef __CERIUM_GPU__ @@ -99,6 +96,11 @@ } } +void +CpuThreads::set_NDRange(void *ndr) { + gpu->set_NDRange(ndr); +} + /** * このCPU からのメールを受信する。 *
--- a/TaskManager/kernel/ppe/CpuThreads.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/kernel/ppe/CpuThreads.h Tue Mar 12 16:52:49 2013 +0900 @@ -3,6 +3,7 @@ #include <pthread.h> #include "Threads.h" +#include "GpuThreads.h" #include "TaskManagerImpl.h" #include "MainScheduler.h" #include "Sem.h" @@ -14,9 +15,10 @@ TaskManagerImpl *manager; SemPtr wait; int useRefDma; + } cpu_thread_arg_t; -class GpuThreads; +//class GpuThreads; class CpuThreads : public Threads { public: @@ -32,7 +34,7 @@ virtual void send_mail(int speid, int num, memaddr *data); // BLOCKING virtual void add_output_tasklist(int command, memaddr buff, int alloc_size); virtual int is_gpu(int cpuid); - + virtual void set_NDRange(void *ndr); private: /* variables */ pthread_t *threads; @@ -41,9 +43,7 @@ int cpu_num; int use_refdma; int id_offset; -#ifdef __CERIUM_GPU__ GpuThreads *gpu; -#endif }; #endif
--- a/TaskManager/kernel/ppe/TaskManager.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/kernel/ppe/TaskManager.cc Tue Mar 12 16:52:49 2013 +0900 @@ -42,6 +42,7 @@ void TaskManager::finish() { + delete ndr; delete m_impl; } @@ -111,6 +112,11 @@ } void +TaskManager::set_NDRange(ND_RANGE_T_PTR ndr) { + m_impl->set_NDRange((void*)ndr); +} + +void TaskManager::error(const char* error_message) { printf("%s \n",error_message); exit(1);
--- a/TaskManager/kernel/ppe/TaskManager.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/kernel/ppe/TaskManager.h Tue Mar 12 16:52:49 2013 +0900 @@ -8,6 +8,12 @@ class Scheduler; class MemList; +typedef struct nd_range { + cl_uint dimension; + size_t gws[3]; + size_t lws[3]; +} ND_RANGE_T, *ND_RANGE_T_PTR; + class TaskManager { public: /* constructor */ @@ -17,6 +23,7 @@ /* variables */ TaskManagerImpl *m_impl; void (*tm_end)(TaskManager *manager); + ND_RANGE_T_PTR ndr; /* user function */ HTaskPtr create_task(int cmd); @@ -29,6 +36,7 @@ int get_cpuNum(); int get_random(); Scheduler *get_scheduler(); + void set_NDRange(ND_RANGE_T_PTR ndr) ; MemList* createMemList(int size, int count); void start_profile() { m_impl->start_profile(); }
--- a/TaskManager/kernel/ppe/TaskManagerImpl.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/kernel/ppe/TaskManagerImpl.cc Tue Mar 12 16:52:49 2013 +0900 @@ -196,7 +196,7 @@ return get_task_name(htask, 0); -} + } const char * TaskManagerImpl::get_task_name(HTaskPtr htask, int index) { if (!htask) return NULL; @@ -382,6 +382,7 @@ tl = next; } } + void error(const char *error_message) {
--- a/TaskManager/kernel/ppe/TaskManagerImpl.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/kernel/ppe/TaskManagerImpl.h Tue Mar 12 16:52:49 2013 +0900 @@ -8,6 +8,7 @@ #include "HTask.h" #include "Scheduler.h" #include "TaskLog.h" +#include <OpenCL/opencl.h> class MemList; extern QueueInfo<TaskQueue> *taskQueuePool ; @@ -15,10 +16,8 @@ extern QueueInfo<TaskList> *taskListPool; extern QueueInfo<TaskLog> *taskLogQueue; - - class TaskManagerImpl { -public: + public: /* variables */ int machineNum; @@ -49,7 +48,7 @@ virtual void append_waitTask(HTaskPtr); virtual void polling() = 0; virtual void print_arch() = 0; - + virtual void set_NDRange(void*) = 0; void check_task_finish(HTaskPtr task, QueueInfo<HTask> *wait_queue); void check_task_list_finish(SchedTask *s, TaskListPtr list, QueueInfo<HTask> *wait_queue); @@ -60,6 +59,7 @@ virtual HTaskPtr create_task(int cmd, memaddr rbuf, long r_size, memaddr wbuf, long w_size,void *from); virtual HTaskPtr create_task_array(int id, int num_task, int num_param, int num_inData, int num_outData,void *from); virtual TaskListPtr createTaskList() = 0; + const char *get_task_name(int cmd); const char *get_task_name(TaskPtr task); const char *get_task_name(SimpleTaskPtr simpletask); @@ -70,49 +70,50 @@ virtual void spawn_task(HTaskPtr); virtual void set_task_cpu(HTaskPtr, CPU_TYPE); void set_taskList(HTaskPtr htask, QueueInfo<TaskList> * taskList); - + void free_htask(HTaskPtr htask) { #if !defined(__SPU__) - if (htask->self) { - htask->flag.no_auto_free = 0; - return; - } - htaskImpl->free_(htask); + if (htask->self) { + htask->flag.no_auto_free = 0; + return; + } + htaskImpl->free_(htask); #endif } void* allocate(int size, int alignment) { - void *buff = 0; - if (size==0) return 0; + void *buff = 0; + if (size==0) return 0; #if defined(__SPU__) || ! defined(HAS_POSIX_MEMALIGN) - buff = malloc(size); + buff = malloc(size); #else - posix_memalign(&buff, alignment, size); + posix_memalign(&buff, alignment, size); #endif - if (buff==0) - get_scheduler()->printf("Can't allocate memory\n"); - return buff; + if (buff==0) + get_scheduler()->printf("Can't allocate memory\n"); + return buff; } - + void* allocate(int size) { - - void *buff = 0; - if (size==0) return 0; + + void *buff = 0; + if (size==0) return 0; #if defined(__SPU__) || ! defined(HAS_POSIX_MEMALIGN) - buff = malloc(size); + buff = malloc(size); #else - posix_memalign(&buff, DEFAULT_ALIGNMENT, size); + posix_memalign(&buff, DEFAULT_ALIGNMENT, size); #endif if (buff==0) - get_scheduler()->printf("Can't allocate memory\n"); - return buff; + get_scheduler()->printf("Can't allocate memory\n"); + return buff; } Scheduler* get_scheduler() { return scheduler; } void set_scheduler(Scheduler *s) { scheduler = s; } + } __attribute__ ((aligned (DEFAULT_ALIGNMENT))); extern void error(const char* error_message); #endif
--- a/TaskManager/kernel/ppe/Threads.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/kernel/ppe/Threads.h Tue Mar 12 16:52:49 2013 +0900 @@ -21,7 +21,7 @@ virtual void send_mail(int speid, int num, memaddr *data) = 0; // BLOCKING virtual void add_output_tasklist(int command, memaddr buff, int alloc_size) = 0; virtual int is_gpu(int cpuid) { return 0; } - + virtual void set_NDRange(void* ndr)=0; /* variables */ pthread_t *threads; int cpu_num;
--- a/TaskManager/kernel/schedule/SchedTask.h Tue Mar 05 06:52:55 2013 +0900 +++ b/TaskManager/kernel/schedule/SchedTask.h Tue Mar 12 16:52:49 2013 +0900 @@ -108,9 +108,9 @@ * swap するだけで良い。size は同じである必要がある。 */ void swap() { - void * tmp = readbuf; - readbuf = writebuf; - writebuf = tmp; + void * tmp = readbuf; + readbuf = writebuf; + writebuf = tmp; }
--- a/example/many_task/sort.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/example/many_task/sort.cc Tue Mar 12 16:52:49 2013 +0900 @@ -203,7 +203,6 @@ s->bsort[i]->wait_for(s->fsort[i+1]); s->bsort[i]->no_auto_free(); s->bsort[i]->spawn(); - printf("task list spawn \n"); } } @@ -212,7 +211,6 @@ if (!all) restart->wait_for(s->fsort[0]); for (int i = 0; i < s->split_num; i++) { s->fsort[i]->spawn(); - printf("task list spawn\n"); } if (sort_count == 1) { // last loop wait for all task
--- a/example/multiply/gpu/Multi.cl Tue Mar 05 06:52:55 2013 +0900 +++ b/example/multiply/gpu/Multi.cl Tue Mar 12 16:52:49 2013 +0900 @@ -1,5 +1,8 @@ __kernel void -add(__global const void *params,__global const float *A, __global const float*B, __global float *C) +multi(__global const void *params,__global const float *A, __global const float*B, __global float *C) { - *C=*A+*B; + int i=get_global_id(0); + + C[i]=A[i]*B[i]; + }
--- a/example/multiply/gpu/task_init.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/example/multiply/gpu/task_init.cc Tue Mar 12 16:52:49 2013 +0900 @@ -12,5 +12,5 @@ void task_init(void) { - GpuSchedRegister(MULTI_TASK, "gpu/Multi.cl","multi"); + GpuSchedRegister(MULTIPLY_TASK, "gpu/Multi.cl","multi"); }
--- a/example/multiply/main.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/example/multiply/main.cc Tue Mar 12 16:52:49 2013 +0900 @@ -3,13 +3,13 @@ #include <string.h> #include "TaskManager.h" #include "Func.h" - +/* typedef struct nd_range { int dimension; size_t gws[3]; size_t lws[3]; } ND_RANGE_T, *ND_RANGE_T_PTR; - +*/ extern void task_init(void); static int task = 1; static int length = DATA_NUM; @@ -30,13 +30,12 @@ void multi_init(TaskManager *manager) -{ +{ HTask *multiply; A = new float[length]; B = new float[length]; C = new float[length]; - for(int i=0; i<length; i++) { A[i]=(float)(i+1000); B[i]=(float)i/10.f; @@ -45,12 +44,15 @@ * Create Task * create_task(Task ID); */ - multiply = manager->create_task(MULTIPLY_TASK); + ND_RANGE_T_PTR ndr = new ND_RANGE_T; ndr->dimension = 1; - ndr->gws[0] = sizeof(C)/sizeof(C[0]); ndr->gws[1] = 1; ndr->gws[2] = 1; - ndr->lws[0] = 1;ndr->lws[1] = 1; ndr->lws[2] = 1; - multiply->set_param(0,(memaddr)ndr); + ndr->gws[0] = 100; ndr->gws[1] = 1; ndr->gws[2] = 1; + ndr->lws[0] = 1; ndr->lws[1] = 1; ndr->lws[2] = 1; + manager->set_NDRange(ndr); + + multiply = manager->create_task(MULTIPLY_TASK); + multiply->set_param(0,(memaddr)&ndr); multiply->nd_range(); multiply->set_cpu(SPE_ANY);
--- a/example/multiply/ppe/Multi.cc Tue Mar 05 06:52:55 2013 +0900 +++ b/example/multiply/ppe/Multi.cc Tue Mar 12 16:52:49 2013 +0900 @@ -2,6 +2,7 @@ #include "SchedTask.h" #include "Multi.h" #include "Func.h" +#include "GpuScheduler.h" /* これは必須 */ SchedDefineTask(Multiply); @@ -14,7 +15,7 @@ A = (float*)s->get_input(rbuf, 0); B = (float*)s->get_input(rbuf, 1); C = (float*)s->get_output(wbuf, 0); - + long length = (long)s->get_param(1); for (int i=0; i<length; i++) { C[i]=A[i]*B[i];