pushing everything

Makefile

+CC      =  nvcc
+
+main: runqueue.o tools.o timeops.o test.o
+	$(CC)   tools.o runqueue.o timeops.o test.o -o  testing  
+
+
+test.o : test.cu
+	$(CC)  -c test.cu -o test.o -dc
+
+timeops.o : timeops.cu
+	$(CC) -c timeops.cu -o timeops.o -dc
+
+runqueue.o : runqueue.cu
+	$(CC)  -c runqueue.cu -o runqueue.o -dc
+
+
+tools.o : tools.cu
+	$(CC) -c tools.cu -o tools.o -dc 
+
+clean:
+	rm -f *.o testing *~

image.c

+#include <SDL/SDL.h>
+#include <assert.h>
+#include <stdint.h>
+#include <stdlib.h>
+
+#define WIDTH 256
+#define HEIGHT 256
+
+static _Bool init_app(const char * name, SDL_Surface * icon, uint32_t flags)
+{
+    atexit(SDL_Quit);
+    if(SDL_Init(flags) < 0)
+        return 0;
+
+    SDL_WM_SetCaption(name, name);
+    SDL_WM_SetIcon(icon, NULL);
+
+    return 1;
+}
+
+static uint8_t * init_data(uint8_t * data)
+{
+    for(size_t i = WIDTH * HEIGHT * 3; i--; )
+        data[i] = (i % 3 == 0) ? (i / 3) % WIDTH :
+            (i % 3 == 1) ? (i / 3) / WIDTH : 0;
+
+    return data;
+}
+
+static _Bool process(uint8_t * data)
+{
+    for(SDL_Event event; SDL_PollEvent(&event);)
+        if(event.type == SDL_QUIT) return 0;
+
+    for(size_t i = 0; i < WIDTH * HEIGHT * 3; i += 1 + rand() % 3)
+      {
+
+	data[i] = rand() % 160;
+	//	printf("%d \n",data[i]);
+	if (data[i]>80)
+	  data[i] = 1;
+	else data[i]= 255;
+      }
+
+    return 1;
+}
+
+static void render(SDL_Surface * sf)
+{
+    SDL_Surface * screen = SDL_GetVideoSurface();
+    if(SDL_BlitSurface(sf, NULL, screen, NULL) == 0)
+        SDL_UpdateRect(screen, 0, 0, 0, 0);
+}
+
+static int filter(const SDL_Event * event)
+{ return event->type == SDL_QUIT; }
+
+#define mask32(BYTE) (*(uint32_t *)(uint8_t [4]){ [BYTE] = 0xff })
+
+int main(int argc, char * argv[])
+{
+    (void)argc, (void)argv;
+    static uint8_t buffer[WIDTH * HEIGHT * 3];
+
+    _Bool ok =
+        init_app("SDL example", NULL, SDL_INIT_VIDEO) &&
+        SDL_SetVideoMode(WIDTH, HEIGHT, 24, SDL_HWSURFACE);
+
+    assert(ok);
+
+    SDL_Surface * data_sf = SDL_CreateRGBSurfaceFrom(
+        init_data(buffer), WIDTH, HEIGHT, 24, WIDTH * 3,
+        mask32(0), mask32(1), mask32(2), 0);
+
+    SDL_SetEventFilter(filter);
+
+
+
+
+    //  To be continued !! 
+    cudaMallocManaged();
+
+    
+    for(; ; SDL_Delay(10))
+      {
+	process(buffer);
+	render(data_sf);
+      }
+    return 0;
+}

image.cu

