Using an algorithm in Cudafy doesn't work when it works on CPU?

Question

I have a working algorithm to perform a 2D perspective transform on an image.

The Algorithm is a follows:

private Bitmap RescaleImage(double TopLX, double TopLY, double TopRX, double TopRY, double LowLX, double LowLY, double LowRX, double LowRY, int width, int height)
{
    byte[] src_bmp = bmp.ToByteArray();
    byte[] dst_bmp = new byte[src_bmp.Length];


    for (int x = 0; x < width; x++)
    {
        for (int y = 0; y < height; y++)
        {
            /*
             * relative position
             */
            double rx = (double)x / width;
            double ry = (double)y / height;

            /*
             * get top and bottom position
             */
            double topX = TopLX + rx * (TopRX - TopLX);
            double topY = TopLY + rx * (TopRY - TopLY);
            double bottomX = LowLX + rx * (LowRX - LowLX);
            double bottomY = LowLY + rx * (LowRY - LowLY);

            /*
             * select center between top and bottom point
             */
            double centerX = topX + ry * (bottomX - topX);
            double centerY = topY + ry * (bottomY - topY);

            /*
             * store result
             */
            // get fractions
            double xf = centerX - (int)centerX;
            double yf = centerY - (int)centerY;

            // 4 colors - we're flipping sides so we can use the distance instead of inverting it later

            byte cTL0, cTL1, cTL2, cTL3, cTR0, cTR1, cTR2, cTR3, cLL0, cLL1, cLL2, cLL3, cLR0, cLR1, cLR2, cLR3;

            cTL0 = src_bmp[(((int)centerY + 1) * (width * 4)) + (((int)centerX + 1) * 4)];
            cTL1 = src_bmp[((((int)centerY + 1) * (width * 4)) + (((int)centerX + 1) * 4)) + 1];
            cTL2 = src_bmp[((((int)centerY + 1) * (width * 4)) + (((int)centerX + 1) * 4)) + 2];

            cTR0 = src_bmp[(((int)centerY + 0) * (width * 4)) + (((int)centerX + 1) * 4)];
            cTR1 = src_bmp[((((int)centerY + 0) * (width * 4)) + (((int)centerX + 1) * 4)) + 1];
            cTR2 = src_bmp[((((int)centerY + 0) * (width * 4)) + (((int)centerX + 1) * 4)) + 2];

            cLL0 = src_bmp[(((int)centerY + 1) * (width * 4)) + (((int)centerX + 0) * 4)];
            cLL1 = src_bmp[((((int)centerY + 1) * (width * 4)) + (((int)centerX + 0) * 4)) + 1];
            cLL2 = src_bmp[((((int)centerY + 1) * (width * 4)) + (((int)centerX + 0) * 4)) + 2];

            cLR0 = src_bmp[(((int)centerY + 0) * (width * 4)) + (((int)centerX + 0) * 4)];
            cLR1 = src_bmp[((((int)centerY + 0) * (width * 4)) + (((int)centerX + 0) * 4)) + 1];
            cLR2 = src_bmp[((((int)centerY + 0) * (width * 4)) + (((int)centerX + 0) * 4)) + 2];

            // 4 distances
            double dTL = Math.Sqrt(xf * xf + yf * yf);
            double dTR = Math.Sqrt((1 - xf) * (1 - xf) + yf * yf);
            double dLL = Math.Sqrt(xf * xf + (1 - yf) * (1 - yf));
            double dLR = Math.Sqrt((1 - xf) * (1 - xf) + (1 - yf) * (1 - yf));

            // 4 parts
            double factor = 1.0 / (dTL + dTR + dLL + dLR);
            dTL *= factor;
            dTR *= factor;
            dLL *= factor;
            dLR *= factor;

            // accumulate parts
            double r = dTL * (double)cTL0 + dTR * (double)cTR0 + dLL * (double)cLL0 + dLR * (double)cLR0;
            double g = dTL * (double)cTL1 + dTR * (double)cTR1 + dLL * (double)cLL1 + dLR * (double)cLR1;
            double b = dTL * (double)cTL2 + dTR * (double)cTR2 + dLL * (double)cLL2 + dLR * (double)cLR2;

            byte c0 = (byte)(r + 0.5);
            byte c1 = (byte)(g + 0.5);
            byte c2 = (byte)(b + 0.5);

            dst_bmp[(y * (width * 4)) + (x * 4)] = c0;
            dst_bmp[((y * (width * 4)) + (x * 4)) + 1] = c1;
            dst_bmp[((y * (width * 4)) + (x * 4)) + 2] = c2;

        }
    }

    Bitmap bmpOut = dst_bmp.ToBitmap(width, height);
    return bmpOut;
}

This works fine and the output is exactly what i want. however i have made a very subtle change to make it run on the GPU using Cudafy:

public void PerformPerspectiveCorrection(PointF TL, PointF TR, PointF LL, PointF LR)
{
    CheckIsSet();
    _gpu.Launch(Width, Height).PerspectiveCorrectionSingleOperation(_gdata.SourceImage, _gdata.ResultImage, TL.X, TL.Y, TR.X, TR.Y, LL.X, LL.Y, LR.X, LR.Y, Width, Height);
}

