Home My Page Projects Code Snippets Project Openings diderot
Summary Activity Tracker Tasks SCM

SCM Repository

[diderot] View of /branches/vis12-cl/src/compiler/cl-target/fragments/c-run.in
ViewVC logotype

View of /branches/vis12-cl/src/compiler/cl-target/fragments/c-run.in

Parent Directory Parent Directory | Revision Log Revision Log


Revision 3137 - (download) (annotate)
Thu Mar 26 14:55:52 2015 UTC (3 years, 11 months ago) by jhr
File size: 2140 byte(s)
working on OpenCL scheduling
//! Run the Diderot program (OpenCL version)
//! \param wrld the world-state of the Diderot program
//! \param maxNSteps the limit on the number of super steps; 0 means unlimited
//! \return the number of steps taken; if an error is encountered during execution,
//!   then this function returns -1.
uint32_t @PREFIX@Run (@PREFIX@World_t *wrld, uint32_t maxNSteps)
{
    if (maxNSteps == 0) maxNSteps = 0xffffffff;  // essentially unlimited

  // work sizes for the program
    size_t globalWorkSize[2];
    size_t localWorkSize[2]; 
    globalWorkSize[0] = wrld->device->numCUs * wrld->device->cuWidth; 
    globalWorkSize[1] = wrld->nWorkers; 
    localWorkSize[0] = wrld->device->cuWidth;
    localWorkSize[1] = 1;

  // iterate until all strands are stable
    uint32_t nSteps = 0;
    uint32_t nActive = wrld->numActive;
    while ((nActive > 0) && (nSteps < maxNSteps)) {
        nSteps++;
      // Run the update kernel on all strands 
        sts = clEnqueueNDRangeKernel(wrld->cmdQ, wrld->UpdateKern, 2, NULL,
                globalWorkSize, localWorkSize, 0, NULL, NULL); 
        if (CheckOCLStatus (wrld, sts, "error in executing update kernel"))
	    return -1;

        sts = clEnqueueTask(wrld->cmdQ, wrld->SchedKern, 0, NULL, NULL); 
        if (CheckOCLStatus (wrld, sts, "error in executing scheduler update kernel before compaction"))
	    return -1;
 
      // Run the compaction kernel on all strands 
        sts = clEnqueueNDRangeKernel(wrld->cmdQ, wrld->CompactKern, 2, NULL,
                globalWorkSize, localWorkSize, 0, NULL, NULL);
        if (CheckOCLStatus (wrld, sts, "error executing compaction kernel"))
	    return -1;

        sts = clEnqueueTask(wrld->cmdQ,wrld->SchedKern, 0, NULL, NULL); 
        if (CheckOCLStatus (wrld, sts, "error in executing scheduler update kernel after compaction"))
	    return -1;

        sts = clEnqueueReadBuffer(wrld->cmdQ, kernelArgs.schedMem, CL_TRUE, 0, sizeof(SchedState_t),
                &scheduler, 0, NULL, NULL); 
        if (CheckOCLStatus (wrld, sts, "error reading back scheduler information"))
	    return -1;
    }

    return nSteps;

} // @PREFIX@Run

root@smlnj-gforge.cs.uchicago.edu
ViewVC Help
Powered by ViewVC 1.0.0