Merge remote-tracking branch 'origin/released'

Conflicts:
	apps/xwininfo/configure.ac
	apps/xwininfo/xwininfo.c
	libX11/configure.ac
	libX11/specs/libX11/AppC.xml
	libX11/specs/libX11/AppD.xml
	libX11/specs/libX11/CH03.xml
	libX11/specs/libX11/CH04.xml
	libX11/specs/libX11/CH05.xml
	libX11/specs/libX11/CH06.xml
	libX11/specs/libX11/CH07.xml
	libX11/specs/libX11/CH08.xml
	libX11/specs/libX11/CH09.xml
	libX11/specs/libX11/CH11.xml
	libX11/specs/libX11/CH12.xml
	libX11/specs/libX11/CH13.xml
	libX11/specs/libX11/CH14.xml
	libX11/specs/libX11/CH15.xml
	libX11/specs/libX11/CH16.xml
	libxcb/configure.ac
	libxcb/src/c_client.py
	libxcb/src/xcb_auth.c
	libxcb/src/xcb_util.c
	mesalib/common.py
	mesalib/configs/linux-dri
	mesalib/docs/GL3.txt
	mesalib/docs/download.html
	mesalib/docs/install.html
	mesalib/include/GL/internal/dri_interface.h
	mesalib/scons/custom.py
	mesalib/scons/gallium.py
	mesalib/src/gallium/auxiliary/util/u_math.h
	mesalib/src/gallium/auxiliary/util/u_vbuf_mgr.c
	mesalib/src/glsl/ast_function.cpp
	mesalib/src/glsl/ast_to_hir.cpp
	mesalib/src/glsl/glcpp/glcpp-parse.y
	mesalib/src/glsl/glsl_parser_extras.cpp
	mesalib/src/glsl/glsl_parser_extras.h
	mesalib/src/glsl/ir.cpp
	mesalib/src/glsl/ir.h
	mesalib/src/glsl/ir_clone.cpp
	mesalib/src/glsl/ir_print_visitor.cpp
	mesalib/src/glsl/ir_validate.cpp
	mesalib/src/glsl/linker.cpp
	mesalib/src/glsl/main.cpp
	mesalib/src/glsl/opt_tree_grafting.cpp
	mesalib/src/mapi/glapi/gen/Makefile
	mesalib/src/mapi/glapi/gen/gl_XML.py
	mesalib/src/mapi/glapi/gen/gl_table.py
	mesalib/src/mapi/glapi/glapi_mapi_tmp.h
	mesalib/src/mapi/glapi/glapi_sparc.S
	mesalib/src/mapi/glapi/glapi_x86-64.S
	mesalib/src/mapi/glapi/glapi_x86.S
	mesalib/src/mapi/glapi/glapitemp.h
	mesalib/src/mapi/glapi/glprocs.h
	mesalib/src/mapi/mapi/u_thread.c
	mesalib/src/mapi/mapi/u_thread.h
	mesalib/src/mesa/SConscript
	mesalib/src/mesa/drivers/common/driverfuncs.c
	mesalib/src/mesa/drivers/windows/gldirect/dglcontext.c
	mesalib/src/mesa/main/api_arrayelt.c
	mesalib/src/mesa/main/bufferobj.c
	mesalib/src/mesa/main/compiler.h
	mesalib/src/mesa/main/dd.h
	mesalib/src/mesa/main/dlist.c
	mesalib/src/mesa/main/enums.c
	mesalib/src/mesa/main/es_generator.py
	mesalib/src/mesa/main/fbobject.c
	mesalib/src/mesa/main/imports.h
	mesalib/src/mesa/main/mtypes.h
	mesalib/src/mesa/main/pbo.c
	mesalib/src/mesa/main/remap_helper.h
	mesalib/src/mesa/main/shared.c
	mesalib/src/mesa/main/texgetimage.c
	mesalib/src/mesa/main/teximage.c
	mesalib/src/mesa/main/uniforms.c
	mesalib/src/mesa/program/ir_to_mesa.cpp
	mesalib/src/mesa/program/prog_optimize.c
	mesalib/src/mesa/program/register_allocate.c
	mesalib/src/mesa/program/register_allocate.h
	mesalib/src/mesa/sources.mak
	mesalib/src/mesa/state_tracker/st_cb_bufferobjects.c
	mesalib/src/mesa/state_tracker/st_cb_texture.c
	mesalib/src/mesa/state_tracker/st_texture.h
	mesalib/src/mesa/swrast/s_context.c
	mesalib/src/mesa/tnl/t_draw.c
	mesalib/src/mesa/vbo/vbo_exec_array.c
	mesalib/src/mesa/vbo/vbo_save_api.c
	pixman/pixman/pixman-arm-common.h
	pixman/pixman/pixman-fast-path.c
	pixman/pixman/pixman-inlines.h
	pixman/test/Makefile.am
	pixman/test/utils.c
	pixman/test/utils.h
	xorg-server/configure.ac
	xorg-server/fb/fbblt.c
	xorg-server/hw/xquartz/darwin.c
	xorg-server/xkeyboard-config/symbols/de