+#include <SDL/SDL.h>
+#include <assert.h>
+#include <stdint.h>
+#include <stdlib.h>
+
+#define WIDTH 256
+#define HEIGHT 256
+
+static _Bool init_app(const char * name, SDL_Surface * icon, uint32_t flags)
+{
+    atexit(SDL_Quit);
+    if(SDL_Init(flags) < 0)
+        return 0;
+
+    SDL_WM_SetCaption(name, name);
+    SDL_WM_SetIcon(icon, NULL);
+
+    return 1;
+}
+
+static uint8_t * init_data(uint8_t * data)
+{
+    for(size_t i = WIDTH * HEIGHT * 3; i--; )
+        data[i] = (i % 3 == 0) ? (i / 3) % WIDTH :
+            (i % 3 == 1) ? (i / 3) / WIDTH : 0;
+
+    return data;
+}
+
+static _Bool process(uint8_t * data)
+{
+    for(SDL_Event event; SDL_PollEvent(&event);)
+        if(event.type == SDL_QUIT) return 0;
+
+    for(size_t i = 0; i < WIDTH * HEIGHT * 3; i += 1 + rand() % 3)
+      {
+
+	data[i] = rand() % 160;
+	//	printf("%d \n",data[i]);
+	if (data[i]>80)
+	  data[i] = 1;
+	else data[i]= 255;
+      }
+
+    return 1;
+}
+
+static void render(SDL_Surface * sf)
+{
+    SDL_Surface * screen = SDL_GetVideoSurface();
+    if(SDL_BlitSurface(sf, NULL, screen, NULL) == 0)
+        SDL_UpdateRect(screen, 0, 0, 0, 0);
+}
+
+static int filter(const SDL_Event * event)
+{ return event->type == SDL_QUIT; }
+
+#define mask32(BYTE) (*(uint32_t *)(uint8_t [4]){ [BYTE] = 0xff })
+
+int main(int argc, char * argv[])
+{
+    (void)argc, (void)argv;
+    static uint8_t buffer[WIDTH * HEIGHT * 3];
+
+    _Bool ok =
+        init_app("SDL example", NULL, SDL_INIT_VIDEO) &&
+        SDL_SetVideoMode(WIDTH, HEIGHT, 24, SDL_HWSURFACE);
+
+    assert(ok);
+
+    SDL_Surface * data_sf = SDL_CreateRGBSurfaceFrom(
+        init_data(buffer), WIDTH, HEIGHT, 24, WIDTH * 3,
+        mask32(0), mask32(1), mask32(2), 0);
+
+    SDL_SetEventFilter(filter);
+
+    for(; ; SDL_Delay(10))
+      {
+	process(buffer);
+	render(data_sf);
+      }
+    return 0;
+}

runqueue.cu

+#include "runqueue.h"
+
+
+
+struct pruda_runqueue_t * create_pruda_runqueue(){
+  struct pruda_runqueue_t * rq = (struct pruda_runqueue_t *)(malloc(sizeof(struct pruda_runqueue_t)));
+  rq->size = 0;
+  return rq;
+}
+
+int add_tail_pruda_task_runqueue(struct pruda_task_t * task, struct pruda_runqueue_t * rq){
+  if (rq == NULL){
+    printf("task queue is empty, exiting \n");
+    exit(-1);
+  }
+  if (rq->size >= MAX_PRUDA_TASK_PER_QUEUE){
+    fprintf( stderr, "Runqueue size exceeded, pruda task adding failed ");
+    return FAIL;
+  }
+  rq->list[rq->size] = task;
+  rq->size++;
+  return SUCCESS;
+}
+
+
+
+int del_tail_pruda_task_runqueue(struct pruda_runqueue_t *rq){
+  if (rq->size <= 0){
+    fprintf( stderr, "Runqueue  is empty, nothing to delete");
+    return FAIL;
+  }
+  rq->size--;
+  return SUCCESS;
+}
+
+
+struct pruda_task_t  * get_tail_pruda_task_runqueue(struct pruda_runqueue_t *rq){
+  return rq->list[rq->size-1];
+}
+
+void destroy_pruda_runqueue(struct pruda_runqueue_t *rq){
+  free(rq);
+}
+
+// runqueue list operations
+struct pruda_runqueue_list_t * create_pruda_runqueues_list(){
+  struct pruda_runqueue_list_t *rql =  (struct pruda_runqueue_list_t *)(malloc(sizeof(struct pruda_runqueue_list_t)));
+  for (int i=0;i< RUNQUEUES_NMB;i++)
+    rql->list[i]=create_pruda_runqueue();
+
+  return rql;
+}
+
+struct pruda_runqueue_t * get_most_priority_queue_fixed_priority(struct pruda_runqueue_list_t *rql){
+   for (int i=0;i< RUNQUEUES_NMB;i++)
+     if (rql->list[i]->size >0)
+       return rql->list[i];
+  return NULL;
+}
+
+
+void print_rq_state(struct pruda_runqueue_list_t *rql){
+  printf("15  %d |  20: %d |  2: %d \n", rql->list[15]->size, rql->list[20]->size,
+	 rql->list[2]->size); 
+}
+
+void destroy_pruda_runqueue_list_t(struct  pruda_runqueue_list_t * rql){
+  for (int i=0;i<RUNQUEUES_NMB;i++)
+    destroy_pruda_runqueue(rql->list[i]);
+}
+
+
+void add_pruda_task_fixed_priority(struct pruda_task_t * tau, struct pruda_runqueue_list_t * rql){
+  add_tail_pruda_task_runqueue(tau,rql->list[tau->priority]);
+}
+
+
+struct pruda_task_t *  get_most_priority_task_fixed_priority(struct pruda_runqueue_list_t * rql){
+  struct pruda_runqueue_t * rq = get_most_priority_queue_fixed_priority(rql);
+  if (rq!=NULL)
+    return rq->list[rq->size-1];
+  return NULL;
+  
+}

