diff --git a/CMakeLists.txt b/CMakeLists.txt
index 7b80ecc..4e15ca2 100644
--- a/CMakeLists.txt
+++ b/CMakeLists.txt
@@ -10,6 +10,7 @@ message(STATUS "OpenCL CXX includes: ${OPENCL_HAS_CPP_BINDINGS}")
 message(STATUS "OpenCL libraries: ${OPENCL_LIBRARIES}")
 
 add_executable(${PROJECT_NAME} "main.c")
+configure_file("saxpy.cl" "cl_kernels/saxpy.cl" COPYONLY)
 
 target_link_libraries(${PROJECT_NAME} ${OPENCL_LIBRARIES})
 include_directories(${PROJECT_NAME} ${OPENCL_INCLUDE_DIR})
diff --git a/main.c b/main.c
index f02d7b7..b528e54 100644
--- a/main.c
+++ b/main.c
@@ -2,8 +2,50 @@
 #include <CL/cl.h>
 #include <stdlib.h>
 #include <stdint.h>
+#include <string.h>
 
-int main()
+
+void check_error(cl_int error)
+{
+	if (error != CL_SUCCESS) {
+		printf("OpenCL call failed with error %d\n", (int)error);
+		exit(error);
+	}
+}
+
+void load_kernel_from_file(char *file, char **src)
+{
+	long fsize;
+	FILE *f = fopen(file, "rb");
+
+	fseek(f, 0, SEEK_END);
+	fsize = ftell(f);
+	fseek(f, 0, SEEK_SET);  //same as rewind(f);
+	*src = (char *)malloc(fsize + 1);
+	fread(*src, fsize, 1, f);
+	fclose(f);
+	(*src)[fsize] = 0;
+}
+
+cl_program create_program(char *src, cl_context context)
+{
+	cl_program program;
+	size_t lengths[1];
+	cl_int error = 0;
+
+	lengths[0] = strlen(src);
+	printf("strlen: %u\n", (unsigned int)lengths[0]);
+	program = clCreateProgramWithSource(context, 1, (const char **)&src,
+					    lengths, &error);
+	check_error(error);
+
+	return program;
+}
+
+float test_a[] = {21.321, 213.213, 3213.23};
+float test_b[] = {21.21, 13.213, 17.2553};
+
+int main(void)
 {
 
 	cl_int error;
@@ -13,31 +55,37 @@ int main()
 	cl_uint device_id_count = 0;
 	cl_context_properties context_properties[4];
 	cl_context context;
+	cl_program program;
+	cl_kernel kernel;
+	cl_mem a_buff, b_buff;
+	char *source_code = NULL;
 
-	clGetPlatformIDs (0, NULL, &platform_id_count);
+	clGetPlatformIDs(0, NULL, &platform_id_count);
 	if (platform_id_count == 0)
 		return -1;
 
-	platform_ids = (cl_platform_id*) malloc(platform_id_count *
+	platform_ids = (cl_platform_id *)malloc(platform_id_count *
 						sizeof(cl_platform_id));
 	if (platform_ids == NULL)
 		return -1;
 
-	clGetPlatformIDs (platform_id_count, platform_ids, NULL);
+	clGetPlatformIDs(platform_id_count, platform_ids, NULL);
 
 	printf("Platforms available: %u\n", (unsigned int)platform_id_count);
 
 
-	clGetDeviceIDs (platform_ids[0], CL_DEVICE_TYPE_ALL, 0, NULL,
+	clGetDeviceIDs(platform_ids[0], CL_DEVICE_TYPE_ALL, 0, NULL,
 			&device_id_count);
-	printf("Device in Platform 0 count: %u\n", (unsigned int)device_id_count);
+	printf("Device in Platform 0 count: %u\n",
+	       (unsigned int)device_id_count);
 
 	if (device_id_count == 0)
 		return -1;
 
-	device_ids = (cl_device_id*) malloc(sizeof(cl_device_id)*device_id_count);
+	device_ids = (cl_device_id *) malloc(sizeof(cl_device_id)
+					    *device_id_count);
 
-	clGetDeviceIDs (platform_ids[0], CL_DEVICE_TYPE_ALL, device_id_count,
+	clGetDeviceIDs(platform_ids[0], CL_DEVICE_TYPE_ALL, device_id_count,
 			device_ids, NULL);
 
 	context_properties[0] = CL_CONTEXT_PLATFORM;
@@ -45,9 +93,32 @@ int main()
 	context_properties[2] = context_properties[3] = 0;
 
 
-	context = clCreateContext (context_properties, device_id_count,
+	context = clCreateContext(context_properties, device_id_count,
 				   device_ids, NULL,
 				   NULL, &error);
+	check_error(error);
+
+	load_kernel_from_file("cl_kernels/saxpy.cl", &source_code);
+	program = create_program(source_code, context);
+
+	/* Compile the source code */
+	check_error(clBuildProgram(program, device_id_count,
+				   device_ids, NULL, NULL, NULL));
+
+
+	kernel = clCreateKernel(program, "SAXPY", &error);
+
+
+	/* Create memory buffers for data */
+	a_buff = clCreateBuffer(context, CL_MEM_READ_ONLY |
+				CL_MEM_COPY_HOST_PTR,
+				sizeof(test_a), test_a, &error);
+	check_error(error);
+
+	b_buff = clCreateBuffer(context, CL_MEM_READ_WRITE |
+				CL_MEM_COPY_HOST_PTR,
+				sizeof(test_b), test_b, &error);
+	check_error(error);
 
 	return 0;
 }
diff --git a/saxpy.cl b/saxpy.cl
new file mode 100644
index 0000000..2a661a0
--- /dev/null
+++ b/saxpy.cl
@@ -0,0 +1,5 @@
+__kernel void SAXPY (__global float* x, __global float* y, float a)
+{
+	const int i = get_global_id (0);
+	y [i] += a * x [i];
+}