1 files changed, 343 insertions, 331 deletions

diff --git a/mesalib/src/mesa/swrast/s_aatritemp.h b/mesalib/src/mesa/swrast/s_aatritemp.h
index 4136df3a7..77b3ae6ec 100644
--- a/mesalib/src/mesa/swrast/s_aatritemp.h
+++ b/mesalib/src/mesa/swrast/s_aatritemp.h
@@ -1,331 +1,343 @@
-/*

- * Mesa 3-D graphics library

- * Version:  7.0.3

- *

- * Copyright (C) 1999-2007  Brian Paul   All Rights Reserved.

- *

- * Permission is hereby granted, free of charge, to any person obtaining a

- * copy of this software and associated documentation files (the "Software"),

- * to deal in the Software without restriction, including without limitation

- * the rights to use, copy, modify, merge, publish, distribute, sublicense,

- * and/or sell copies of the Software, and to permit persons to whom the

- * Software is furnished to do so, subject to the following conditions:

- *

- * The above copyright notice and this permission notice shall be included

- * in all copies or substantial portions of the Software.

- *

- * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS

- * OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,

- * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL

- * BRIAN PAUL BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER LIABILITY, WHETHER IN

- * AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR IN

- * CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE.

- */

-

-

-/*

- * Antialiased Triangle Rasterizer Template

- *

- * This file is #include'd to generate custom AA triangle rasterizers.

- * NOTE: this code hasn't been optimized yet.  That'll come after it

- * works correctly.

- *

- * The following macros may be defined to indicate what auxillary information

- * must be copmuted across the triangle:

- *    DO_Z         - if defined, compute Z values

- *    DO_ATTRIBS   - if defined, compute texcoords, varying, etc.

- */

-

-/*void triangle( struct gl_context *ctx, GLuint v0, GLuint v1, GLuint v2, GLuint pv )*/