runqueue.h

+#ifndef PRUDA_RUNQUEUE_H
+#define PRUDA_RUNQUEUE_H
+
+#define RUNQUEUES_NMB 32
+#define MAX_PRUDA_TASK_PER_QUEUE 32
+
+
+#define SUCCESS 1
+#define FAIL 0 
+
+
+#include <stdlib.h>
+#include <stdio.h>
+
+#include "tools.h"
+
+
+
+struct pruda_runqueue_t{
+  struct pruda_task_t * list[MAX_PRUDA_TASK_PER_QUEUE];
+  int size; 
+};
+
+struct pruda_runqueue_list_t{
+  struct pruda_runqueue_t  * list[RUNQUEUES_NMB];
+};
+
+
+
+// run queue operations
+struct pruda_runqueue_t *  create_pruda_runqueue();
+int add_tail_pruda_task_runqueue(struct pruda_task_t *, struct pruda_runqueue_t *);
+int del_tail_pruda_task_runqueue(struct pruda_runqueue_t *);
+struct pruda_task_t * get_tail_pruda_task_runqueue(struct pruda_runqueue_t *);
+void destroy_pruda_runqueue(struct pruda_runqueue_t *);
+
+// runqueue list operations
+struct pruda_runqueue_list_t * create_pruda_runqueues_list();
+void destroy_prudarunqueue_list_t(struct pruda_runqueue_list_t *);
+
+
+void add_pruda_task_fixed_priority(struct pruda_task_t * tau, struct pruda_runqueue_list_t * rql);
+struct pruda_task_t *  get_most_priority_task_fixed_priority(struct pruda_runqueue_list_t * rql);
+struct pruda_runqueue_t * get_most_priority_queue_fixed_priority(struct pruda_runqueue_list_t *);
+
+
+
+void print_rq_state(struct pruda_runqueue_list_t *rql);
+#endif

test.cu

+#include "runqueue.h"
+#include "tools.h"
+
+
+
+#include <stdio.h>
+#include <stdlib.h>
+#include <unistd.h>
+
+
+
+// // Kernel function to add the elements of two arrays
+// __global__ void add()
+// {
+//   __allocate_to_sm(0);
+//   printf("I am processed by %lu\n",(unsigned long)__get_smid());
+// }
+
+
+
+
+__global__ void kernel_a_code(){
+    printf(" ******** this is the kernel a  code \n");
+    int res = 0; 
+    for (int i=0;i<500000000;i++){
+      res +=i * 5;
+    }
+}
+
+
+
+__global__  void kernel_b_code(){
+printf(" ****************************** this is the kernel b code \n");
+    int res = 0; 
+    for (int i=0;i<500000000;i++){
+      res +=i * 5;
+    }
+}
+
+__global__  void kernel_c_code(){
+printf(" ***************************************** this is the kernel c code \n");
+    int res = 0; 
+    for (int i=0;i<500000000;i++){
+      res +=i * 5;
+    }
+}
+
+
+
+int main(int argc, char  ** argv){
+
+
+
+  struct pruda_task_t * p_task_a = create_pruda_task(0, kernel_a_code, 600000, 6000000, 15, 1, 1);
+  struct pruda_task_t * p_task_b = create_pruda_task(1, kernel_b_code, 600000, 6000000, 20, 1, 1);
+  struct pruda_task_t * p_task_c = create_pruda_task(2, kernel_c_code, 600000, 6000000, 2, 1, 1);
+
+
+  init_scheduler(SINGLE, FP);
+  add_pruda_task(p_task_a);
+  add_pruda_task(p_task_b);
+  add_pruda_task(p_task_c);
+  
+  printf("Sched initialed, creating cpu threads \n");
+
+  create_cpu_threads();
+
+  sleep(20);
+ 
+}

