updated CUDA hybridAPI and added a CUDA hello example
[charm.git] / src / arch / cuda / hybridAPI / cuda-hybrid-api.cu
1 /* 
2  * cuda-hybrid-api.cu
3  *
4  * by Lukasz Wesolowski
5  * 04.01.2008
6  *
7  * an interface for execution on the GPU
8  *
9  * description: 
10  * -user enqueues one or more work requests to the work
11  * request queue (wrQueue) to be executed on the GPU
12  * - a converse function (gpuProgressFn) executes periodically to
13  * offload work requests to the GPU one at a time
14  *
15  */
16
17 #include "wrqueue.h"
18 #include "cuda-hybrid-api.h"
19
20 workRequestQueue *wrQueue = NULL; 
21
22 /*
23   TO DO
24   stream 1 - kernel execution
25   stream 2 - memory setup
26   stream 3 - memory copies
27 */
28
29 /* setupMemory
30    set up memory on the gpu for this kernel's execution */
31 void setupMemory(workRequest *wr) {
32
33   cudaMalloc((void **)&(wr->readWriteDevicePtr), wr->readWriteLen);
34   cudaMalloc((void **)&(wr->readOnlyDevicePtr), wr->readOnlyLen); 
35   cudaMalloc((void **)&(wr->writeOnlyDevicePtr), wr->writeOnlyLen);
36
37   cudaMemcpy(wr->readWriteDevicePtr, wr->readWriteHostPtr, wr->readWriteLen, 
38                   cudaMemcpyHostToDevice); 
39   cudaMemcpy(wr->readOnlyDevicePtr, wr->readOnlyHostPtr, wr->readOnlyLen, 
40                   cudaMemcpyHostToDevice); 
41
42
43 /* cleanupMemory
44    free memory no longer needed on the gpu */ 
45 void cleanupMemory(workRequest *wr) {
46
47   cudaMemcpy(wr->readWriteHostPtr, wr->readWriteDevicePtr, wr->readWriteLen, cudaMemcpyDeviceToHost); 
48   cudaMemcpy(wr->writeOnlyHostPtr, wr->writeOnlyDevicePtr, wr->writeOnlyLen, cudaMemcpyHostToDevice); 
49
50   cudaFree(wr->readWriteDevicePtr); 
51   cudaFree(wr->readOnlyDevicePtr); 
52   cudaFree(wr->writeOnlyDevicePtr);
53
54 }
55
56 /* kernelSelect
57    a switch statement defined by the user to allow the library to execute
58    the correct kernel */ 
59 void kernelSelect(workRequest *wr);
60
61 /* initHybridAPI
62    initializes the work request queue
63 */
64 void initHybridAPI() {
65   initWRqueue(&wrQueue); 
66 }
67
68 /* gpuProgressFn
69    called periodically to check if the current kernel has completed,
70    and invoke subsequent kernel */
71 void gpuProgressFn() {
72   if (wrQueue == NULL) {
73     return; 
74   }
75
76   while (!isEmpty(wrQueue)) {
77     workRequest *wr = head(wrQueue); 
78     
79     if (wr->executing == 0) {
80       setupMemory(wr); 
81       kernelSelect(wr); 
82       // cudaEventRecord(wr->completionEvent, 0);
83       wr->executing = 1; 
84       return; 
85     }  
86     // else if (cudaEventQuery(wr->completionEvent) == cudaSuccess ) {      
87     else if (cudaStreamQuery(0) == cudaSuccess ) {      
88       cleanupMemory(wr);
89       dequeue(wrQueue);
90       wr->callbackFn();
91     }
92       
93   }
94 }
95
96 /* exitHybridAPI
97    cleans up and deletes memory allocated for the queue
98 */
99 void exitHybridAPI() {
100   deleteWRqueue(wrQueue); 
101 }