-{

-   const SWcontext *swrast = SWRAST_CONTEXT(ctx);

-   const GLfloat *p0 = v0->attrib[FRAG_ATTRIB_WPOS];

-   const GLfloat *p1 = v1->attrib[FRAG_ATTRIB_WPOS];

-   const GLfloat *p2 = v2->attrib[FRAG_ATTRIB_WPOS];

-   const SWvertex *vMin, *vMid, *vMax;

-   GLint iyMin, iyMax;

-   GLfloat yMin, yMax;

-   GLboolean ltor;

-   GLfloat majDx, majDy;  /* major (i.e. long) edge dx and dy */

-   

-   SWspan span;

-   

-#ifdef DO_Z

-   GLfloat zPlane[4];

-#endif

-   GLfloat rPlane[4], gPlane[4], bPlane[4], aPlane[4];

-#if defined(DO_ATTRIBS)

-   GLfloat attrPlane[FRAG_ATTRIB_MAX][4][4];

-   GLfloat wPlane[4];  /* win[3] */

-#endif

-   GLfloat bf = SWRAST_CONTEXT(ctx)->_BackfaceCullSign;

-   

-   (void) swrast;

-

-   INIT_SPAN(span, GL_POLYGON);

-   span.arrayMask = SPAN_COVERAGE;

-

-   /* determine bottom to top order of vertices */

-   {

-      GLfloat y0 = v0->attrib[FRAG_ATTRIB_WPOS][1];

-      GLfloat y1 = v1->attrib[FRAG_ATTRIB_WPOS][1];

-      GLfloat y2 = v2->attrib[FRAG_ATTRIB_WPOS][1];

-      if (y0 <= y1) {

-	 if (y1 <= y2) {

-	    vMin = v0;   vMid = v1;   vMax = v2;   /* y0<=y1<=y2 */

-	 }

-	 else if (y2 <= y0) {

-	    vMin = v2;   vMid = v0;   vMax = v1;   /* y2<=y0<=y1 */

-	 }

-	 else {

-	    vMin = v0;   vMid = v2;   vMax = v1;  bf = -bf; /* y0<=y2<=y1 */

-	 }

-      }

-      else {

-	 if (y0 <= y2) {

-	    vMin = v1;   vMid = v0;   vMax = v2;  bf = -bf; /* y1<=y0<=y2 */

-	 }

-	 else if (y2 <= y1) {

-	    vMin = v2;   vMid = v1;   vMax = v0;  bf = -bf; /* y2<=y1<=y0 */

-	 }

-	 else {

-	    vMin = v1;   vMid = v2;   vMax = v0;   /* y1<=y2<=y0 */

-	 }

-      }

-   }

-

-   majDx = vMax->attrib[FRAG_ATTRIB_WPOS][0] - vMin->attrib[FRAG_ATTRIB_WPOS][0];

-   majDy = vMax->attrib[FRAG_ATTRIB_WPOS][1] - vMin->attrib[FRAG_ATTRIB_WPOS][1];

-

-   /* front/back-face determination and cullling */

-   {

-      const GLfloat botDx = vMid->attrib[FRAG_ATTRIB_WPOS][0] - vMin->attrib[FRAG_ATTRIB_WPOS][0];

-      const GLfloat botDy = vMid->attrib[FRAG_ATTRIB_WPOS][1] - vMin->attrib[FRAG_ATTRIB_WPOS][1];

-      const GLfloat area = majDx * botDy - botDx * majDy;

-      /* Do backface culling */

-      if (area * bf < 0 || area == 0 || IS_INF_OR_NAN(area))

-	 return;

-      ltor = (GLboolean) (area < 0.0F);

-

-      span.facing = area * swrast->_BackfaceSign > 0.0F;

-   }

-

-   /* Plane equation setup:

-    * We evaluate plane equations at window (x,y) coordinates in order

-    * to compute color, Z, fog, texcoords, etc.  This isn't terribly

-    * efficient but it's easy and reliable.

-    */

-#ifdef DO_Z

-   compute_plane(p0, p1, p2, p0[2], p1[2], p2[2], zPlane);

-   span.arrayMask |= SPAN_Z;

-#endif

-   if (ctx->Light.ShadeModel == GL_SMOOTH) {

-      compute_plane(p0, p1, p2, v0->color[RCOMP], v1->color[RCOMP], v2->color[RCOMP], rPlane);

-      compute_plane(p0, p1, p2, v0->color[GCOMP], v1->color[GCOMP], v2->color[GCOMP], gPlane);

-      compute_plane(p0, p1, p2, v0->color[BCOMP], v1->color[BCOMP], v2->color[BCOMP], bPlane);

-      compute_plane(p0, p1, p2, v0->color[ACOMP], v1->color[ACOMP], v2->color[ACOMP], aPlane);

-   }

-   else {

-      constant_plane(v2->color[RCOMP], rPlane);

-      constant_plane(v2->color[GCOMP], gPlane);

-      constant_plane(v2->color[BCOMP], bPlane);

-      constant_plane(v2->color[ACOMP], aPlane);

-   }

-   span.arrayMask |= SPAN_RGBA;

-#if defined(DO_ATTRIBS)

-   {

-      const GLfloat invW0 = v0->attrib[FRAG_ATTRIB_WPOS][3];

-      const GLfloat invW1 = v1->attrib[FRAG_ATTRIB_WPOS][3];

-      const GLfloat invW2 = v2->attrib[FRAG_ATTRIB_WPOS][3];

-      compute_plane(p0, p1, p2, invW0, invW1, invW2, wPlane);

-      span.attrStepX[FRAG_ATTRIB_WPOS][3] = plane_dx(wPlane);

-      span.attrStepY[FRAG_ATTRIB_WPOS][3] = plane_dy(wPlane);

-      ATTRIB_LOOP_BEGIN

-         GLuint c;

-         if (swrast->_InterpMode[attr] == GL_FLAT) {

-            for (c = 0; c < 4; c++) {

-               constant_plane(v2->attrib[attr][c] * invW2, attrPlane[attr][c]);

-            }

-         }

-         else {

-            for (c = 0; c < 4; c++) {

-               const GLfloat a0 = v0->attrib[attr][c] * invW0;

-               const GLfloat a1 = v1->attrib[attr][c] * invW1;

-               const GLfloat a2 = v2->attrib[attr][c] * invW2;

-               compute_plane(p0, p1, p2, a0, a1, a2, attrPlane[attr][c]);

-            }

-         }

-         for (c = 0; c < 4; c++) {

-            span.attrStepX[attr][c] = plane_dx(attrPlane[attr][c]);

-            span.attrStepY[attr][c] = plane_dy(attrPlane[attr][c]);

-         }

-      ATTRIB_LOOP_END

-   }

-#endif

-

-   /* Begin bottom-to-top scan over the triangle.

-    * The long edge will either be on the left or right side of the

-    * triangle.  We always scan from the long edge toward the shorter

-    * edges, stopping when we find that coverage = 0.  If the long edge

-    * is on the left we scan left-to-right.  Else, we scan right-to-left.

-    */

-   yMin = vMin->attrib[FRAG_ATTRIB_WPOS][1];

-   yMax = vMax->attrib[FRAG_ATTRIB_WPOS][1];

-   iyMin = (GLint) yMin;

-   iyMax = (GLint) yMax + 1;

-

-   if (ltor) {

-      /* scan left to right */

-      const GLfloat *pMin = vMin->attrib[FRAG_ATTRIB_WPOS];

-      const GLfloat *pMid = vMid->attrib[FRAG_ATTRIB_WPOS];

-      const GLfloat *pMax = vMax->attrib[FRAG_ATTRIB_WPOS];

-      const GLfloat dxdy = majDx / majDy;

-      const GLfloat xAdj = dxdy < 0.0F ? -dxdy : 0.0F;

-      GLfloat x = pMin[0] - (yMin - iyMin) * dxdy;

-      GLint iy;

-      for (iy = iyMin; iy < iyMax; iy++, x += dxdy) {

-         GLint ix, startX = (GLint) (x - xAdj);

-         GLuint count;

-         GLfloat coverage = 0.0F;

-

-         /* skip over fragments with zero coverage */

-         while (startX < MAX_WIDTH) {

-            coverage = compute_coveragef(pMin, pMid, pMax, startX, iy);

-            if (coverage > 0.0F)

-               break;

-            startX++;

-         }

-

-         /* enter interior of triangle */

-         ix = startX;

-

-#if defined(DO_ATTRIBS)

-         /* compute attributes at left-most fragment */

-         span.attrStart[FRAG_ATTRIB_WPOS][3] = solve_plane(ix + 0.5F, iy + 0.5F, wPlane);

-         ATTRIB_LOOP_BEGIN

-            GLuint c;

-            for (c = 0; c < 4; c++) {

-               span.attrStart[attr][c] = solve_plane(ix + 0.5F, iy + 0.5F, attrPlane[attr][c]);

-            }

-         ATTRIB_LOOP_END

-#endif

-

-         count = 0;

-         while (coverage > 0.0F) {

-            /* (cx,cy) = center of fragment */

-            const GLfloat cx = ix + 0.5F, cy = iy + 0.5F;

-            SWspanarrays *array = span.array;

-            array->coverage[count] = coverage;

-#ifdef DO_Z

-            array->z[count] = (GLuint) solve_plane(cx, cy, zPlane);

-#endif

-            array->rgba[count][RCOMP] = solve_plane_chan(cx, cy, rPlane);

-            array->rgba[count][GCOMP] = solve_plane_chan(cx, cy, gPlane);

-            array->rgba[count][BCOMP] = solve_plane_chan(cx, cy, bPlane);

-            array->rgba[count][ACOMP] = solve_plane_chan(cx, cy, aPlane);

-            ix++;

-            count++;

-            coverage = compute_coveragef(pMin, pMid, pMax, ix, iy);

-         }

-         

-         if (ix <= startX)

-            continue;

-         

-         span.x = startX;

-         span.y = iy;

-         span.end = (GLuint) ix - (GLuint) startX;

-         _swrast_write_rgba_span(ctx, &span);

-      }

-   }

-   else {

-      /* scan right to left */

-      const GLfloat *pMin = vMin->attrib[FRAG_ATTRIB_WPOS];

-      const GLfloat *pMid = vMid->attrib[FRAG_ATTRIB_WPOS];

-      const GLfloat *pMax = vMax->attrib[FRAG_ATTRIB_WPOS];

-      const GLfloat dxdy = majDx / majDy;

-      const GLfloat xAdj = dxdy > 0 ? dxdy : 0.0F;

-      GLfloat x = pMin[0] - (yMin - iyMin) * dxdy;

-      GLint iy;

-      for (iy = iyMin; iy < iyMax; iy++, x += dxdy) {

-         GLint ix, left, startX = (GLint) (x + xAdj);

-         GLuint count, n;

-         GLfloat coverage = 0.0F;

-         

-         /* make sure we're not past the window edge */

-         if (startX >= ctx->DrawBuffer->_Xmax) {

-            startX = ctx->DrawBuffer->_Xmax - 1;

-         }

-

-         /* skip fragments with zero coverage */

-         while (startX > 0) {

-            coverage = compute_coveragef(pMin, pMax, pMid, startX, iy);

-            if (coverage > 0.0F)

-               break;

-            startX--;

-         }

-         

-         /* enter interior of triangle */

-         ix = startX;

-         count = 0;

-         while (coverage > 0.0F) {

-            /* (cx,cy) = center of fragment */

-            const GLfloat cx = ix + 0.5F, cy = iy + 0.5F;

-            SWspanarrays *array = span.array;

-            ASSERT(ix >= 0);

-            array->coverage[ix] = coverage;

-#ifdef DO_Z

-            array->z[ix] = (GLuint) solve_plane(cx, cy, zPlane);

-#endif

-            array->rgba[ix][RCOMP] = solve_plane_chan(cx, cy, rPlane);

-            array->rgba[ix][GCOMP] = solve_plane_chan(cx, cy, gPlane);

-            array->rgba[ix][BCOMP] = solve_plane_chan(cx, cy, bPlane);

-            array->rgba[ix][ACOMP] = solve_plane_chan(cx, cy, aPlane);

-            ix--;

-            count++;

-            coverage = compute_coveragef(pMin, pMax, pMid, ix, iy);

-         }

-         

-#if defined(DO_ATTRIBS)

-         /* compute attributes at left-most fragment */

-         span.attrStart[FRAG_ATTRIB_WPOS][3] = solve_plane(ix + 1.5F, iy + 0.5F, wPlane);

-         ATTRIB_LOOP_BEGIN

-            GLuint c;

-            for (c = 0; c < 4; c++) {

-               span.attrStart[attr][c] = solve_plane(ix + 1.5F, iy + 0.5F, attrPlane[attr][c]);

-            }

-         ATTRIB_LOOP_END

-#endif

-

-         if (startX <= ix)

-            continue;

-

-         n = (GLuint) startX - (GLuint) ix;

-

-         left = ix + 1;

-

-         /* shift all values to the left */

-         /* XXX this is temporary */

-         {

-            SWspanarrays *array = span.array;

-            GLint j;

-            for (j = 0; j < (GLint) n; j++) {

-               array->coverage[j] = array->coverage[j + left];

-               COPY_CHAN4(array->rgba[j], array->rgba[j + left]);

-#ifdef DO_Z

-               array->z[j] = array->z[j + left];

-#endif

-            }

-         }

-

-         span.x = left;

-         span.y = iy;

-         span.end = n;

-         _swrast_write_rgba_span(ctx, &span);

-      }

-   }

-}

