Fix any warning found by gcc -Wall and clang -Wall

18 files changed, 81 insertions, 70 deletions

diff --git a/src/api/api_event.cpp b/src/api/api_event.cpp
index 14f6ac0..6e114b2 100644
--- a/src/api/api_event.cpp
+++ b/src/api/api_event.cpp
@@ -15,7 +15,7 @@ clWaitForEvents(cl_uint             num_events,
     // Check the events in the list
     cl_context global_ctx = 0;
 
-    for (int i=0; i<num_events; ++i)
+    for (cl_uint i=0; i<num_events; ++i)
     {
         if (!event_list[i]->isA(Coal::Object::T_Event))
             return CL_INVALID_EVENT;
@@ -32,7 +32,7 @@ clWaitForEvents(cl_uint             num_events,
     }
 
     // Wait for the events
-    for (int i=0; i<num_events; ++i)
+    for (cl_uint i=0; i<num_events; ++i)
     {
         event_list[i]->waitForStatus(Coal::Event::Complete);
     }
diff --git a/src/api/api_kernel.cpp b/src/api/api_kernel.cpp
index 41332df..6c200ce 100644
--- a/src/api/api_kernel.cpp
+++ b/src/api/api_kernel.cpp
@@ -97,7 +97,7 @@ clCreateKernelsInProgram(cl_program     program,
     else
     {
         // Copy the kernels
-        for (int i=0; i<ks.size(); ++i)
+        for (size_t i=0; i<ks.size(); ++i)
         {
             kernels[i] = (cl_kernel)ks[i];
         }
diff --git a/src/api/api_program.cpp b/src/api/api_program.cpp
index 09151bc..1e32c9b 100644
--- a/src/api/api_program.cpp
+++ b/src/api/api_program.cpp
@@ -89,7 +89,7 @@ clCreateProgramWithBinary(cl_context            context,
     if (*errcode_ret != CL_SUCCESS)
         return 0;
 
-    for (int i=0; i<num_devices; ++i)
+    for (cl_uint i=0; i<num_devices; ++i)
     {
         bool found = false;
 
@@ -102,7 +102,7 @@ clCreateProgramWithBinary(cl_context            context,
             return 0;
         }
 
-        for (int j=0; j<context_num_devices; ++j)
+        for (cl_uint j=0; j<context_num_devices; ++j)
         {
             if (device_list[i] == context_devices[j])
             {
@@ -203,11 +203,11 @@ clBuildProgram(cl_program           program,
         if (result != CL_SUCCESS)
             return result;
 
-        for (int i=0; i<num_devices; ++i)
+        for (cl_uint i=0; i<num_devices; ++i)
         {
             bool found = false;
 
-            for (int j=0; j<context_num_devices; ++j)
+            for (cl_uint j=0; j<context_num_devices; ++j)
             {
                 if (device_list[i] == context_devices[j])
                 {
diff --git a/src/core/commandqueue.cpp b/src/core/commandqueue.cpp
index 9e6aeca..67cf7ae 100644
--- a/src/core/commandqueue.cpp
+++ b/src/core/commandqueue.cpp
@@ -257,7 +257,7 @@ void CommandQueue::pushEventsOnDevice()
 
         event_wait_list = event->waitEvents(count);
 
-        for (int i=0; i<count; ++i)
+        for (cl_uint i=0; i<count; ++i)
         {
             if (event_wait_list[i]->status() != Event::Complete)
             {
@@ -342,7 +342,7 @@ Event::Event(CommandQueue *parent,
              cl_int *errcode_ret)
 : Object(Object::T_Event, parent),
   p_num_events_in_wait_list(num_events_in_wait_list), p_event_wait_list(0),
-  p_device_data(0), p_status(status)
+  p_status(status), p_device_data(0)
 {
     // Initialize the locking machinery
     pthread_cond_init(&p_state_change_cond, 0);
@@ -364,7 +364,7 @@ Event::Event(CommandQueue *parent,
     }
 
     // Check that none of the events in event_wait_list is in an error state
-    for (int i=0; i<num_events_in_wait_list; ++i)
+    for (cl_uint i=0; i<num_events_in_wait_list; ++i)
     {
         if (event_wait_list[i] == 0)
         {
@@ -394,7 +394,7 @@ Event::Event(CommandQueue *parent,
     }
 
     // Explore the events we are waiting on and reference them
-    for (int i=0; i<num_events_in_wait_list; ++i)
+    for (cl_uint i=0; i<num_events_in_wait_list; ++i)
     {
         clRetainEvent((cl_event)event_wait_list[i]);
 
@@ -416,7 +416,7 @@ void Event::freeDeviceData()
 
 Event::~Event()
 {
-    for (int i=0; i<p_num_events_in_wait_list; ++i)
+    for (cl_uint i=0; i<p_num_events_in_wait_list; ++i)
         clReleaseEvent((cl_event)p_event_wait_list[i]);
 
     if (p_event_wait_list)
diff --git a/src/core/compiler.cpp b/src/core/compiler.cpp
index 138d680..dcfd694 100644
--- a/src/core/compiler.cpp
+++ b/src/core/compiler.cpp
@@ -17,8 +17,8 @@
 using namespace Coal;
 
 Compiler::Compiler(DeviceInterface *device)
-: p_log_stream(p_log), p_log_printer(0), p_device(device), p_module(0),
-  p_optimize(true)
+: p_device(device), p_module(0), p_optimize(true), p_log_stream(p_log),
+  p_log_printer(0)
 {
 
 }
diff --git a/src/core/context.cpp b/src/core/context.cpp
index 54c3d66..09ef4bc 100644
--- a/src/core/context.cpp
+++ b/src/core/context.cpp
@@ -22,8 +22,8 @@ Context::Context(const cl_context_properties *properties,
                  void *user_data,
                  cl_int *errcode_ret)
 : Object(Object::T_Context, 0), p_properties(0), p_pfn_notify(pfn_notify),
-  p_props_len(0), p_user_data(user_data), p_platform(0), p_devices(0),
-  p_num_devices(0)
+  p_user_data(user_data), p_devices(0), p_num_devices(0), p_props_len(0),
+  p_platform(0)
 {
     if (!p_pfn_notify)
         p_pfn_notify = &default_pfn_notify;
@@ -94,7 +94,7 @@ Context::Context(const cl_context_properties *properties,
         return;
     }
 
-    for (int i=0; i<num_devices; ++i)
+    for (cl_uint i=0; i<num_devices; ++i)
     {
         cl_device_id device = devices[i];
 
@@ -183,7 +183,7 @@ cl_int Context::info(cl_context_info param_name,
 
 bool Context::hasDevice(DeviceInterface *device) const
 {
-    for (int i=0; i<p_num_devices; ++i)
+    for (unsigned int i=0; i<p_num_devices; ++i)
         if (p_devices[i] == device)
             return true;
 
diff --git a/src/core/cpu/device.cpp b/src/core/cpu/device.cpp
index 31f7aae..10cd037 100644
--- a/src/core/cpu/device.cpp
+++ b/src/core/cpu/device.cpp
@@ -23,8 +23,8 @@
 using namespace Coal;
 
 CPUDevice::CPUDevice()
-: DeviceInterface(), p_cores(0), p_workers(0), p_stop(false),
-  p_num_events(0), p_initialized(false)
+: DeviceInterface(), p_cores(0), p_num_events(0), p_workers(0), p_stop(false),
+  p_initialized(false)
 {
 
 }
@@ -65,7 +65,7 @@ void CPUDevice::init()
     // Create worker threads
     p_workers = (pthread_t *)std::malloc(numCPUs() * sizeof(pthread_t));
 
-    for (int i=0; i<numCPUs(); ++i)
+    for (unsigned int i=0; i<numCPUs(); ++i)
     {
         pthread_create(&p_workers[i], 0, &worker, this);
     }
@@ -86,7 +86,7 @@ CPUDevice::~CPUDevice()
     pthread_cond_broadcast(&p_events_cond);
     pthread_mutex_unlock(&p_events_mutex);
 
-    for (int i=0; i<numCPUs(); ++i)
+    for (unsigned int i=0; i<numCPUs(); ++i)
     {
         pthread_join(p_workers[i], 0);
     }
@@ -188,6 +188,8 @@ void CPUDevice::freeEventDeviceData(Event *event)
             if (cpu_e)
                 delete cpu_e;
         }
+        default:
+            break;
     }
 }
 
diff --git a/src/core/cpu/kernel.cpp b/src/core/cpu/kernel.cpp
index ca281b2..61ef87f 100644
--- a/src/core/cpu/kernel.cpp
+++ b/src/core/cpu/kernel.cpp
@@ -188,7 +188,7 @@ llvm::Function *CPUKernel::callFunction(std::vector<void *> &freeLocal)
     llvm::SmallVector<llvm::Value *, 8> args;
 
     // Add each kernel arg to args
-    for (int i=0; i<p_kernel->numArgs(); ++i)
+    for (unsigned int i=0; i<p_kernel->numArgs(); ++i)
     {
         const Kernel::Arg &a = p_kernel->arg(i);
         llvm::Constant *arg_constant = 0;
@@ -358,8 +358,6 @@ CPUKernelEvent::~CPUKernelEvent()
 
 bool CPUKernelEvent::reserve()
 {
-    int rs;
-
     // Lock, this will be unlocked in takeInstance()
     pthread_mutex_lock(&p_mutex);
 
@@ -412,7 +410,7 @@ CPUKernelWorkGroup *CPUKernelEvent::takeInstance()
 CPUKernelWorkGroup::CPUKernelWorkGroup(CPUKernel *kernel, KernelEvent *event,
                                        CPUKernelEvent *cpu_event,
                                        const size_t *work_group_index)
-: p_kernel(kernel), p_event(event), p_cpu_event(cpu_event),
+: p_kernel(kernel), p_cpu_event(cpu_event), p_event(event),
   p_work_dim(event->work_dim())
 {
 
@@ -465,7 +463,7 @@ bool CPUKernelWorkGroup::run()
     // We may have some cleanup to do
     if (free_after)
     {
-        for (int i=0; i<local_to_free.size(); ++i)
+        for (size_t i=0; i<local_to_free.size(); ++i)
         {
             std::free(local_to_free[i]);
         }
diff --git a/src/core/cpu/worker.cpp b/src/core/cpu/worker.cpp
index b1344ef..62fc2f0 100644
--- a/src/core/cpu/worker.cpp
+++ b/src/core/cpu/worker.cpp
@@ -16,7 +16,7 @@ using namespace Coal;
 void *worker(void *data)
 {
     CPUDevice *device = (CPUDevice *)data;
-    bool stop = false, last_run;
+    bool stop = false;
     cl_int errcode;
     Event *event;
 
diff --git a/src/core/events.cpp b/src/core/events.cpp
index 9c0b169..b487227 100644
--- a/src/core/events.cpp
+++ b/src/core/events.cpp
@@ -87,7 +87,7 @@ bool BufferEvent::isSubBufferAligned(const MemObject *buffer,
 
     size_t mask = 0;
 
-    for (int i=0; i<align; ++i)
+    for (cl_uint i=0; i<align; ++i)
         mask = 1 | (mask << 1);
 
     if (((SubBuffer *)buffer)->offset() & mask)
@@ -545,7 +545,7 @@ NativeKernelEvent::NativeKernelEvent(CommandQueue *parent,
         std::memcpy((void *)p_args, (void *)args, cb_args);
 
         // Replace memory objects with global pointers
-        for (int i=0; i<num_mem_objects; ++i)
+        for (cl_uint i=0; i<num_mem_objects; ++i)
         {
             const MemObject *buffer = mem_list[i];
             const char *loc = (const char *)args_mem_loc[i];
@@ -599,7 +599,7 @@ KernelEvent::KernelEvent(CommandQueue *parent,
                          const Event **event_wait_list,
                          cl_int *errcode_ret)
 : Event(parent, Queued, num_events_in_wait_list, event_wait_list, errcode_ret),
-  p_kernel(kernel), p_work_dim(work_dim)
+  p_work_dim(work_dim), p_kernel(kernel)
 {
     if (*errcode_ret != CL_SUCCESS) return;
 
@@ -668,7 +668,7 @@ KernelEvent::KernelEvent(CommandQueue *parent,
     // Populate work_offset, work_size and local_work_size
     size_t work_group_size = 1;
 
-    for (int i=0; i<work_dim; ++i)
+    for (cl_uint i=0; i<work_dim; ++i)
     {
         if (global_work_offset)
         {
@@ -726,7 +726,7 @@ KernelEvent::KernelEvent(CommandQueue *parent,
     }
 
     // Check arguments (buffer alignment, image size, ...)
-    for (int i=0; i<kernel->numArgs(); ++i)
+    for (unsigned int i=0; i<kernel->numArgs(); ++i)
     {
         const Kernel::Arg &a = kernel->arg(i);
 
diff --git a/src/core/kernel.cpp b/src/core/kernel.cpp
index 2a5a90b..f5511bf 100644
--- a/src/core/kernel.cpp
+++ b/src/core/kernel.cpp
@@ -41,7 +41,7 @@ Kernel::~Kernel()
 
 const Kernel::DeviceDependent &Kernel::deviceDependent(DeviceInterface *device) const
 {
-    for (int i=0; i<p_device_dependent.size(); ++i)
+    for (size_t i=0; i<p_device_dependent.size(); ++i)
     {
         const DeviceDependent &rs = p_device_dependent[i];
 
@@ -54,7 +54,7 @@ const Kernel::DeviceDependent &Kernel::deviceDependent(DeviceInterface *device)
 
 Kernel::DeviceDependent &Kernel::deviceDependent(DeviceInterface *device)
 {
-    for (int i=0; i<p_device_dependent.size(); ++i)
+    for (size_t i=0; i<p_device_dependent.size(); ++i)
     {
         DeviceDependent &rs = p_device_dependent[i];
 
@@ -85,7 +85,7 @@ cl_int Kernel::addFunction(DeviceInterface *device, llvm::Function *function,
     if (!append && p_args.size() != f->getNumParams())
         return CL_INVALID_KERNEL_DEFINITION;
 
-    for (int i=0; i<f->getNumParams(); ++i)
+    for (unsigned int i=0; i<f->getNumParams(); ++i)
     {
         llvm::Type *arg_type = f->getParamType(i);
         Arg::Kind kind = Arg::Invalid;
@@ -113,7 +113,7 @@ cl_int Kernel::addFunction(DeviceInterface *device, llvm::Function *function,
                     llvm::cast<llvm::StructType>(arg_type);
                 llvm::StringRef struct_name = struct_type->getName();
 
-                if (struct_name == "imade2d")
+                if (struct_name == "image2d")
                 {
                     kind = Arg::Image2D;
                     file = Arg::Global;
@@ -275,7 +275,7 @@ const Kernel::Arg &Kernel::arg(unsigned int index) const
 
 bool Kernel::argsSpecified() const
 {
-    for (int i=0; i<p_args.size(); ++i)
+    for (size_t i=0; i<p_args.size(); ++i)
     {
         if (!p_args[i].defined())
             return false;
@@ -412,8 +412,8 @@ cl_int Kernel::workGroupInfo(DeviceInterface *device,
  * Kernel::Arg
  */
 Kernel::Arg::Arg(unsigned short vec_dim, File file, Kind kind)
-: p_vec_dim(vec_dim), p_file(file), p_kind(kind), p_defined(false),
-  p_runtime_alloc(0), p_data(0)
+: p_vec_dim(vec_dim), p_file(file), p_kind(kind), p_data(0), p_defined(false),
+  p_runtime_alloc(0)
 {
 
 }
@@ -480,6 +480,8 @@ size_t Kernel::Arg::valueSize() const
         case Image3D:
             return sizeof(cl_mem);
     }
+
+    return 0;
 }
 
 unsigned short Kernel::Arg::vecDim() const
diff --git a/src/core/memobject.cpp b/src/core/memobject.cpp
index 7d78e14..4c14ba9 100644
--- a/src/core/memobject.cpp
+++ b/src/core/memobject.cpp
@@ -15,8 +15,8 @@ using namespace Coal;
 
 MemObject::MemObject(Context *ctx, cl_mem_flags flags, void *host_ptr,
                      cl_int *errcode_ret)
-: Object(Object::T_MemObject, ctx), p_flags(flags), p_host_ptr(host_ptr),
-  p_dtor_callback(0), p_devicebuffers(0), p_num_devices(0)
+: Object(Object::T_MemObject, ctx), p_num_devices(0), p_flags(flags),
+  p_host_ptr(host_ptr), p_devicebuffers(0), p_dtor_callback(0)
 {
     // Check the flags value
     const cl_mem_flags all_flags = CL_MEM_READ_WRITE | CL_MEM_WRITE_ONLY |
@@ -63,7 +63,7 @@ MemObject::~MemObject()
     if (p_devicebuffers)
     {
         // Also delete our children in the device
-        for (int i=0; i<p_num_devices; ++i)
+        for (unsigned int i=0; i<p_num_devices; ++i)
             delete p_devicebuffers[i];
 
         std::free((void *)p_devicebuffers);
@@ -133,7 +133,7 @@ cl_int MemObject::init()
     // Create a DeviceBuffer for each device
     unsigned int failed_devices = 0;
 
-    for (int i=0; i<p_num_devices; ++i)
+    for (unsigned int i=0; i<p_num_devices; ++i)
     {
         DeviceInterface *device = devices[i];
 
@@ -203,7 +203,7 @@ void *MemObject::host_ptr() const
 
 DeviceBuffer *MemObject::deviceBuffer(DeviceInterface *device) const
 {
-    for (int i=0; i<p_num_devices; ++i)
+    for (unsigned int i=0; i<p_num_devices; ++i)
     {
         if (p_devicebuffers[i]->device() == device)
             return p_devicebuffers[i];
@@ -248,7 +248,7 @@ cl_int MemObject::info(cl_mem_info param_name,
                        size_t *param_value_size_ret) const
 {
     void *value = 0;
-    int value_length = 0;
+    size_t value_length = 0;
     class SubBuffer *subbuf = (class SubBuffer *)this;
 
     union {
@@ -368,8 +368,8 @@ MemObject::Type Buffer::type() const
 
 SubBuffer::SubBuffer(class Buffer *parent, size_t offset, size_t size,
                      cl_mem_flags flags, cl_int *errcode_ret)
-: MemObject((Context *)parent->parent(), flags, 0, errcode_ret), p_size(size),
-  p_offset(offset), p_parent(parent)
+: MemObject((Context *)parent->parent(), flags, 0, errcode_ret), p_offset(offset),
+  p_size(size), p_parent(parent)
 {
     if (size == 0)
     {
@@ -590,7 +590,7 @@ cl_int Image2D::imageInfo(cl_image_info param_name,
                          size_t *param_value_size_ret) const
 {
     void *value = 0;
-    int value_length = 0;
+    size_t value_length = 0;
     class Image3D *image3D = (class Image3D *)this;
 
     union {
diff --git a/src/core/program.cpp b/src/core/program.cpp
index f64b97a..eeb5f48 100644
--- a/src/core/program.cpp
+++ b/src/core/program.cpp
@@ -37,6 +37,10 @@ using namespace Coal;
 Program::Program(Context *ctx)
 : Object(Object::T_Program, ctx), p_type(Invalid), p_state(Empty)
 {
+    p_null_device_dependent.compiler = 0;
+    p_null_device_dependent.device = 0;
+    p_null_device_dependent.linked_module = 0;
+    p_null_device_dependent.program = 0;
 }
 
 Program::~Program()
@@ -57,7 +61,7 @@ void Program::setDevices(cl_uint num_devices, DeviceInterface * const*devices)
 {
     p_device_dependent.resize(num_devices);
 
-    for (int i=0; i<num_devices; ++i)
+    for (cl_uint i=0; i<num_devices; ++i)
     {
         DeviceDependent &dep = p_device_dependent[i];
 
@@ -70,24 +74,28 @@ void Program::setDevices(cl_uint num_devices, DeviceInterface * const*devices)
 
 Program::DeviceDependent &Program::deviceDependent(DeviceInterface *device)
 {
-    for (int i=0; i<p_device_dependent.size(); ++i)
+    for (size_t i=0; i<p_device_dependent.size(); ++i)
     {
         DeviceDependent &rs = p_device_dependent[i];
 
         if (rs.device == device || (!device && p_device_dependent.size() == 1))
             return rs;
     }
+
+    return p_null_device_dependent;
 }
 
 const Program::DeviceDependent &Program::deviceDependent(DeviceInterface *device) const
 {
-    for (int i=0; i<p_device_dependent.size(); ++i)
+    for (size_t i=0; i<p_device_dependent.size(); ++i)
     {
         const DeviceDependent &rs = p_device_dependent[i];
 
         if (rs.device == device || (!device && p_device_dependent.size() == 1))
             return rs;
     }
+
+    return p_null_device_dependent;
 }
 
 DeviceProgram *Program::deviceDependentProgram(DeviceInterface *device) const
@@ -128,14 +136,14 @@ Kernel *Program::createKernel(const std::string &name, cl_int *errcode_ret)
     Kernel *rs = new Kernel(this);
 
     // Add a function definition for each device
-    for (int i=0; i<p_device_dependent.size(); ++i)
+    for (size_t i=0; i<p_device_dependent.size(); ++i)
     {
         bool found = false;
         DeviceDependent &dep = p_device_dependent[i];
         const std::vector<llvm::Function *> &kernels = kernelFunctions(dep);
 
         // Find the one with the good name
-        for (int j=0; j<kernels.size(); ++j)
+        for (size_t j=0; j<kernels.size(); ++j)
         {
             llvm::Function *func = kernels[j];
 
@@ -177,7 +185,7 @@ std::vector<Kernel *> Program::createKernels(cl_int *errcode_ret)
     // Create the kernel for each function name
     // It returns an error if the signature is not the same for every device
     // or if the kernel isn't found on all the devices.
-    for (int i=0; i<kernels.size(); ++i)
+    for (size_t i=0; i<kernels.size(); ++i)
     {
         cl_int result = CL_SUCCESS;
         Kernel *kernel = createKernel(kernels[i]->getNameStr(), &result);
@@ -201,7 +209,7 @@ cl_int Program::loadSources(cl_uint count, const char **strings,
     p_source = std::string(embed_stdlib_h);
 
     // Merge all strings into one big one
-    for (int i=0; i<count; ++i)
+    for (cl_uint i=0; i<count; ++i)
     {
         size_t len = 0;
         const char *data = strings[i];
@@ -239,7 +247,7 @@ cl_int Program::loadBinaries(const unsigned char **data, const size_t *lengths,
     setDevices(num_devices, device_list);
 
     // Load the data
-    for (int i=0; i<num_devices; ++i)
+    for (cl_uint i=0; i<num_devices; ++i)
     {
         DeviceDependent &dep = deviceDependent(device_list[i]);
 
@@ -288,7 +296,7 @@ cl_int Program::build(const char *options,
         setDevices(num_devices, device_list);
     }
 
-    for (int i=0; i<num_devices; ++i)
+    for (cl_uint i=0; i<num_devices; ++i)
     {
         DeviceDependent &dep = deviceDependent(device_list[i]);
 
@@ -362,7 +370,7 @@ cl_int Program::build(const char *options,
         std::vector<std::string> api_s;     // Needed to keep valid data in api
         const std::vector<llvm::Function *> &kernels = kernelFunctions(dep);
 
-        for (int j=0; j<kernels.size(); ++j)
+        for (size_t j=0; j<kernels.size(); ++j)
         {
             std::string s = kernels[j]->getNameStr();
 
@@ -444,7 +452,7 @@ cl_int Program::info(cl_program_info param_name,
             break;
 
         case CL_PROGRAM_DEVICES:
-            for (int i=0; i<p_device_dependent.size(); ++i)
+            for (size_t i=0; i<p_device_dependent.size(); ++i)
             {
                 const DeviceDependent &dep = p_device_dependent[i];
 
@@ -464,7 +472,7 @@ cl_int Program::info(cl_program_info param_name,
             break;
 
         case CL_PROGRAM_BINARY_SIZES:
-            for (int i=0; i<p_device_dependent.size(); ++i)
+            for (size_t i=0; i<p_device_dependent.size(); ++i)
             {
                 const DeviceDependent &dep = p_device_dependent[i];
 
@@ -487,7 +495,7 @@ cl_int Program::info(cl_program_info param_name,
             if (!param_value || param_value_size < value_length)
                 return CL_INVALID_VALUE;
 
-            for (int i=0; i<p_device_dependent.size(); ++i)
+            for (size_t i=0; i<p_device_dependent.size(); ++i)
             {
                 const DeviceDependent &dep = p_device_dependent[i];
                 unsigned char *dest = binaries[i];
diff --git a/src/core/program.h b/src/core/program.h
index c8ae9c6..3101415 100644
--- a/src/core/program.h
+++ b/src/core/program.h
@@ -87,6 +87,7 @@ class Program : public Object
         };
 
         std::vector<DeviceDependent> p_device_dependent;
+        DeviceDependent p_null_device_dependent;
 
         void setDevices(cl_uint num_devices, DeviceInterface * const*devices);
         DeviceDependent &deviceDependent(DeviceInterface *device);
diff --git a/src/core/sampler.cpp b/src/core/sampler.cpp
index bdf0f14..8322c3b 100644
--- a/src/core/sampler.cpp
+++ b/src/core/sampler.cpp
@@ -118,7 +118,7 @@ cl_int Sampler::checkImageAvailability() const
 
         if (!image_support)
         {
-            std:free((void *)devices);
+            std::free((void *)devices);
             return CL_INVALID_OPERATION;
         }
     }
diff --git a/tests/test_commandqueue.cpp b/tests/test_commandqueue.cpp
index c8d9834..1fac3eb 100644
--- a/tests/test_commandqueue.cpp
+++ b/tests/test_commandqueue.cpp
@@ -334,7 +334,9 @@ START_TEST (test_events)
         "unable to get timing information about a profiling-enabled event"
     );
     fail_if(
-        !(timing_queued <= timing_submit <= timing_start <= timing_end),
+        !(timing_queued <= timing_submit &&
+          timing_submit <= timing_start &&
+          timing_start <= timing_end),
         "something went wrong with the timings : they are unordered"
     );
 
diff --git a/tests/test_kernel.cpp b/tests/test_kernel.cpp
index 98a2898..134b88f 100644
--- a/tests/test_kernel.cpp
+++ b/tests/test_kernel.cpp
@@ -14,7 +14,7 @@ static const char source[] =
     "    a[i] = simple_function(f) * b[i];\n"
     "}\n"
     "\n"
-    "__kernel void kernel2(__global int *buf) {\n"
+    "__kernel void kernel2(__global unsigned int *buf) {\n"
     "    size_t i = get_global_id(0);\n"
     "\n"
     "    buf[i % 256] = 2 * (i % 256);\n"
@@ -29,10 +29,9 @@ static void native_kernel(void *args)
     };
 
     struct ags *data = (struct ags *)args;
-    int i;
 
     // Not
-    for (int i=0; i<data->buffer_size; ++i)
+    for (size_t i=0; i<data->buffer_size; ++i)
     {
         data->buffer[i] = ~data->buffer[i];
     }
@@ -53,7 +52,7 @@ START_TEST (test_compiled_kernel)
     const char *src = source;
     size_t program_len = sizeof(source);
 
-    int buffer[256];
+    unsigned int buffer[256];
 
     result = clGetDeviceIDs(platform, CL_DEVICE_TYPE_DEFAULT, 1, &device, 0);
     fail_if(
@@ -143,7 +142,7 @@ START_TEST (test_compiled_kernel)
     );
 
     ok = true;
-    for (int i=0; i<local_size; ++i)
+    for (size_t i=0; i<local_size; ++i)
     {
         if (buffer[i] !=  2 * i)
         {
diff --git a/tests/test_program.cpp b/tests/test_program.cpp
index 15dcde8..0635af2 100644
--- a/tests/test_program.cpp
+++ b/tests/test_program.cpp
@@ -152,7 +152,6 @@ START_TEST (test_program_build_info)
     cl_int result;
 
     const char *src = program_source;
-    size_t program_len = sizeof(program_source);
 
     result = clGetDeviceIDs(platform, CL_DEVICE_TYPE_DEFAULT, 1, &device, 0);
     fail_if(