[Cudafy]
private static void PerspectiveCorrectionSingleOperation(GThread thread, byte[] src_bmp, byte[] dst_bmp, double TopLX, double TopLY, double TopRX, double TopRY, double LowLX, double LowLY, double LowRX, double LowRY, int width, int height)
{

    int x = thread.blockIdx.x;
    int y = thread.threadIdx.x;

    /*
     * relative position
     */
    double rx = (double)x / width;
    double ry = (double)y / height;

    /*
     * get top and bottom position
     */
    double topX = TopLX + rx * (TopRX - TopLX);
    double topY = TopLY + rx * (TopRY - TopLY);
    double bottomX = LowLX + rx * (LowRX - LowLX);
    double bottomY = LowLY + rx * (LowRY - LowLY);

    /*
     * select center between top and bottom point
     */
    double centerX = topX + ry * (bottomX - topX);
    double centerY = topY + ry * (bottomY - topY);

    /*
     * store result
     */
    // get fractions
    double xf = centerX - (int)centerX;
    double yf = centerY - (int)centerY;

    // 4 colors - we're flipping sides so we can use the distance instead of inverting it later

    byte cTL0, cTL1, cTL2, cTR0, cTR1, cTR2, cLL0, cLL1, cLL2, cLR0, cLR1, cLR2;

    cTL0 = src_bmp[(((int)centerY + 1) * (width * 4)) + (((int)centerX + 1) * 4)];
    cTL1 = src_bmp[((((int)centerY + 1) * (width * 4)) + (((int)centerX + 1) * 4)) + 1];
    cTL2 = src_bmp[((((int)centerY + 1) * (width * 4)) + (((int)centerX + 1) * 4)) + 2];

    cTR0 = src_bmp[(((int)centerY + 0) * (width * 4)) + (((int)centerX + 1) * 4)];
    cTR1 = src_bmp[((((int)centerY + 0) * (width * 4)) + (((int)centerX + 1) * 4)) + 1];
    cTR2 = src_bmp[((((int)centerY + 0) * (width * 4)) + (((int)centerX + 1) * 4)) + 2];

    cLL0 = src_bmp[(((int)centerY + 1) * (width * 4)) + (((int)centerX + 0) * 4)];
    cLL1 = src_bmp[((((int)centerY + 1) * (width * 4)) + (((int)centerX + 0) * 4)) + 1];
    cLL2 = src_bmp[((((int)centerY + 1) * (width * 4)) + (((int)centerX + 0) * 4)) + 2];

    cLR0 = src_bmp[(((int)centerY + 0) * (width * 4)) + (((int)centerX + 0) * 4)];
    cLR1 = src_bmp[((((int)centerY + 0) * (width * 4)) + (((int)centerX + 0) * 4)) + 1];
    cLR2 = src_bmp[((((int)centerY + 0) * (width * 4)) + (((int)centerX + 0) * 4)) + 2];

    // 4 distances
    double dTL = Math.Sqrt(xf * xf + yf * yf);
    double dTR = Math.Sqrt((1 - xf) * (1 - xf) + yf * yf);
    double dLL = Math.Sqrt(xf * xf + (1 - yf) * (1 - yf));
    double dLR = Math.Sqrt((1 - xf) * (1 - xf) + (1 - yf) * (1 - yf));

    // 4 parts
    double factor = 1.0 / (dTL + dTR + dLL + dLR);
    dTL *= factor;
    dTR *= factor;
    dLL *= factor;
    dLR *= factor;

    // accumulate parts
    double r = dTL * (double)cTL0 + dTR * (double)cTR0 + dLL * (double)cLL0 + dLR * (double)cLR0;
    double g = dTL * (double)cTL1 + dTR * (double)cTR1 + dLL * (double)cLL1 + dLR * (double)cLR1;
    double b = dTL * (double)cTL2 + dTR * (double)cTR2 + dLL * (double)cLL2 + dLR * (double)cLR2;

    byte c0 = (byte)(r + 0.5);
    byte c1 = (byte)(g + 0.5);
    byte c2 = (byte)(b + 0.5);

    dst_bmp[(y * (width * 4)) + (x * 4)] = c0;
    dst_bmp[((y * (width * 4)) + (x * 4)) + 1] = c1;
    dst_bmp[((y * (width * 4)) + (x * 4)) + 2] = c2;
}

The byte[] i get back is all 0's. I have tried directly applying a value (255) to all bytes in dst_bmp as well and it seems to only perform the operations for only one row of pixels (1280 bytes as the first row is 320px and there are 4 byte per px).

Any ideas? This is infuriating!

Is bmp in byte[] src_bmp = bmp.ToByteArray(); is a bitmap type variable? — Md Sifatul Islam, Jan 19 '17 at 07:56

score 2 · Accepted Answer · answered Jul 22 '15 at 12:43

Found the answer, wasn't to do with my algorithm at all! i was passing float values into the doubles for the Cudafy method. This cannot happen as the GPU is running the CUDA "blind" and thus will access the floats as if they are double without casting them before allocating the type of memory for the values.

Changed the double parameters to floats, worked like a charm.

Using an algorithm in Cudafy doesn't work when it works on CPU?

1 Answers1