-

-

-#undef DO_Z

-#undef DO_ATTRIBS

-#undef DO_OCCLUSION_TEST

+/*
+ * Mesa 3-D graphics library
+ * Version:  7.0.3
+ *
+ * Copyright (C) 1999-2007  Brian Paul   All Rights Reserved.
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a
+ * copy of this software and associated documentation files (the "Software"),
+ * to deal in the Software without restriction, including without limitation
+ * the rights to use, copy, modify, merge, publish, distribute, sublicense,
+ * and/or sell copies of the Software, and to permit persons to whom the
+ * Software is furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included
+ * in all copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS
+ * OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
+ * BRIAN PAUL BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER LIABILITY, WHETHER IN
+ * AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR IN
+ * CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE.
+ */
+
+
+/*
+ * Antialiased Triangle Rasterizer Template
+ *
+ * This file is #include'd to generate custom AA triangle rasterizers.
+ * NOTE: this code hasn't been optimized yet.  That'll come after it
+ * works correctly.
+ *
+ * The following macros may be defined to indicate what auxillary information
+ * must be copmuted across the triangle:
+ *    DO_Z         - if defined, compute Z values
+ *    DO_ATTRIBS   - if defined, compute texcoords, varying, etc.
+ */
+
+/*void triangle( struct gl_context *ctx, GLuint v0, GLuint v1, GLuint v2, GLuint pv )*/
+{
+   const SWcontext *swrast = SWRAST_CONTEXT(ctx);
+   const GLfloat *p0 = v0->attrib[FRAG_ATTRIB_WPOS];
+   const GLfloat *p1 = v1->attrib[FRAG_ATTRIB_WPOS];
+   const GLfloat *p2 = v2->attrib[FRAG_ATTRIB_WPOS];
+   const SWvertex *vMin, *vMid, *vMax;
+   GLint iyMin, iyMax;
+   GLfloat yMin, yMax;
+   GLboolean ltor;
+   GLfloat majDx, majDy;  /* major (i.e. long) edge dx and dy */
+   
+   SWspan span;
+   
+#ifdef DO_Z
+   GLfloat zPlane[4];
+#endif
+   GLfloat rPlane[4], gPlane[4], bPlane[4], aPlane[4];
+#if defined(DO_ATTRIBS)
+   GLfloat attrPlane[FRAG_ATTRIB_MAX][4][4];
+   GLfloat wPlane[4];  /* win[3] */
+#endif
+   GLfloat bf = SWRAST_CONTEXT(ctx)->_BackfaceCullSign;
+   
+   (void) swrast;
+
+   INIT_SPAN(span, GL_POLYGON);
+   span.arrayMask = SPAN_COVERAGE;
+
+   /* determine bottom to top order of vertices */
+   {
+      GLfloat y0 = v0->attrib[FRAG_ATTRIB_WPOS][1];
+      GLfloat y1 = v1->attrib[FRAG_ATTRIB_WPOS][1];
+      GLfloat y2 = v2->attrib[FRAG_ATTRIB_WPOS][1];
+      if (y0 <= y1) {
+	 if (y1 <= y2) {
+	    vMin = v0;   vMid = v1;   vMax = v2;   /* y0<=y1<=y2 */
+	 }
+	 else if (y2 <= y0) {
+	    vMin = v2;   vMid = v0;   vMax = v1;   /* y2<=y0<=y1 */
+	 }
+	 else {
+	    vMin = v0;   vMid = v2;   vMax = v1;  bf = -bf; /* y0<=y2<=y1 */
+	 }
+      }
+      else {
+	 if (y0 <= y2) {
+	    vMin = v1;   vMid = v0;   vMax = v2;  bf = -bf; /* y1<=y0<=y2 */
+	 }
+	 else if (y2 <= y1) {
+	    vMin = v2;   vMid = v1;   vMax = v0;  bf = -bf; /* y2<=y1<=y0 */
+	 }
+	 else {
+	    vMin = v1;   vMid = v2;   vMax = v0;   /* y1<=y2<=y0 */
+	 }
+      }
+   }
+
+   majDx = vMax->attrib[FRAG_ATTRIB_WPOS][0] - vMin->attrib[FRAG_ATTRIB_WPOS][0];
+   majDy = vMax->attrib[FRAG_ATTRIB_WPOS][1] - vMin->attrib[FRAG_ATTRIB_WPOS][1];
+
+   /* front/back-face determination and cullling */
+   {
+      const GLfloat botDx = vMid->attrib[FRAG_ATTRIB_WPOS][0] - vMin->attrib[FRAG_ATTRIB_WPOS][0];
+      const GLfloat botDy = vMid->attrib[FRAG_ATTRIB_WPOS][1] - vMin->attrib[FRAG_ATTRIB_WPOS][1];
+      const GLfloat area = majDx * botDy - botDx * majDy;
+      /* Do backface culling */
+      if (area * bf < 0 || area == 0 || IS_INF_OR_NAN(area))
+	 return;
+      ltor = (GLboolean) (area < 0.0F);
+
+      span.facing = area * swrast->_BackfaceSign > 0.0F;
+   }
+
+   /* Plane equation setup:
+    * We evaluate plane equations at window (x,y) coordinates in order
+    * to compute color, Z, fog, texcoords, etc.  This isn't terribly
+    * efficient but it's easy and reliable.
+    */
+#ifdef DO_Z
+   compute_plane(p0, p1, p2, p0[2], p1[2], p2[2], zPlane);
+   span.arrayMask |= SPAN_Z;
+#endif
+   if (ctx->Light.ShadeModel == GL_SMOOTH) {
+      compute_plane(p0, p1, p2, v0->color[RCOMP], v1->color[RCOMP], v2->color[RCOMP], rPlane);
+      compute_plane(p0, p1, p2, v0->color[GCOMP], v1->color[GCOMP], v2->color[GCOMP], gPlane);
+      compute_plane(p0, p1, p2, v0->color[BCOMP], v1->color[BCOMP], v2->color[BCOMP], bPlane);
+      compute_plane(p0, p1, p2, v0->color[ACOMP], v1->color[ACOMP], v2->color[ACOMP], aPlane);
+   }
+   else {
+      constant_plane(v2->color[RCOMP], rPlane);
+      constant_plane(v2->color[GCOMP], gPlane);
+      constant_plane(v2->color[BCOMP], bPlane);
+      constant_plane(v2->color[ACOMP], aPlane);
+   }
+   span.arrayMask |= SPAN_RGBA;
+#if defined(DO_ATTRIBS)
+   {
+      const GLfloat invW0 = v0->attrib[FRAG_ATTRIB_WPOS][3];
+      const GLfloat invW1 = v1->attrib[FRAG_ATTRIB_WPOS][3];
+      const GLfloat invW2 = v2->attrib[FRAG_ATTRIB_WPOS][3];
+      compute_plane(p0, p1, p2, invW0, invW1, invW2, wPlane);
+      span.attrStepX[FRAG_ATTRIB_WPOS][3] = plane_dx(wPlane);
+      span.attrStepY[FRAG_ATTRIB_WPOS][3] = plane_dy(wPlane);
+      ATTRIB_LOOP_BEGIN
+         GLuint c;
+         if (swrast->_InterpMode[attr] == GL_FLAT) {
+            for (c = 0; c < 4; c++) {
+               constant_plane(v2->attrib[attr][c] * invW2, attrPlane[attr][c]);
+            }
+         }
+         else {
+            for (c = 0; c < 4; c++) {
+               const GLfloat a0 = v0->attrib[attr][c] * invW0;
+               const GLfloat a1 = v1->attrib[attr][c] * invW1;
+               const GLfloat a2 = v2->attrib[attr][c] * invW2;
+               compute_plane(p0, p1, p2, a0, a1, a2, attrPlane[attr][c]);
+            }
+         }
+         for (c = 0; c < 4; c++) {
+            span.attrStepX[attr][c] = plane_dx(attrPlane[attr][c]);
+            span.attrStepY[attr][c] = plane_dy(attrPlane[attr][c]);
+         }
+      ATTRIB_LOOP_END
+   }
+#endif
+
+   /* Begin bottom-to-top scan over the triangle.
+    * The long edge will either be on the left or right side of the
+    * triangle.  We always scan from the long edge toward the shorter
+    * edges, stopping when we find that coverage = 0.  If the long edge
+    * is on the left we scan left-to-right.  Else, we scan right-to-left.
+    */
+   yMin = vMin->attrib[FRAG_ATTRIB_WPOS][1];
+   yMax = vMax->attrib[FRAG_ATTRIB_WPOS][1];
+   iyMin = (GLint) yMin;
+   iyMax = (GLint) yMax + 1;
+
+   if (ltor) {
+      /* scan left to right */
+      const GLfloat *pMin = vMin->attrib[FRAG_ATTRIB_WPOS];
+      const GLfloat *pMid = vMid->attrib[FRAG_ATTRIB_WPOS];
+      const GLfloat *pMax = vMax->attrib[FRAG_ATTRIB_WPOS];
+      const GLfloat dxdy = majDx / majDy;
+      const GLfloat xAdj = dxdy < 0.0F ? -dxdy : 0.0F;
+      GLint iy;
+#ifdef _OPENMP
+#pragma omp parallel for schedule(dynamic) private(iy) firstprivate(span)
+#endif
+      for (iy = iyMin; iy < iyMax; iy++) {
+         GLfloat x = pMin[0] - (yMin - iy) * dxdy;
+         GLint ix, startX = (GLint) (x - xAdj);
+         GLuint count;
+         GLfloat coverage = 0.0F;
+
+#ifdef _OPENMP
+         /* each thread needs to use a different (global) SpanArrays variable */
+         span.array = SWRAST_CONTEXT(ctx)->SpanArrays + omp_get_thread_num();
+#endif
+         /* skip over fragments with zero coverage */
+         while (startX < MAX_WIDTH) {
+            coverage = compute_coveragef(pMin, pMid, pMax, startX, iy);
+            if (coverage > 0.0F)
+               break;
+            startX++;
+         }
+
+         /* enter interior of triangle */
+         ix = startX;
+
+#if defined(DO_ATTRIBS)
+         /* compute attributes at left-most fragment */
+         span.attrStart[FRAG_ATTRIB_WPOS][3] = solve_plane(ix + 0.5F, iy + 0.5F, wPlane);
+         ATTRIB_LOOP_BEGIN
+            GLuint c;
+            for (c = 0; c < 4; c++) {
+               span.attrStart[attr][c] = solve_plane(ix + 0.5F, iy + 0.5F, attrPlane[attr][c]);
+            }
+         ATTRIB_LOOP_END
+#endif
+
+         count = 0;
+         while (coverage > 0.0F) {
+            /* (cx,cy) = center of fragment */
+            const GLfloat cx = ix + 0.5F, cy = iy + 0.5F;
+            SWspanarrays *array = span.array;
+            array->coverage[count] = coverage;
+#ifdef DO_Z
+            array->z[count] = (GLuint) solve_plane(cx, cy, zPlane);
+#endif
+            array->rgba[count][RCOMP] = solve_plane_chan(cx, cy, rPlane);
+            array->rgba[count][GCOMP] = solve_plane_chan(cx, cy, gPlane);
+            array->rgba[count][BCOMP] = solve_plane_chan(cx, cy, bPlane);
+            array->rgba[count][ACOMP] = solve_plane_chan(cx, cy, aPlane);
+            ix++;
+            count++;
+            coverage = compute_coveragef(pMin, pMid, pMax, ix, iy);
+         }
+         
+         if (ix > startX) {
+            span.x = startX;
+            span.y = iy;
+            span.end = (GLuint) ix - (GLuint) startX;
+            _swrast_write_rgba_span(ctx, &span);
+         }
+      }
+   }
+   else {
+      /* scan right to left */
+      const GLfloat *pMin = vMin->attrib[FRAG_ATTRIB_WPOS];
+      const GLfloat *pMid = vMid->attrib[FRAG_ATTRIB_WPOS];
+      const GLfloat *pMax = vMax->attrib[FRAG_ATTRIB_WPOS];
+      const GLfloat dxdy = majDx / majDy;
+      const GLfloat xAdj = dxdy > 0 ? dxdy : 0.0F;
+      GLint iy;
+#ifdef _OPENMP
+#pragma omp parallel for schedule(dynamic) private(iy) firstprivate(span)
+#endif
+      for (iy = iyMin; iy < iyMax; iy++) {
+         GLfloat x = pMin[0] - (yMin - iy) * dxdy;
+         GLint ix, left, startX = (GLint) (x + xAdj);
+         GLuint count, n;
+         GLfloat coverage = 0.0F;
+         
+#ifdef _OPENMP
+         /* each thread needs to use a different (global) SpanArrays variable */
+         span.array = SWRAST_CONTEXT(ctx)->SpanArrays + omp_get_thread_num();
+#endif
+         /* make sure we're not past the window edge */
+         if (startX >= ctx->DrawBuffer->_Xmax) {
+            startX = ctx->DrawBuffer->_Xmax - 1;
+         }
+
+         /* skip fragments with zero coverage */
+         while (startX > 0) {
+            coverage = compute_coveragef(pMin, pMax, pMid, startX, iy);
+            if (coverage > 0.0F)
+               break;
+            startX--;
+         }
+         
+         /* enter interior of triangle */
+         ix = startX;
+         count = 0;
+         while (coverage > 0.0F) {
+            /* (cx,cy) = center of fragment */
+            const GLfloat cx = ix + 0.5F, cy = iy + 0.5F;
+            SWspanarrays *array = span.array;
+            ASSERT(ix >= 0);
+            array->coverage[ix] = coverage;
+#ifdef DO_Z
+            array->z[ix] = (GLuint) solve_plane(cx, cy, zPlane);
+#endif
+            array->rgba[ix][RCOMP] = solve_plane_chan(cx, cy, rPlane);
+            array->rgba[ix][GCOMP] = solve_plane_chan(cx, cy, gPlane);
+            array->rgba[ix][BCOMP] = solve_plane_chan(cx, cy, bPlane);
+            array->rgba[ix][ACOMP] = solve_plane_chan(cx, cy, aPlane);
+            ix--;
+            count++;
+            coverage = compute_coveragef(pMin, pMax, pMid, ix, iy);
+         }
+         
+#if defined(DO_ATTRIBS)
+         /* compute attributes at left-most fragment */
+         span.attrStart[FRAG_ATTRIB_WPOS][3] = solve_plane(ix + 1.5F, iy + 0.5F, wPlane);
+         ATTRIB_LOOP_BEGIN
+            GLuint c;
+            for (c = 0; c < 4; c++) {
+               span.attrStart[attr][c] = solve_plane(ix + 1.5F, iy + 0.5F, attrPlane[attr][c]);
+            }
+         ATTRIB_LOOP_END
+#endif
+
+         if (startX > ix) {
+            n = (GLuint) startX - (GLuint) ix;
+
+            left = ix + 1;
+
+            /* shift all values to the left */
+            /* XXX this is temporary */
+            {
+               SWspanarrays *array = span.array;
+               GLint j;
+               for (j = 0; j < (GLint) n; j++) {
+                  array->coverage[j] = array->coverage[j + left];
+                  COPY_CHAN4(array->rgba[j], array->rgba[j + left]);
+#ifdef DO_Z
+                  array->z[j] = array->z[j + left];
+#endif
+               }
+            }
+
+            span.x = left;
+            span.y = iy;
+            span.end = n;
+            _swrast_write_rgba_span(ctx, &span);
+         }
+      }
+   }
+}
+
+
+#undef DO_Z
+#undef DO_ATTRIBS
+#undef DO_OCCLUSION_TEST