timeops.cu

+#include "timeops.h"
+
+int cmp_spec(struct timespec *a, struct timespec *b){
+return 0;
+
+}
+
+void  add_spec_us(struct timespec * s, long time_us){
+  s->tv_nsec += time_us * 1000;
+  while (s->tv_nsec >= 1000000000) {
+    s->tv_nsec = s->tv_nsec - 1000000000;
+    s->tv_sec += 1;
+  }
+}

timeops.h

+#ifndef TIMEOPS_H
+#define TIMEOPS_H
+
+#include <time.h>
+#include <stdio.h>
+
+
+
+int cmp_spec(struct timespec *a, struct timespec *b);
+void add_spec_us(struct timespec * s, long time_us);
+
+
+#endif 

tools.cu

+#include "tools.h"
+
+
+static __device__ __inline__ uint32_t __get_smid(){    
+  uint32_t smid;    
+  asm volatile("mov.u32 %0, %%smid;" : "=r"(smid));    
+  return smid;
+  }
+
+static __device__ __inline__ int  __check_to_sm(uint32_t sm){     
+  return __get_smid()==sm;
+  }
+
+static __device__ __inline__ void  __allocate_to_sm(int sm){    
+       if (!__check_to_sm(sm))
+       	  asm("exit;");
+  }
+
+
+
+
+struct scheduler_t  * scheduler; 
+
+// Houssam : need to declare indexes methods 
+void pruda_alloc_sm(int sm){}
+int  pruda_get_sm(){
+  return 0;
+}
+int  pruda_check_sm(int sm){
+  return 0;
+}
+void pruda_thread_exit(){}
+void pruda_kernel_abort(){}
+
+
+void init_scheduler(int strategy, int policy){
+
+  scheduler = (struct scheduler_t *) (malloc(sizeof(struct scheduler_t)));
+
+  scheduler->strategy = strategy;
+  scheduler->policy = policy;
+
+  scheduler->tq  = create_pruda_runqueue();
+  scheduler->rql = create_pruda_runqueues_list();
+
+
+  // Houssam: This part need to be redefined so to dynamically be parametrized  
+  scheduler->sm0rq= create_pruda_runqueue();
+  scheduler->sm1rq=create_pruda_runqueue();
+
+
+ scheduler->mut = PTHREAD_MUTEX_INITIALIZER;
+  switch (policy)
+    {
+    case EDF:	
+      scheduler->pruda_subscribe = &pruda_subscribe_edf;
+      scheduler->pruda_resched = &pruda_resched_edf;
+      break;
+      
+    case FP:
+      scheduler->pruda_subscribe = &pruda_subscribe_fp;
+      scheduler->pruda_resched = &pruda_resched_fp;
+      break;
+      
+    default:
+      printf("Unknown scheduling policy, exiting ...  \n");
+      exit(-1);
+    }
+   
+
+  cudaStreamCreate(&(scheduler->hsq));
+  cudaStreamCreate(&(scheduler->lsq));
+  // Houssam : Need to set the priority between streams 
+  scheduler->lsq_free = 0;
+  scheduler->hsq_free = 0;
+
+}
+
+
+void pruda_subscribe_fp(struct pruda_task_t *tau){
+
+  // pthread_mutex_lock(&(scheduler->mut));
+  if (tau->priority < 0 || tau->priority >= RUNQUEUES_NMB ){
+    printf("Task priority out of Range, exiting \n");
+    exit(-1);
+  }
+  add_pruda_task_fixed_priority(tau,scheduler->rql);
+  // pthread_mutex_unlock(&(scheduler->mut));
+  scheduler->pruda_resched();
+}
+
+
+
+
+void pruda_resched_fp_single(){
+  pthread_mutex_lock(&(scheduler->mut));
+  if (scheduler->lsq_free == 1) {
+     pthread_mutex_unlock(&(scheduler->mut));
+    return;
+  }
+  print_rq_state(scheduler->rql);
+  struct pruda_runqueue_t * rq_h = get_most_priority_queue_fixed_priority(scheduler->rql);
+  
+
+  if (rq_h != NULL) {
+    
+    struct pruda_task_t * mp  = rq_h->list[rq_h->size-1];
+    del_tail_pruda_task_runqueue(rq_h);
+
+    mp->str = &(scheduler->lsq);
+    (*(mp->kernel_func))<<<mp->gs,mp->bs,0,(*(mp->str)) >>>();
+    cudaError_t code2= cudaGetLastError(); 
+    if (code2 != cudaSuccess) 
+      {
+	printf("Running error: %s \n", cudaGetErrorString(code2));
+	exit(-1);
+      }
+    sem_post(&(mp->wait_exec));
+    scheduler->lsq_free =1;
+  }
+  pthread_mutex_unlock(&(scheduler->mut));
+}
+
+void pruda_resched_fp_multiple(){}
+void pruda_resched_multiproc(){}
+
+void pruda_resched_fp(){  
+  switch (scheduler->strategy)
+    {
+    case SINGLE:	
+      pruda_resched_fp_single();
+      break;
+    case MULTIPLE:
+      pruda_resched_fp_multiple();
+      break;
+    case MULTIPROC:
+      pruda_resched_multiproc();
+      break;
+    default:
+      printf("Unknown strategy, exiting ...  \n");
+      exit(-1);
+    }
+}
+
+void pruda_subscribe_edf(struct pruda_task_t *tau){
+
+  // Houssam : need to be completed init params 
+}
+
+void pruda_resched_edf(){
+  // Houssam : need to be completed init params 
+}
+
+
+void reset_pruda_task_queue(){
+  // Houssam : need to be completed init params 
+  //return reset_pruda_runqueue(scheduler->tq);
+}
+
+int add_pruda_task(struct pruda_task_t *tau){
+  return add_tail_pruda_task_runqueue(tau, scheduler->tq);
+}
+
+int del_tail_pruda_task_from_tq(){
+  return del_tail_pruda_task_runqueue(scheduler->tq); 
+}
+
+struct pruda_task_t * create_pruda_task(int id, kernel_t kernel_func, long period_us,
+					long deadline_us, long priority, int bs, int gs){
+
+  
+  struct pruda_task_t * task = (struct pruda_task_t *)(malloc(sizeof(struct pruda_task_t)));
+  
+  task->id = id;
+  task->kernel_func=kernel_func;
+  task->period_us=period_us;
+  task->deadline_us=deadline_us;
+  task->priority=priority;
+  task->bs=bs;
+  task->gs=gs;
+  task->str = NULL;
+  sem_init(&(task->wait_exec), 0, 0); 
+
+  return task; 
+}
+
+void print_pruda_task(const struct pruda_task_t *task){
+  if (!task){
+    printf("Task is an empty pointer \n");
+    return;
+  }
+  printf("[Task: id= %d, T= %lu , D= %lu, P= %lu, gs= %d, bs= %d ] \n",
+	 task->id, task->period_us, task->deadline_us, task->priority,
+	 task->gs, task->bs);
+}
+
+
+
+
+void *pruda_task(void *args){
+
+  struct pruda_task_t *task = (struct pruda_task_t *)(args);
+  struct timespec next;
+
+
+
+  pthread_barrier_wait(&(scheduler->sync_all_barrier));
+
+   while (1){
+
+    
+    clock_gettime(CLOCK_REALTIME, &next);
+    add_spec_us(&next,task->period_us);
+
+    scheduler->pruda_subscribe(task);
+
+    // Sync task execution 
+    sem_wait(&(task->wait_exec));
+    cudaError_t code= cudaStreamSynchronize(*(task->str));
+    if (code != cudaSuccess) 
+      {
+	printf("Stream synchronization error : %s \n", cudaGetErrorString(code));
+	exit(-1);
+      }	
+    if (task->str == &(scheduler->hsq)){
+	scheduler->hsq_free = 0;
+      }
+    else {
+      scheduler->lsq_free = 0;
+    }
+    scheduler->pruda_resched();
+
+    
+    clock_nanosleep(CLOCK_REALTIME,TIMER_ABSTIME,&next,NULL);
+    
+  }
+}
+
+
+
+void create_cpu_threads(){
+
+  pthread_barrier_init(&(scheduler->sync_all_barrier),NULL, scheduler->tq->size);
+  for (int i=0;i<scheduler->tq->size;i++){
+
+  
+  pthread_attr_init(&(scheduler->tq->list[i]->attr));
+  pthread_attr_setschedpolicy(&(scheduler->tq->list[i]->attr), SCHED_FIFO);
+  scheduler->tq->list[i]->param.sched_priority =  scheduler->tq->list[i]->priority;
+  pthread_attr_setschedparam(&(scheduler->tq->list[i]->attr), &(scheduler->tq->list[i]->param));   
+
+   
+  pthread_create(&(scheduler->tq->list[i]->th), &(scheduler->tq->list[i]->attr), pruda_task, scheduler->tq->list[i]);
+  }
+}
+
+
+
+int pruda_task_queue_size(){
+  return scheduler->tq->size;
+}

