can single work items kernels run in parallel on same device

Any type of kernel can run in parallel with another, as long as they are invoked in a separate queue, and no event is used to forcibly sentimentalize them; the key point here is that they must run in a different queue and you should not force the host to wait for each kernel execution separately using commands like clFlush() or clFinish(), or by waiting on events. You can, and probably should, wait for an event associated with each kernel invocation, or use clFinish() on every single queue you have, after invoking all the kernels in the host, to make sure all kernels have finished execution, to be then able to read the data back from the device.

I've tried everything I can think of but I still cannot get my kernel tasks to run in parallel.

__kernel void sumN1(const double start,
                    const double stop,
                    const double step,
                 __global double *z) {
  //// get index of the work item
  //int index = get_global_id(0);
  //init result
  double sum = 0.0;
  for (double i=start; i<=stop; i+=step) {
    sum += i;
  }
  z = sum;
} // sumN1
... eight identical kernels ('cept the kernel name) - in the same .cl file
__kernel void sumN8(const double start,
                    const double stop,
                    const double step,
                 __global double *z) {
  //// get index of the work item
  //int index = get_global_id(0);
  //init result
  double sum = 0.0;
  for (double i=start; i<=stop; i+=step) {
    sum += i;
  }
  z = sum;
} // sumN8

////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
//
// main - sumN
//
// - sumN num 
//
int main(int argc, char *argv) {
      cl_device_id  accDeviceID;                //device ID of first acceleator device
    cl_platform_id  accPlatformID;              //platform ID of platform w/first accelerator device
            cl_mem  answer;                  //array to collect results
       std::string  binary_file;                //name of OpenCL program
        cl_context  context;                    //OpenCL context for this application
    cl_device_type  deviceType;                 //type of a device (CPU, GPU, ACCELERATOR)
          cl_event  event;                      //wait event synchronization handle used by OpenCL API
              bool  foundintel = false;         //indicates that Intel FPGA card was found
              char  info_text;   //value of some returned text information
              bool  isACC = false;              //flag to remember that an accelerator has been found
         cl_kernel  kernel;                  //OpenCL kernal for this applicaiton
               int  kerns = 0;                  //number of kernels to use
           cl_uint  numDevices;                 //number of OpenCL computing devices for a platform
           cl_uint  numPlatforms;               //number of OpenCL platforms (typically 1)
            double  number = 0.0;               //number to compute sum to
        cl_program  program;                    //OpenCL program for this application
  cl_command_queue  queue;                   //OpenCL command queue for this application
            double  result = { 0.0 };        //result of the summation computation
            size_t  size;                       //size of returned information from OpenCL API
            double  start = 1.0;                //start of summing
            cl_int  status;                     //return code used by OpenCL API
            double  step  = 1.0;                //step  of summing
            double  stop  = 1.0;                //end   of summing
            cl_int  task_done;                  //info from event query
          cl_event  task_event;              //events from tasks
.... some code omitted here that handled input args and platform, device setup
  ////////////////////////////////////////
  // OpenCL context
  context = clCreateContext(NULL, 1, &accDeviceID, NULL, NULL, &status);
  exitOnFail(status, "create context");
  ////////////////////////////////////////
  // OpenCL command queue
  for ( int kz=0; kz<kerns; kz++) {
    queue = clCreateCommandQueue(context, accDeviceID, 0, &status);
    exitOnFail(status, "create command queue");
  }
  ////////////////////////////////////////
  // Create the program for all device. Use the first device as the
  // representative device (assuming all device are of the same type).
  binary_file = getBoardBinaryFile("sumN", accDeviceID);
  program = createProgramFromBinary(context, binary_file.c_str(), &accDeviceID, 1);
  ////////////////////////////////////////
  // Build the program that was just created.
  status = clBuildProgram(program, 0, NULL, "", NULL, NULL);
  exitOnFail(status, "Failed to build program");
  const double start_time = getCurrentTimestamp();
  ////////////////////////////////////////
  // create the kernel
  // Create the kernel - name passed in here must match kernel name in the
  // original CL file, that was compiled into an AOCX file using the AOC tool
  char kernel_name = "sumNx";  // Kernel name, as defined in the CL file
  for ( int kz=0; kz<kerns; kz++) {
    sprintf(kernel_name, "sumN%d", kz+1);  // generate the Kernel name, as defined in the CL file
    kernel = clCreateKernel(program, kernel_name, &status);
    exitOnFail(status, "Failed to create kernel");
    // Set the kernel argument (argument 0)
    status = clSetKernelArg(kernel, 0, sizeof(cl_double), &start);
    exitOnFail(status, "Failed to set kernel arg 0");
    // Set the kernel argument (argument 1)
    status = clSetKernelArg(kernel, 1, sizeof(cl_double), &stop);
    exitOnFail(status, "Failed to set kernel arg 0");
    // Set the kernel argument (argument 2)
    status = clSetKernelArg(kernel, 2, sizeof(cl_double), &step);
    exitOnFail(status, "Failed to set kernel arg 0");
    // last OpenCL argument: memory buffer object for result
    answer = clCreateBuffer(context, CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR, sizeof(cl_double), &result, &status);
    exitOnFail(status, "create buffer for answer");
    // set 4th argument
    status = clSetKernelArg(kernel, 3, sizeof(cl_mem), &answer);
    exitOnFail(status, "set kernel argument answer");
  }
  // Launch the kernels
  for ( int kz=0; kz<kerns; kz++) {
    status = clEnqueueTask(queue, kernel, 0, NULL, &task_event);
    exitOnFail(status, "Failed to launch kernel");
  }
  int total_done = 0;
  int its_done = { 0 };
  while (total_done < kerns) {
    for ( int kz=0; kz<kerns; kz++) {
      if ( its_done == 0 ) {
        status = clGetEventInfo(task_event, CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(cl_int), &task_done, NULL);
        if (task_done == CL_COMPLETE) {
          printf("Task:%d complete (%0.3f ms)\n", kz, (getCurrentTimestamp() - start_time) * 1.0e3 );
          total_done++;
          its_done = 1;
        } else {
          //printf("Task:%d incomplete\n", kz);
        }
      } // if kz task not done
    } // foreach task event
  } // wait for kernels to complete
//// Wait for command queue to complete pending events
//for ( int kz=0; kz<kerns; kz++) {
//  status = clFinish(queue);
//  exitOnFail(status, "Failed to finish");
//}
  const double end_time = getCurrentTimestamp();
  // Wall-clock time taken.
  printf("\nTime: %0.3f ms (%0.3f ms / kernel)\n", (end_time - start_time) * 1e3, (end_time - start_time) * 1e3 / (double)kerns );
  for ( int kz=0; kz<kerns; kz++) {
    printf("Sum 0-%f (step %f) = %f\n", number, step, result);
  }
  // Free the resources allocated
  cleanup();
  if(kernel) {
    for ( int kz=0; kz<kerns; kz++) {
      clReleaseKernel(kernel);
    }
  }
  if(program) {
    clReleaseProgram(program);
  }
  if(queue) {
    for ( int kz=0; kz<kerns; kz++) {
      clReleaseCommandQueue(queue);
    }
  }
  if(context) {
    clReleaseContext(context);
  }
  exit(0);
} // main

Try putting "start_time = getCurrentTimestamp();" before the kernel invocation loop. I have a feeling your kernel run time is so short that total time is being dominated by the clCreateBuffer() call. Note that on most hardware, using host pointer will result in the OpenCL runtime actually allocating and transferring the whole buffer to device memory. Also considering increasing your input size so that your total run time is at least a few seconds.

I did try the longer runs. The result is the same (i.e. no apparent parallelism).

And what about moving the "start_time = getCurrentTimestamp()" call?

  const double start_time = getCurrentTimestamp();
  // Launch the kernels
  for ( int kz=0; kz<kerns; kz++) {
    status = clEnqueueTask(queue, kernel, 0, NULL, &task_event);
    exitOnFail(status, "Failed to launch kernel");
  }

I cannot really think of anything else, and I don't see any particular issues in your host code.

What about:

Maybe a suggestion, not sure if it is an issue, but you could try using CL_MEM_COPY_HOST_PTR instead of CL_MEM_USE_HOST_PTR which will allow multiple copies of the input data to be generated for each cl_mem object rather than having them all point to the same chunk of allocated memory. Also adding the 'restrict' flag to your global variables in the kernel to let it know that no other pointers to the same data are modifying the data. You will need to do an enqueue read buffer to get the data back out since they aren't mapped.