<div>"Good news", I believe both problems may indeed be related and of alignment. At least the first one.</div><div><br></div><div>If I change the alignment in kernel_type.h in simple ways [*] I get the Z problem fixed, but the viewport broken in situations it was working before.</div>
<div>If I go to camera fly mode the viewport is fine ('dirty' but fine) until I stop then I get the (2) problem. So what are the rules for alignment here?</div><div><br></div><div>[*] <a href="http://www.pasteall.org/27124">http://www.pasteall.org/27124</a></div>
<div><br></div><div>Regarding (1) camera behaves as if Z = 0:</div><div>-------------------------------------------------------------</div><div><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">> Here's also a small test, I'm not totally confident the differentials </span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">are aligned well in the struct, you could test this patch to see if </span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">that's somehow related, but it's only a random guess:</span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; "><br>
</span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">> </span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; "><a href="http://www.pasteall.org/27115/diff" target="_blank" style="color: rgb(103, 117, 58); ">http://www.pasteall.org/27115/diff</a></span></div>
<div>It doesn't do much. but see above.</div><div><br></div><div><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">> The way I would debug this is try to find exactly the point where things are different between the CPU and GPU, which value or operation is different between the two.</span></div>
<div><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">I was trying to do a different approach: I was trying to take out/by pass whatever I could (e.g. use identity matrixes, raster_x=raster_y=0.0f ...) that would still produce a different result between cpu/gpu. I think I need a better understanding of the rendering pipeline in cycle to look at that further. You gotta agree that this is a quite peculiar bug.</span></div>
<div><br></div><div><br></div><div>Regarding (2) viewport problem:</div><div>----------------------------------------------</div><div><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">Does F12 rendering work (with resolutions that fail in the viewport) </span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">or is it just viewport rendering? ...</span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; "><br>
</span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">> </span><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; "><a href="http://www.pasteall.org/27116/diff" target="_blank" style="color: rgb(103, 117, 58); ">http://www.pasteall.org/27116/diff</a></span></div>
<div>The dumped image is as bad as the viewport one. A note: the image is not 'broken' right away. While the light calculation is still dirty, the image is correct. Only after a few (set_tile() resolution > 8 or so) moments the viewport breaks.</div>
<div><br></div><div><span class="Apple-style-span" style="border-collapse: collapse; color: rgb(32, 32, 32); font-family: 'Droid Sans', arial, sans-serif; font-size: 13px; ">> Another thing you could test is the workgroup_size in device_opencl.cpp in path_trace and tonemap.</span></div>
workgroup_size is calculated as 16. If I change it to 1, 2, 4, 8, 16 it produces the same result.<div>If I change it to 32 OpenCL throw an error on me (OpenCL error (-54): Invalid work group size)<div><br></div><div>Thanks,</div>
<div>Dalai</div><div><br><div class="gmail_quote">2011/12/6 Dalai Felinto <span dir="ltr"><<a href="mailto:dfelinto@gmail.com">dfelinto@gmail.com</a>></span><br><blockquote class="gmail_quote" style="margin:0 0 0 .8ex;border-left:1px #ccc solid;padding-left:1ex;">
<div>Hi,</div><div class="im"><div><br></div><div>> What does this mean exactly, that Pcamera.z is always 0?</div></div><div><a href="http://www.pasteall.org/pic/22112" target="_blank">http://www.pasteall.org/pic/22112</a></div>
<div>I can rotate the camera, move it in x and y and is all good. As soon as I move it outsize Z=0 the view no longer match.</div>
<div>(rastertocamera and raster_x and raster_y seems right though, same for the transform operations)</div><div class="im"><div><br></div><div>>Does F12 rendering work (with resolutions that fail in the viewport) or is it just viewport rendering?</div>
</div><div>F12 never renders ok. it's always black to me (with CPU renders fine).</div><div><br></div><div>I'll test the other things later and get back to you.</div><div>Thanks for the fast reply.</div><div><br>
</div><font color="#888888"><div>
Dalai</div></font><div><div></div><div class="h5"><br><div class="gmail_quote">2011/12/6 Brecht Van Lommel <span dir="ltr"><<a href="mailto:brechtvanlommel@pandora.be" target="_blank">brechtvanlommel@pandora.be</a>></span><br>
<blockquote class="gmail_quote" style="margin:0 0 0 .8ex;border-left:1px #ccc solid;padding-left:1ex">
Hi,<br>
<div><br>
On Wed, Dec 7, 2011 at 12:49 AM, Dalai Felinto <<a href="mailto:dfelinto@gmail.com" target="_blank">dfelinto@gmail.com</a>> wrote:<br>
> (1) camera is always in Z 0.<br>
> <a href="http://www.pasteall.org/blend/10154" target="_blank">http://www.pasteall.org/blend/10154</a><br>
> (CPU matches blender camera, GPU matches blender camera but with Z = 0)<br>
<br>
</div>What does this mean exactly, that Pcamera.z is always 0? If that is<br>
the case, either rastertocamera is wrong, or (raster_x, raster_y) is<br>
wrong, or something goes wrong in the execution of transform(). But<br>
it's also not clear to me how ray->D can then match the cpu version,<br>
since Pcamera is used to compute it, and Z being 0 there would break<br>
that?<br>
<br>
The way I would debug this is try to find exactly the point where<br>
things are different between the CPU and GPU, which value or operation<br>
is different between the two. For example if it's inside transform(),<br>
copy the code from that into camera_sample_perspective and find out<br>
which value or operation is the culprit.<br>
<br>
Here's also a small test, I'm not totally confident the differentials<br>
are aligned well in the struct, you could test this patch to see if<br>
that's somehow related, but it's only a random guess:<br>
<a href="http://www.pasteall.org/27115/diff" target="_blank">http://www.pasteall.org/27115/diff</a><br>
<div><br>
> (2) the viewport 'crashes' for a lot of combinations of widthxheight@tiles.<br>
> Sometimes reducing the tiles fixes it, sometimes changing the width,<br>
> sometimes the height, sometimes praying ... ;)<br>
> <a href="http://www.pasteall.org/pic/show.php?id=22109" target="_blank">http://www.pasteall.org/pic/show.php?id=22109</a><br>
<br>
</div>Does F12 rendering work (with resolutions that fail in the viewport)<br>
or is it just viewport rendering? In case of the latter, the problem<br>
is either in tonemapping or viewport opengl drawing. To test if it is<br>
the opengl drawing, this line can be added to dump the image after<br>
tonemapping:<br>
<a href="http://www.pasteall.org/27116/diff" target="_blank">http://www.pasteall.org/27116/diff</a><br>
<br>
I'd also do tests with tile size set to e.g. 10000, that eliminates a<br>
variable since only image width/height matter then.<br>
<br>
Another thing you could test is the workgroup_size in<br>
device_opencl.cpp in path_trace and tonemap. You could try setting<br>
that to fixed values like 4, 8, 16. It might also be interesting to<br>
print out the workgroup_size that is computed, maybe it has a strange<br>
value.<br>
<div><br>
> (3) full shader may not be working [I'm not sure about this, but let's<br>
> assuming the "AO" render is good for now]<br>
<br>
</div>That's indeed intentional at the moment.<br>
<br>
Brecht.<br>
_______________________________________________<br>
Bf-cycles mailing list<br>
<a href="mailto:Bf-cycles@blender.org" target="_blank">Bf-cycles@blender.org</a><br>
<a href="http://lists.blender.org/mailman/listinfo/bf-cycles" target="_blank">http://lists.blender.org/mailman/listinfo/bf-cycles</a><br>
</blockquote></div><br>
</div></div></blockquote></div><br></div></div>