1 Reply Latest reply on Nov 16, 2012 3:36 PM by binying

    OpenCL program compilation failure

    dikobraz

      OpenCL reports an error, trying to compile this kernel:

       

      __constant sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_NEAREST;

       

      __kernel

      void raycastFine(__read_only  image2d_t    imgIn,

                       float3                    someParam1,

                       float3                    someParam2,

                       int3                      someParam3,

                       int3                      someParam4,

                       __write_only image2d_t    imgOut,

                       uint2                     size)

      {

          int2 tcoord;

          tcoord.x = get_global_id(0);

          tcoord.y = get_global_id(1);

          if (tcoord.x >= size.x || tcoord.y >= size.y) {

              return;

          }

         

          float4 t = read_imagef(imgIn, sampler, tcoord);

          if (t.w == 0.0f)

          {

              write_imagef(imgOut, tcoord, 0.0f);

              return;

          }

      }

       

      In AMD KernelAnalyzer 2:

       

      ========== Build: started ==========

      Error: clBuildProgram had an unhandled exception.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

      Error: NULL BuiltProgram Handle.

       

      ========== Build: 0 of 17 succeeded ==========

       

      And by just changing order of formal parameters this can be fixed. Am I missing something?

       

      __constant sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_NEAREST;

       

      __kernel

      void raycastFine(__read_only  image2d_t    imgIn,

                       __write_only image2d_t    imgOut,

                       float3                    someParam1,

                       float3                    someParam2,

                       int3                      someParam3,

                       int3                      someParam4,

                       uint2                     size)

      {

          int2 tcoord;

          tcoord.x = get_global_id(0);

          tcoord.y = get_global_id(1);

          if (tcoord.x >= size.x || tcoord.y >= size.y) {

              return;

          }

         

          float4 t = read_imagef(imgIn, sampler, tcoord);

          if (t.w == 0.0f)

          {

              write_imagef(imgOut, tcoord, 0.0f);

              return;

          }

      }

       

      P.S: Where to post bug report for OpenCL?

        • Re: OpenCL program compilation failure
          binying

          Interestingly, this compiles well in kernelAnalyzer.

           

          Note that I changed "write_imagef(imgOut, tcoord, 0.0f);" to "write_imagef(imgOut, tcoord, t.w);" where t.w == 0.0f.

               


           

          ------------------------------------------------------------

          // Enter your kernel in this window

           

           

           

           

          __constant sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_NEAREST;

           

          __kernel

          void raycastFine(__read_only  image2d_t    imgIn,

                           float3                    someParam1,

                           float3                    someParam2,

                           int3                      someParam3,

                           int3                      someParam4,

                           __write_only image2d_t    imgOut,

                           uint2                     size)

          {

              int2 tcoord;

              tcoord.x = get_global_id(0);

              tcoord.y = get_global_id(1);

              if (tcoord.x >= size.x || tcoord.y >= size.y) {

                  return;

              }

             float4 t = read_imagef(imgIn, sampler, tcoord);

             // float4 t = read_imagef(imgIn, sampler, tcoord);

              if (t.w == 0.0f)

              {

                  //write_imagef(imgOut, tcoord, 0.0f);

                   write_imagef(imgOut, tcoord,t.w);

                  return;

              }

          }