tools.h

+#ifndef PRUDA_TOOLS_H
+#define PRUDA_TOOLS_H
+
+
+#include <time.h>
+#include <pthread.h>
+#include <semaphore.h>
+#include <cstdint>
+
+
+#include "runqueue.h"
+#include "timeops.h"
+
+
+
+
+#define SINGLE 1
+#define MULTIPLE 2
+#define MULTIPROC 3
+
+
+#define EDF 4
+#define FP 5
+
+
+
+typedef void (*kernel_t)();
+
+
+struct pruda_task_t {
+  int id;
+  kernel_t kernel_func;
+  long period_us;
+  long deadline_us;
+  long priority;
+  int bs;
+  int gs;
+
+
+  struct sched_param param;
+
+  // CPU Thread params
+  pthread_t th;
+  pthread_attr_t attr;
+
+
+  sem_t wait_exec;
+
+  // stream
+  cudaStream_t *str; 
+};
+
+
+
+struct scheduler_t {
+  int strategy;
+  int policy;
+
+  // Scheduling functionaliteis 
+  void (*pruda_subscribe)(struct pruda_task_t* );
+  void (*pruda_resched)();
+  
+
+  // All pruda tasks queue
+  struct pruda_runqueue_t *tq;
+
+  // pruda runqueue 
+  struct pruda_runqueue_list_t * rql;
+
+
+  // Houssam: This part need to be redefined so to dynamically be parametrized  
+  // SM runqueues
+  struct pruda_runqueue_t *sm0rq;
+  struct pruda_runqueue_t *sm1rq;
+
+  // Cuda streams 
+    cudaStream_t hsq;
+    cudaStream_t lsq;
+    int hsq_free;
+    int lsq_free;
+
+
+
+  // protect resched from multiple accesses
+  pthread_mutex_t mut; 
+
+
+
+  pthread_barrier_t sync_all_barrier;
+
+  
+
+
+};
+
+// need to declare indexes methods
+
+void pruda_alloc_sm(int sm);
+int pruda_get_sm();
+int pruda_check_sm(int sm);
+void pruda_thread_exit();
+void pruda_kernel_abort();
+
+
+
+
+struct pruda_task_t * create_pruda_task(int id, kernel_t kernel_func, long period_us, long deadline_us, long priority, int bs, int gs);
+
+void init_scheduler(int strategy, int policy);
+
+
+void pruda_subscribe_fp(struct pruda_task_t *);
+void pruda_resched_fp();
+
+void pruda_subscribe_edf(struct pruda_task_t *);
+void pruda_resched_edf();
+
+
+
+void reset_pruda_task_queue();
+int add_pruda_task(struct pruda_task_t *);
+int del_tail_pruda_task_from_tq();
+
+
+void print_pruda_task(const struct pruda_task_t *task);
+
+void *pruda_task(void *args);
+
+
+void create_cpu_threads();
+
+
+int pruda_task_queue_size();
+
+
+#endif