Skip to content
Snippets Groups Projects
Commit 39f38a62 authored by rtekin's avatar rtekin
Browse files

scheduler primitives - to complete

parent a2f29a62
No related branches found
No related tags found
No related merge requests found
#include <cuda.h>
##include "rtkernel.hpp"
/****************** TYPE DEFINITION ******************/
stuct p_kernel {
kernel_t kernel;
int priority;
int gridsize;
int blocksize;
int period;
void* arg;
};
typedef struct p_kernel p_kernel_t;
typedef int policy_t;
/*
* define a new type `scheduler_t`, a data structure for 2 informations :
+ a kernel execution (and eventually memory operations) control method
(one of the actual 3 strategies)
+ a scheduling policy
*/
struct scheduler {
policy_t policy;
method_t method;
};
typedef struct scheduler scheduler_t;
/****************** VARIABLE DEFINITION ******************/
static hash_set <p_kernel_t> addedKernel;
/* Associates priority to the list of kernels to launch on the GPU
that have this priority
*/
static map <int, list<p_kernel_t> > kernelsToLaunch; /*! qu'il n'y ai pas de getjob avant le remove -> semaphore ?*/
/* A var to contains the following informations :
- GPU task scheduling policy
- kernel execution control strategy
*/
static scheduler_t scheduler;
/****************** METHOD DEFINITION ******************/
void removeJob(kernel_t kernelToRemove){
//lock
//-> supprime kernelToRemove de la liste correspondante à sa prio
//si cette liste devient vide supprime la paire (prio, liste)
//unlock
return;
}
p_kernel_t getJob(){
p_kernel_t pkernel;
/*according to the scheduling policy contained in scheduler.policy :
switch (scheduler.policy) ...
*/
switch (scheduler.method) {
case SAMESTREAM:
//lock
//SI kernelsToLaunch non vide et kernelsToLaunch.begin()->second non vide
pkernel = kernelsToLaunch.begin()->second.begin();
//unlock
removeJob(pkernel);
break;
case SEVERALSTREAMS:
break;
case CHOSENSM:
break;
default:
fprintf(stderr, "ERROR : INVALID strategy, i.e not one of : %s\n", "SAMESTREAM, SEVERALSTREAMS");
exit(-1);
}
return pkernel;
}
void *launchKernels(void *arg){
p_kernel_t pkernel;
while(1) {
pkernel = getJob();
pkernel.kernel<<<pkernel.gridsize, pkernel.blocksize, 0, getStream()>>>(pkernel.arg);
}
}
void suscribe(/*kernel_t kernel, int priority*/p_kernel_t *pk){
switch (scheduler.method) {
case SAMESTREAM:
kernelsToLaunch[priority].push_back(pkernel);
break;
case SEVERALSTREAMS:
break;
case CHOSENSM:
break;
default:
fprintf(stderr, "ERROR : INVALID strategy, i.e not one of : %s\n", "SAMESTREAM, SEVERALSTREAMS");
exit(-1);
}
//il y a consommer
return;
}
void *pruda_task(void *arg){
struct timespec next;
p_kernel_t *pk = (p_kernel_t *) arg;
while(1){
clock_gettime(CLOCK_MONOTONIC, &next);
suscribe(/*pk->kernel, pk->priority*/pk);
timespec_addto(next, pk->period);
clock_nanosleep(CLOCK_MONOTONIC, 0, &next, 0);
}
}
void start_kernels(scheduler_t scheduler){
//lance les threads cpu pruda_task et lauchKernels -> un thread de chaque
// par p_kernel_t de `addedKernel`
}
void init_scheduler(method_t method, policy_t policy;){
scheduler.method = method;
scheduler.policy = policy;
start_kernels(scheduler);
}
void addRtKernel(kernel_t kernel, int priority, int gridsize, int blocksize, int period, void* arg) {
p_kernel_t pkernel;
pkernel.kernel = kernel;
pkernel.priority = priority;
pkernel.gridsize = gridsize;
pkernel.blocksize = blocksize;
pkernel.period = period;
pkernel.arg = arg;
addedKernel.insert(pkernel);
return;
}
0% Loading or .
You are about to add 0 people to the discussion. Proceed with caution.
Please register or to comment