--- a/fft_fixstart/main.cc	Mon Feb 04 03:43:17 2013 +0900
+++ b/fft_fixstart/main.cc	Tue Feb 05 13:32:08 2013 +0900
@@ -104,19 +104,19 @@
 
     /* Reverse bit ordering */
     setWorkSize(gws, lws, n, n);
-    ret = clEnqueueNDRangeKernel(queue, brev, 2, NULL, gws, lws, 0, NULL, NULL);
+    ret = clEnqueueTask(queue, brev, 0, NULL, NULL);
 
     /* Perform Butterfly Operations*/
     setWorkSize(gws, lws, n/2, n);
     for (iter=1; iter <= m; iter++) {
         ret = clSetKernelArg(bfly, 4, sizeof(cl_int), (void *)&iter);
-        ret = clEnqueueNDRangeKernel(queue, bfly, 2, NULL, gws, lws, 0, NULL, &kernelDone);
+        ret = clEnqueueTask(queue, bfly, 0, NULL, NULL);
         ret = clWaitForEvents(1, &kernelDone);
     }
 
     if (direction == inverse) {
         setWorkSize(gws, lws, n, n);
-        ret = clEnqueueNDRangeKernel(queue, norm, 2, NULL, gws, lws, 0, NULL, &kernelDone);
+        ret = clEnqueueTask(queue, norm, 0, NULL, NULL);
         ret = clWaitForEvents(1, &kernelDone);
     }
 
@@ -137,8 +137,10 @@
             filename = argv[i+1];
         } else if (strcmp(argv[i], "-cpu") == 0) {
             device_type = CL_DEVICE_TYPE_CPU;
+            printf("cpu!\n");
         } else if (strcmp(argv[i], "-gpu") == 0) {
             device_type = CL_DEVICE_TYPE_GPU;
+            printf("gpu!\n");
         }
     }
     if ( (argc == 1)||(filename==0)) {
@@ -225,7 +227,8 @@
     st_time = getTime();
     /* Get platform/device  */
     ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
-    ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_DEFAULT, 1, &device_id, &ret_num_devices);
+
+    ret = clGetDeviceIDs( platform_id, device_type, 1, &device_id, &ret_num_devices);
 
     /* Create OpenCL context */
     context = clCreateContext(NULL, 1, &device_id, NULL, NULL, &ret);
@@ -266,7 +269,7 @@
     ret = clSetKernelArg(sfac, 0, sizeof(cl_mem), (void *)&wmobj);
     ret = clSetKernelArg(sfac, 1, sizeof(cl_int), (void *)&n);
     setWorkSize(gws, lws, n/2, 1);
-    ret = clEnqueueNDRangeKernel(queue, sfac, 1, NULL, gws, lws, 0, NULL, NULL);
+    ret = clEnqueueTask(queue, sfac, 0, NULL, NULL);
 
     /* Butterfly Operation */
     fftCore(rmobj, xmobj, wmobj, m, forward);
@@ -276,7 +279,7 @@
     ret = clSetKernelArg(trns, 1, sizeof(cl_mem), (void *)&rmobj);
     ret = clSetKernelArg(trns, 2, sizeof(cl_int), (void *)&n);
     setWorkSize(gws, lws, n, n);
-    ret = clEnqueueNDRangeKernel(queue, trns, 2, NULL, gws, lws, 0, NULL, NULL);
+    ret = clEnqueueTask(queue, trns, 0, NULL, NULL);
 
     /* Butterfly Operation */
     fftCore(rmobj, xmobj, wmobj, m, forward);
@@ -287,7 +290,7 @@
     ret = clSetKernelArg(hpfl, 1, sizeof(cl_int), (void *)&n);
     ret = clSetKernelArg(hpfl, 2, sizeof(cl_int), (void *)&radius);
     setWorkSize(gws, lws, n, n);
-    ret = clEnqueueNDRangeKernel(queue, hpfl, 2, NULL, gws, lws, 0, NULL, NULL);
+    ret = clEnqueueTask(queue, hpfl, 0, NULL, NULL);
 
     /* Inverse FFT */
 
@@ -298,7 +301,7 @@
     ret = clSetKernelArg(trns, 0, sizeof(cl_mem), (void *)&rmobj);
     ret = clSetKernelArg(trns, 1, sizeof(cl_mem), (void *)&xmobj);
     setWorkSize(gws, lws, n, n);
-    ret = clEnqueueNDRangeKernel(queue, trns, 2, NULL, gws, lws, 0, NULL, NULL);
+    ret = clEnqueueTask(queue, trns, 0, NULL, NULL);
 
     /* Butterfly Operation */